Comment mesurer correctement les temps CUDA?

J’essaie de mesurer correctement les temps d’exécution parallèles et séquentiels, mais je doute en raison de:

Supposons que nous ayons le code suivant:

//get the time clock_t start,finish; double totaltime; start = clock(); double *d_A, *d_B, *d_X; cudaMalloc((void**)&d_A, sizeof(double) * Width * Width); cudaMalloc((void**)&d_B, sizeof(double) * Width); cudaMalloc((void**)&d_X, sizeof(double) * Width); cudaMemcpy(d_A, A, sizeof(double) * Width * Width, cudaMemcpyHostToDevice); cudaMemcpy(d_B, B, sizeof(double) * Width, cudaMemcpyHostToDevice); do_parallel_matmul<<>>(d_A, d_B, d_X, Width); cudaMemcpy(X, d_X, sizeof(double) * Width, cudaMemcpyDeviceToHost); finish = clock(); totaltime=(double)(finish-start)/CLOCKS_PER_SEC; printf("%f", totaltime); 

Ce temps est beaucoup plus long que le temps séquentiel mesuré comme suit:

 clock_t start,finish; double totaltime; start = clock(); do_seq_matmult(); finish = clock(); totaltime=(double)(finish-start)/CLOCKS_PER_SEC; printf("%f", totaltime); 

Donc, je ne sais pas si je devrais mesurer le temps du kernel CUDA uniquement comme suit:

 clock_t start,finish; double totaltime; start = clock(); do_parallel_matmul(); finish = clock(); totaltime=(double)(finish-start)/CLOCKS_PER_SEC; printf("%f", totaltime); 

et éviter les copies de mémoire entre l’hôte et le périphérique …

Je pose la question ci-dessus car je dois soumettre une comparaison entre les exécutions parallèles et les exécutions séquentielles … Mais si je mesure les copies en mémoire dans CUDA, il n’y a pas de grande différence entre CUDA et C …

MODIFIER:

 void do_seq_matmult(const double *A, const double *X, double *resul, const int tam) { *resul = 0; for(int i = 0; i < tam; i++) { for(int j = 0; j < tam; j++) { if(i != j) *resul += A[i * tam + j] * X[j]; } } } __global__ void do_parallel_matmul( double * mat_A, double * vec, double * rst, int dim) { int rowIdx = threadIdx.x + blockIdx.x * blockDim.x; // Get the row Index int aIdx; while( rowIdx < dim) { rst[rowIdx] = 0; // clean the value at first for (int i = 0; i < dim; i++) { aIdx = rowIdx * dim + i; // Get the index for the element a_{rowIdx, i} rst[rowIdx] += (mat_A[aIdx] * vec[i] ); // do the multiplication } rowIdx += gridDim.x * blockDim.x; } __syncthreads(); } 

    Quelques idées:

    1. Il n’est pas juste de programmer l’allocation de mémoire de périphérique et de la comparer à un processeur sans l’allocation de mémoire par l’hôte.

    2. Si cudaMalloc((void**)&d_A, sizeof(double) * Width * Width); CUDA est le premier appel qui inclura la création de contexte CUDA, ce qui pourrait représenter une surcharge importante.

    3. Cudamemcpy n’est pas une bonne comparaison CPU / GPU, car elle dépendra de la bande passante PCI-e du système. Par contre, si vous voyez le kernel comme une accélération du sharepoint vue du CPU, vous devrez inclure le memcpy. Pour optimiser la bande passante PCI-e, utilisez une mémoire à locking de page.

    4. Si votre application doit exécuter la multiplication plusieurs fois, vous avez la possibilité de masquer la plus grande partie de la mémoire en superposant la copie avec l’exécution du kernel. C’est encore mieux sur une unité Tesla où vous avez deux moteurs DMA.

    5. Pour synchroniser le kernel lui-même, vous devrez synchroniser le processeur avec le processeur graphique avant d’arrêter le chronomètre, sinon vous ne ferez que chronométrer le lancement du kernel et non son exécution. L’appel d’un kernel à partir de la CPU est asynchrone. SI vous voulez chronométrer l’exécution du kernel sur le GPU, utilisez cudaEvents.

    6. Exécutez plusieurs threads sur le GPU pour obtenir une comparaison juste.

    7. Améliorer le kernel, vous pouvez faire mieux.

    Vous utilisez la mauvaise fonction pour vos mesures. clock mesure le temps que votre processus a passé sur votre processeur et non le temps passé sur wallclock.

    Jetez un coup d’œil à la bibliothèque High Precision Timer , elle utilise des fonctions de synchronisation liées au système d’exploitation pour mesurer le temps.

    Il utilise un ensemble de fonctions pouvant vous donner une précision micro-seconde .

    Si vous utilisez Windows, vous devez utiliser QueryPerformanceFrequency et QueryPerformanceCounter sous Linux: gettimeofday()

    C’est très léger et facile à utiliser. Disponible pour Windows et Linux.