Al sincronizar los núcleos CUDA, lo siguiente no funciona porque el núcleo no bloquea la ejecución del programa de la CPU mientras se ejecuta:

start timer
kernel<<<g,b>>>();
end timer

He visto tres formas básicas de (con éxito) sincronizar los núcleos CUDA:

(1) Dos registros de eventos CUDA.

float responseTime; //result will be in milliseconds
cudaEvent_t start; cudaEventCreate(&start); cudaEventRecord(start); cudaEventSynchronize(start);
cudaEvent_t stop;  cudaEventCreate(&stop);
kernel<<<g,b>>>();
cudaEventRecord(stop); cudaEventSynchronize(stop);
cudaEventElapsedTime(&responseTime, start, stop); //responseTime = elapsed time

(2) Un CUDA eventRecord.

float start = read_timer(); //helper function on CPU, in milliseconds
cudaEvent_t stop;  cudaEventCreate(&stop);
kernel<<<g,b>>>();
cudaEventRecord(stop); cudaEventSynchronize(stop);
float responseTime = read_timer() - start;

(3) deviceSynchronize en lugar de eventRecord. (Probablemente solo sea útil cuando se usa la programación en una sola transmisión).

float start = read_timer(); //helper function on CPU, in milliseconds
kernel<<<g,b>>>();
cudaDeviceSynchronize();
float responseTime = read_timer() - start;

Verifiqué experimentalmente que estas tres estrategias producen el mismo resultado de tiempo.

Preguntas:

Probablemente puedas usar tu imaginación para descubrir quéread_timer() hace. Sin embargo, no hace daño proporcionar una implementación de ejemplo:

double read_timer(){
    struct timeval start;
    gettimeofday( &start, NULL ); //you need to include <sys/time.h>
    return (double)((start.tv_sec) + 1.0e-6 * (start.tv_usec))*1000; //milliseconds
}