CUDA et GPU à distance

CUDA est bon pour tout le monde, tant qu'il y a une carte vidéo de Nvidia à portée de main. Mais que faire lorsqu'il n'y a pas de carte graphique Nvidia sur votre ordinateur portable préféré? Ou avez-vous besoin de réaliser le développement dans une machine virtuelle?

J'essaierai d'envisager dans cet article une solution telle que le framework rCUDA (Remote CUDA), qui aidera lorsqu'il y a une carte graphique Nvidia, mais elle n'est pas installée sur la machine sur laquelle les applications CUDA sont censées être lancées. Pour ceux qui sont intéressés, bienvenue au chat.

Qu'est-ce que rCUDA

rCUDA (Remote CUDA) est un framework qui implémente l'API CUDA, vous permettant d'utiliser une carte vidéo située sur la machine distante pour l'informatique CUDA sans apporter de modifications à votre code. Développé à l'Université Polytechnique de Valence ( équipe rcuda ).

Limitations

Seuls les systèmes GNU / Linux sont actuellement pris en charge, cependant, les développeurs promettent la prise en charge de Windows à l'avenir. La version actuelle de rCUDA, 18.03beta, est compatible avec CUDA 5-8, c'est-à-dire que CUDA 9 n'est pas pris en charge. Les développeurs ont déclaré une compatibilité totale avec l'API CUDA, à l'exception des graphiques.

Cas d'utilisation possibles

1. L'exécution d'applications CUDA dans une machine virtuelle lors du transfert d'une carte vidéo est gênante ou impossible, par exemple, lorsque la carte vidéo est occupée par un hôte ou lorsqu'il existe plusieurs machines virtuelles.
2. Ordinateur portable sans carte graphique discrète.
3. Le désir d'utiliser plusieurs cartes vidéo (clustering). Théoriquement, vous pouvez utiliser toutes les cartes vidéo disponibles dans l'équipe, y compris conjointement.

Brève instruction

Configuration de test

Des tests ont été effectués sur la configuration suivante:

Serveur:
Ubuntu 16.04, GeForce GTX 660

Client:
Une machine virtuelle avec Ubuntu 16.04 sur un ordinateur portable sans carte graphique discrète.

Obtenir rCUDA

L'étape la plus difficile. Malheureusement, pour le moment, le seul moyen d'obtenir votre copie de ce cadre est de remplir le formulaire de demande approprié sur le site officiel. Cependant, les développeurs promettent de répondre dans un délai de 1 à 2 jours. Dans mon cas, ils m'ont envoyé une distribution le même jour.

Installer CUDA

Vous devez d'abord installer CUDA Toolkit sur le serveur et le client (même si le client n'a pas de carte vidéo nvidia). Pour ce faire, vous pouvez le télécharger sur le site officiel ou utiliser le référentiel. L'essentiel est d'utiliser une version pas supérieure à 8. Dans cet exemple, le programme d'installation .run du site officiel est utilisé .

chmod +x cuda_8.0.61_375.26_linux.run ./cuda_8.0.61_375.26_linux.run 

Important! Sur le client, vous devez refuser d'installer le pilote nvidia. Par défaut, la boîte à outils CUDA sera disponible dans / usr / local / cuda /. Installez les échantillons CUDA, vous en aurez besoin.

Installer rCUDA

Nous décompresserons l'archive reçue des développeurs dans notre répertoire personnel sur le serveur et sur le client.

 tar -xvf rCUDA*.tgz -C ~/ mv ~/rCUDA* ~/rCUDA 

Vous devez effectuer ces actions à la fois sur le serveur et sur le client.

Démarrage du démon rCUDA sur le serveur

 export PATH=$PATH/usr/local/cuda/bin export LD_LIBRARY_PATH=$LD_LIBRARY_PATH:/usr/local/cuda/lib64:/home/<XXX>/rCUDA/lib/cudnn cd ~/rCUDA/bin ./rCUDAd 

Remplacez <XXX> par votre nom d'utilisateur. Utilisez ./rCUDAd -iv si vous voulez voir la sortie détaillée.

Configuration du client

