Pourquoi la mémoire épinglée de CUDA est-elle si rapide ?

Question

J'observe des accélérations substantielles dans le transfert de données lorsque j'utilise la mémoire épinglée pour les transferts de données CUDA. Sous Linux, l'appel système sous-jacent pour y parvenir est mlock. Dans la page de manuel de mlock, il est indiqué que le verrouillage de la page empêche sa sortie par swapping :

mlock() verrouille les pages dans la plage d'adresses commençant à addr et continuant sur len octets. Toutes les pages qui contiennent une partie de la plage d'adresses spécifiée sont garanties d'être résidentes en RAM lorsque l'appel revient avec succès ;

Lors de mes tests, j'avais quelques gigaoctets de mémoire libre sur mon système, il n'y avait donc aucun risque que les pages de mémoire puissent être échangées, mais j'ai quand même observé une accélération. Quelqu'un peut-il m'expliquer ce qu'il se passe réellement ici ? toute idée ou information est la bienvenue.

Answer 1

Pilote CUDA vérifie si la plage de mémoire est verrouillée ou non, il utilisera alors un chemin de code différent. La mémoire verrouillée est stockée dans la mémoire physique (RAM), de sorte que le dispositif peut la récupérer sans l'aide du CPU (DMA, alias copie asynchrone ; le dispositif n'a besoin que de la liste des pages physiques). La mémoire non verrouillée peut générer un défaut de page lors de l'accès, et elle n'est pas seulement stockée dans la mémoire (par exemple, elle peut être dans le swap), donc le pilote doit accéder à chaque page de la mémoire non verrouillée, la copier dans un tampon épinglé et la transmettre au DMA (copie synchrone, page par page).

Comme décrit ici http://forums.nvidia.com/index.php?showtopic=164661

La mémoire hôte utilisée par l'appel asynchrone mem copy doit être verrouillée par cudaMallocHost ou cudaHostAlloc.

Je vous recommande également de consulter les manuels cudaMemcpyAsync et cudaHostAlloc sur le site developer.download.nvidia.com. HostAlloc dit que le pilote cuda peut détecter la mémoire pinned :

Le pilote suit les plages de mémoire virtuelle allouées avec cette fonction (cudaHostAlloc) et accélère automatiquement les appels aux fonctions telles que cudaMemcpy().

Answer 2

CUDA utilise le DMA pour transférer la mémoire épinglée au GPU. La mémoire hôte paginable ne peut pas être utilisée avec la DMA car elle peut résider sur le disque. Si la mémoire n'est pas épinglée (c.-à-d. verrouillée par page), elle est d'abord copiée dans un tampon "d'attente" verrouillé par page, puis copiée vers le GPU par DMA. Ainsi, en utilisant la mémoire épinglée, vous gagnez du temps pour copier de la mémoire hôte paginable à la mémoire hôte verrouillée par page.

Answer 3

Si les pages de mémoire n'avaient pas encore été accédées, elles étaient probablement n'ont jamais été échangés pour commencer . En particulier, les pages nouvellement allouées seront des copies virtuelles de la "page zéro" universelle et n'auront pas d'instanciation physique tant qu'elles n'auront pas été écrites. De même, les nouvelles cartes de fichiers sur le disque resteront purement sur le disque jusqu'à ce qu'elles soient lues ou écrites.

Answer 4

Une note verbeuse sur la copie de pages non verrouillées vers des pages verrouillées.

Cela pourrait être extrêmement coûteux si les pages non verrouillées sont échangées par le système d'exploitation sur un système occupé avec une mémoire vive limitée. Le défaut de page sera alors déclenché pour charger les pages dans la RAM du CPU par le biais d'opérations d'entrée-sortie de disque coûteuses.

L'épinglage de pages peut aussi causer destruction de la mémoire virtuelle sur un système où la RAM du CPU est précieuse. Si le thrashing se produit, le débit du CPU peut être fortement dégradé.