Cuda Streams的概述（四）-- 同步

同步

同步的APIs

同步所有的事情

//阻塞host端，直到所有的CUDA调用完成。
cudaDeviceSynchronize();

同步主机端特定的流

//阻塞host端，直到流里的CUDA调用完成。
cudaStreamSynchronize(stream); 

通过主机端或设备用events

CUDA Events

当操作在流中发生时，提供一个信号机制。这个对profiling和同步是很有用的。
Event有一个boolean状态：

• 发生
• 未发生
• 注意：默认状态=发生

管理Event

//创建一个event
cudaEventCreate(&event);
//销毁一个event
cudaEventDestroy(&event);
//禁用定时来提高性能并避免同步问题
cudaEventCreateWithFlags(&ev, cudaEventDisableTiming);
//把event状态设为未发生
//将event加入队列
//当event到达流的前面时，event状态设为发生
cudaEventRecord(&event, stream);

使用event进行同步

//如果event发生了，返回CUDA_SUCCESS；
cudaEventQuery(event);
//阻塞host端，直到流完成所有的未完成的调用
cudaEventSynchronize(event);
//阻塞流直到event发生
//仅在此次调用之后阻塞启动
//不会阻塞主机端
cudaStreamWaitEvent(stream, event);

一个常见的多线程错误
在cudaEventRecord之前调用cudaEventSynchronize

CUDA_LAUNCH_BLOCKING

CUDA_LAUNCH_BLOCKING是强制同步的环境变量。

• export CUDA_LAUNCH_BLOCKING=1
• 使在主机端，所有CUDA操作是同步的。

也可以用于调试竞争条件。

• 如果在设置了CUDA_LAUNCH_BLOCKING的情况下成功运行，但没有竞争条件的情况下就无法运行。

回顾

在主机端实现同步可以通过：

cudaDeviceSynchronize();
cudaStreamSynchronize(stream);
cudaEventSynchronize(event)

在流之间的同步可以通过：

cudaStreamWaitEvent(stream,event)

通过CUDA_LAUNCH_BLOCKING来确定竞争条件。