Ouvrons le terminal sur le client, dans lequel nous exécuterons le code CUDA à l'avenir. Côté client, nous devons "remplacer" les bibliothèques CUDA standard par des bibliothèques rCUDA, pour lesquelles nous ajoutons les chemins d'accès appropriés à la variable d'environnement LD_LIBRARY_PATH. Nous devons également spécifier le nombre de serveurs et leurs adresses (dans mon exemple, ce sera un).

 export PATH=$PATH/usr/local/cuda/bin export LD_LIBRARY_PATH=/home/<XXX>/rCUDA/lib/:$LD_LIBRARY_PATH export RCUDA_DEVICE_COUNT=1 #    (),     export RCUDA_DEVICE_0=<IP  >:0 #     

Assemblage et lancement

Essayons de construire et d'exécuter quelques exemples.

Exemple 1

Commençons par un exemple simple de deviceQuery qui affiche simplement les paramètres CUDA pour un appareil compatible, c'est-à-dire, dans notre cas, le GTX660 distant.

 cd <YYY>/NVIDIA_CUDA-8.0_Samples/1_Utilities/deviceQuery make EXTRA_NVCCFLAGS=--cudart=shared 

Important! Sans EXTRA_NVCCFLAGS = - cudart = shared, le miracle ne fonctionnera pas
Remplacez <YYY> par le chemin que vous avez spécifié pour les échantillons CUDA lors de l'installation de CUDA.

Exécutez l'exemple assemblé:

 ./deviceQuery 

Si vous avez tout fait correctement, le résultat sera quelque chose comme ceci:

 ./deviceQuery Starting... CUDA Device Query (Runtime API) version (CUDART static linking) Detected 1 CUDA Capable device(s) Device 0: "GeForce GTX 660" CUDA Driver Version / Runtime Version 9.0 / 8.0 CUDA Capability Major/Minor version number: 3.0 Total amount of global memory: 1994 MBytes (2090991616 bytes) ( 5) Multiprocessors, (192) CUDA Cores/MP: 960 CUDA Cores GPU Max Clock rate: 1072 MHz (1.07 GHz) Memory Clock rate: 3004 Mhz Memory Bus Width: 192-bit L2 Cache Size: 393216 bytes Maximum Texture Dimension Size (x,y,z) 1D=(65536), 2D=(65536, 65536), 3D=(4096, 4096, 4096) Maximum Layered 1D Texture Size, (num) layers 1D=(16384), 2048 layers Maximum Layered 2D Texture Size, (num) layers 2D=(16384, 16384), 2048 layers Total amount of constant memory: 65536 bytes Total amount of shared memory per block: 49152 bytes Total number of registers available per block: 65536 Warp size: 32 Maximum number of threads per multiprocessor: 2048 Maximum number of threads per block: 1024 Max dimension size of a thread block (x,y,z): (1024, 1024, 64) Max dimension size of a grid size (x,y,z): (2147483647, 65535, 65535) Maximum memory pitch: 2147483647 bytes Texture alignment: 512 bytes Concurrent copy and kernel execution: Yes with 1 copy engine(s) Run time limit on kernels: Yes Integrated GPU sharing Host Memory: No Support host page-locked memory mapping: Yes Alignment requirement for Surfaces: Yes Device has ECC support: Disabled Device supports Unified Addressing (UVA): Yes Device PCI Domain ID / Bus ID / location ID: 0 / 1 / 0 Compute Mode: < Default (multiple host threads can use ::cudaSetDevice() with device simultaneously) > deviceQuery, CUDA Driver = CUDART, CUDA Driver Version = 9.0, CUDA Runtime Version = 8.0, NumDevs = 1, Device0 = GeForce GTX 660 Result = PASS 

La chose la plus importante que nous devrions voir:

Device0 = GeForce GTX 660
Résultat = PASS

Super! Nous avons réussi à construire et exécuter l'application CUDA sur une machine sans carte graphique discrète, en utilisant à cet effet une carte vidéo installée sur un serveur distant.

Important! Si la sortie de l'application commence par des lignes du formulaire:

 mlock error: Cannot allocate memory rCUDA warning: 1007.461 mlock error: Cannot allocate memory 

cela signifie qu'il est nécessaire d'ajouter les lignes suivantes au fichier "/etc/security/limits.conf" sur le serveur et sur le client:

 * hard memlock unlimited * soft memlock unlimited 

Ainsi, vous autoriserez tous les utilisateurs (*) mémoire de blocage illimitée (illimitée) (memlock). Il serait encore mieux de remplacer * par l'utilisateur souhaité, et au lieu d'un nombre illimité, choisissez des droits moins lourds.

Exemple 2

