多个 CUDA 内核的时间测量

Question

如何测量多个 CUDA 内核的特定时间？我试过这种方法。

   cudaEvent_t tic1, tic2, tic3, tic4;
   float time1, time2;
   cudaEventCreate(&tic1);
   cudaEventCreate(&tic2);
   cudaEventCreate(&tic3);
   cudaEventCreate(&tic4);

// launching the 1st kernel to execute on GPU
   cudaEventRecord(tic1, 0);
   kernel1<<<>>>(d_A, d_y, k, d_flag); 
   checkCuda( cudaPeekAtLastError() );  
   checkCuda(cudaDeviceSynchronize());

// calculating elapsed time on 2nd kernel               
   cudaEventRecord(tic2, 0);
   cudaEventElapsedTime(&time1, tic1, tic2);
   cudaEventSynchronize(tic2);

// launching the 2nd kernel to execute on GPU
   cudaEventRecord(tic3, 0);
   kernel2<<<>>>(d_A_sys, d_y_sys, k, d_flag); 
   checkCuda( cudaPeekAtLastError() );  
   checkCuda(cudaDeviceSynchronize());

// calculating elapsed time on 2nd kernel            
   cudaEventRecord(tic4, 0);
   cudaEventElapsedTime(&time2, tic3, tic4);

   cudaEventSynchronize(tic4);

我得到的第一个结果是正确的。但是对于第二个内核，时间显示为零。

Answer 1

要对时间和持续时间进行复杂（且准确）的测量，您不应依赖 CUDA 事件，而应使用 nVIDIA 的分析机制及其分析器应用程序 nvprof，或使用可视分析器应用程序 nvvp。

阅读 Parallel4All 博客上的这篇文章：

CUDA Pro Tip: nvprof is Your Handy Universal GPU Profiler

基本上，您启动可视化配置文件、加载可执行文件、设置命令行参数并生成可以浏览的图形时间线；它看起来像这样：

或者您在启用 nvprof 的情况下运行您的可执行文件（不使用可视化分析器），它会生成所有内核执行时间、CUDA API 调用等的 csv/xml 列表，和/或分析结果文件然后可以在可视化分析器中打开。