使用cudaEvent计时错误

使用cudaEvent计时错误

最近刚入坑cuda学习，遇到很多小问题却耽误很多时间。比如用cudaEvent来计算gpu时间时总是算不出来，今天偶然看了一眼代码终于发现了问题，原来是计算时间之前缺少了cudaEventSynchronize()这个代码。基本形式如下：

cudaError_t cudaStatus;
cudaEvent_t start, stop;
checkCudaErrors(cudaEventCreate(&start));
checkCudaErrors(cudaEventCreate(&stop));
checkCudaErrors(cudaEventRecord(start, 0));
///
///gpu指令
///
checkCudaErrors(cudaEventRecord(stop, 0)）;
checkCudaErrors(cudaEventSynchronize(stop)）;
 float time;
checkCudaErrors(cudaEventElapsedTime(&time, start, stop));
 std::cout << "time spend: " << time << std::endl;

checkCudaErrors宏定义自行百度
一般的解释是说：通过这个函数来阻塞CPU线程，等待GPU执行完成。大概意思就是这样，然后我又查了一些资料，发现对于GPU的同步过程也是挺有讲究的。

cudaEventSynchronize()
cudaStreamSynchronize()
cudaDeviceSynchronize()
cudaThreadSynchronize()

当捕获一个cudaEvent在一个cuda流时，每个事件都是串行执行的，并默认0为空流，即cudaEventRecord(start, 0)如果当前gpu比较忙，则需要cudaEventQuery()或cudaStreamWaitEvent()等待捕获完成，然后再通过cudaEventSynchronize()或cudaStreamSynchronize()阻塞当前流。它的分辨率可以达到半微秒。
使用cudaDeviceSynchronize()一般用于多个流之间的通信，所以用它来代替事件同步是不可以的，而它带来的问题就是它会使gpu管道暂停，从而阻塞所有流的执行。即使不使用事件或采用cpu计时时，在gpu启用后也需要它来确保计算结果的正确。和cudaThreadSynchronize()类似，不推荐使用。
现在我们通过cudaEvent除了查看gpu的运行时间，还可以去计算它的带宽，吞吐量等。见 https://devblogs.nvidia.com/how-implement-performance-metrics-cuda-cc/#disqus_thread