计时 CUDA 流

Question

我在理解这段代码时遇到了一些问题（运行在非 Hyper-Q 兼容的 GPU 上）：

CHECK(cudaEventRecord(start, 0));

// dispatch job with depth first ordering
for (int i = 0; i < n_streams; i++)
{
    kernel_1<<<grid, block, 0, streams[i]>>>();
    kernel_2<<<grid, block, 0, streams[i]>>>();
    kernel_3<<<grid, block, 0, streams[i]>>>();
    kernel_4<<<grid, block, 0, streams[i]>>>();
}

// record stop event
CHECK(cudaEventRecord(stop, 0));
CHECK(cudaEventSynchronize(stop));

// calculate elapsed time
CHECK(cudaEventElapsedTime(&elapsed_time, start, stop));
printf("Measured time for parallel execution = %.3fs\n",
       elapsed_time / 1000.0f);

整个程序也可以找到here。

问题是：在计算经过时间之前，我们如何确定所有其他内核都已完成？

非空流与启动的其他流之间似乎没有任何同步。即使停止事件是队列中的最后一个事件，根据我的理解，其他流中的其他内核似乎仍然是运行.

Answer 1

我在 Professional CUDA C Programming 中找到了答案：

Events specified on the default stream apply to all preceding operations in all CUDA streams.

换句话说，默认的非 NULL 流中的事件在所有流中的所有先前操作完成后完成。

计时 CUDA 流

Timing CUDA Streams

cuda