Essayons maintenant quelque chose de plus intéressant. Nous testerons l'implémentation du produit scalaire des vecteurs en utilisant la mémoire partagée et la synchronisation ("Technologie CUDA dans les exemples" Sanders J. Kendrot E. 5.3.1).

Dans cet exemple, nous calculons le produit scalaire de deux vecteurs de dimension 33 * 1024, en comparant la réponse avec le résultat obtenu sur le CPU.

 #include <stdio.h> #define imin(a,b) (a<b?a:b) const int N = 33 * 1024; const int threadsPerBlock = 256; const int blocksPerGrid = imin(32, (N+threadsPerBlock-1) / threadsPerBlock); __global__ void dot(float* a, float* b, float* c) { __shared__ float cache[threadsPerBlock]; int tid = threadIdx.x + blockIdx.x * blockDim.x; int cacheIndex = threadIdx.x; float temp = 0; while (tid < N){ temp += a[tid] * b[tid]; tid += blockDim.x * gridDim.x; } // set the cache values cache[cacheIndex] = temp; // synchronize threads in this block __syncthreads(); // for reductions, threadsPerBlock must be a power of 2 // because of the following code int i = blockDim.x/2; while (i != 0){ if (cacheIndex < i) cache[cacheIndex] += cache[cacheIndex + i]; __syncthreads(); i /= 2; } if (cacheIndex == 0) c[blockIdx.x] = cache[0]; } int main (void) { float *a, *b, c, *partial_c; float *dev_a, *dev_b, *dev_partial_c; // allocate memory on the cpu side a = (float*)malloc(N*sizeof(float)); b = (float*)malloc(N*sizeof(float)); partial_c = (float*)malloc(blocksPerGrid*sizeof(float)); // allocate the memory on the gpu cudaMalloc((void**)&dev_a, N*sizeof(float)); cudaMalloc((void**)&dev_b, N*sizeof(float)); cudaMalloc((void**)&dev_partial_c, blocksPerGrid*sizeof(float)); // fill in the host memory with data for(int i=0; i<N; i++) { a[i] = i; b[i] = i*2; } // copy the arrays 'a' and 'b' to the gpu cudaMemcpy(dev_a, a, N*sizeof(float), cudaMemcpyHostToDevice); cudaMemcpy(dev_b, b, N*sizeof(float), cudaMemcpyHostToDevice); dot<<<blocksPerGrid, threadsPerBlock>>>(dev_a, dev_b, dev_partial_c); // copy the array 'c' back from the gpu to the cpu cudaMemcpy(partial_c,dev_partial_c, blocksPerGrid*sizeof(float), cudaMemcpyDeviceToHost); // finish up on the cpu side c = 0; for(int i=0; i<blocksPerGrid; i++) { c += partial_c[i]; } #define sum_squares(x) (x*(x+1)*(2*x+1)/6) printf("GPU - %.6g \nCPU - %.6g\n", c, 2*sum_squares((float)(N-1))); // free memory on the gpu side cudaFree(dev_a); cudaFree(dev_b); cudaFree(dev_partial_c); // free memory on the cpu side free(a); free(b); free(partial_c); } 

Construire et exécuter:

 /usr/local/cuda/bin/nvcc --cudart=shared dotProd.cu -o dotProd ./dotProd 

Ce résultat nous dit que tout va bien pour nous:

GPU - 2.57236e + 13
CPU - 2.57236e + 13

Exemple 3

Exécutez un autre test CUDA-matrixMulCUBLAS standard (multiplication matricielle).

 cd < YYY>/NVIDIA_CUDA-8.0_Samples/0_Simple/matrixMulCUBLAS make EXTRA_NVCCFLAGS=--cudart=shared ./matrixMulCUBLAS 

Intéressant pour nous:

Performance = 436,24 GFlop / s,
Comparaison de CUBLAS Matrix Multiply avec les résultats du processeur: PASS

La sécurité

Je n'ai trouvé aucune mention d'une méthode d'autorisation dans la documentation de rCUDA. Je pense qu'en ce moment, la chose la plus simple qui puisse être faite est d'ouvrir l'accès au port souhaité (8308) uniquement à partir d'une adresse spécifique.

En utilisant iptables, cela ressemblera à ceci:

 iptables -A INPUT -m state --state NEW -p tcp -s < > --dport 8308 -j ACCEPT 

Pour le reste, je laisse le problème de sécurité au-delà de la portée de cet article.