CUDA cudaMemcpy Structure de tableaux

J’aimerais nettoyer les parameters des kernelx CUDA dans mon projet.


Maintenant, un kernel a besoin de 3 tableaux uint32_t , ce qui conduit à un code assez laid: (id signifie le thread global id et valX est une valeur arbitraire)

 __global__ void some_kernel(uint32_t * arr1, uint32_t * arr2, uint32_t * arr3){arr1[id] = val1; arr2[id] = val2; arr3[id] = val3;} 

Je voudrais sorround tous ces tableaux avec une structure:

 typedef struct S{uint_32_t arr1, uint_32_t arr2, uint_32_t arr3, uint32_t size} S; 

où taille désigne la longueur de chaque arrX à l’intérieur de la structure.

Ce que j’aimerais avoir, c’est quelque chose comme:

 __global__ void some_kernel(S * s){s->arr1[id] = val1; s->arr2[id] = val2; s->arr3[id] = val3;} 

À quoi ressembleraient un cudaMalloc et un cudaMemcpy pour une structure comme celle-ci? Y a-t-il des inconvénients de performances que je ne vois pas encore?

Merci d’avance!

Vous avez au moins deux options. Les talonniers ont déjà fait un excellent choix, mais je vais vous présenter l’approche «apprendre à la dure».

Tout d’abord, votre définition de struct:

 typedef struct S { uint32_t *arr1; uint32_t *arr2; uint32_t *arr3; uint32_t size; } S; 

… et la définition du kernel (avec une variable globale, mais vous n’avez pas besoin de suivre avec ce modèle):

 const int size = 10000; __global__ void some_kernel(S *s) { int id = blockIdx.x * blockDim.x + threadIdx.x; if (id < size) { s->arr1[id] = 1; // val1 s->arr2[id] = 2; // val2 s->arr3[id] = 3; // val3 } } 

Remarquez que if vous empêche de courir en dehors des limites.

Ensuite, nous arrivons avec une fonction qui prépare les données, exécute le kernel et affiche un résultat. La première partie concerne l’affectation des données:

 uint32_t *host_arr1, *host_arr2, *host_arr3; uint32_t *dev_arr1, *dev_arr2, *dev_arr3; // Allocate and fill host data host_arr1 = new uint32_t[size](); host_arr2 = new uint32_t[size](); host_arr3 = new uint32_t[size](); // Allocate device data cudaMalloc((void **) &dev_arr1, size * sizeof(*dev_arr1)); cudaMalloc((void **) &dev_arr2, size * sizeof(*dev_arr2)); cudaMalloc((void **) &dev_arr3, size * sizeof(*dev_arr3)); // Allocate helper struct on the device S *dev_s; cudaMalloc((void **) &dev_s, sizeof(*dev_s)); 

Ce n’est rien de spécial, vous venez d’allouer trois tableaux et struct. Ce qui semble plus intéressant, c’est comment gérer la copie de telles données dans un périphérique:

 // Copy data from host to device cudaMemcpy(dev_arr1, host_arr1, size * sizeof(*dev_arr1), cudaMemcpyHostToDevice); cudaMemcpy(dev_arr2, host_arr2, size * sizeof(*dev_arr2), cudaMemcpyHostToDevice); cudaMemcpy(dev_arr3, host_arr3, size * sizeof(*dev_arr3), cudaMemcpyHostToDevice); // NOTE: Binding pointers with dev_s cudaMemcpy(&(dev_s->arr1), &dev_arr1, sizeof(dev_s->arr1), cudaMemcpyHostToDevice); cudaMemcpy(&(dev_s->arr2), &dev_arr2, sizeof(dev_s->arr2), cudaMemcpyHostToDevice); cudaMemcpy(&(dev_s->arr3), &dev_arr3, sizeof(dev_s->arr3), cudaMemcpyHostToDevice); 

En plus de la copie ordinaire du tableau que vous avez remarqué, il est également nécessaire de les “lier” avec la structure. Pour cela, vous devez passer une adresse de pointeur. En conséquence, seuls ces pointeurs sont copiés.

Prochain appel du kernel, recopiez les données sur l’hôte et imprimez les résultats:

 // Call kernel some_kernel<<<10000/256 + 1, 256>>>(dev_s); // block size need to be a multiply of 256 // Copy result to host: cudaMemcpy(host_arr1, dev_arr1, size * sizeof(*host_arr1), cudaMemcpyDeviceToHost); cudaMemcpy(host_arr2, dev_arr2, size * sizeof(*host_arr2), cudaMemcpyDeviceToHost); cudaMemcpy(host_arr3, dev_arr3, size * sizeof(*host_arr3), cudaMemcpyDeviceToHost); // Print some result std::cout << host_arr1[size-1] << std::endl; std::cout << host_arr2[size-1] << std::endl; std::cout << host_arr3[size-1] << std::endl; 

N'oubliez pas que dans tout code sérieux, vous devez toujours rechercher les erreurs des appels CUDA API.