时间戳

首页 > 解决方案 > 混合隐式和显式 CUDA 流

c++ - 混合隐式和显式 CUDA 流

问题描述

我有一些 CUDA 8.0 代码（编辑：我继承的，不是我写的），基本上看起来像这样：

cudaMemcpy(devInputData, ..., cudaMemcpyHostToDevice);
kernelThings<<<GRIDS, BLOCKS, 0, myStream>>>(devInputData);
cudaDeviceSynchronize();

cudaMemcpy()在没有流的情况下同步工作，所以据我所知，这段代码是正确的。

如果我使用 编译CUDA_API_PER_THREAD_DEFAULT_STREAM，这段代码是否仍然安全？我认为不，cudaMemcpy()现在是异步发生的，因此内核有可能在cudaMemcpy()完成之前启动。但是，查看Nsight分析器，我发现没有重叠 - 从文字上我看到：

[Memcpy HtoD]
                  [kernelThings]

两个函数之间有 16 微秒的间隔。此行为在应用程序中重复多次。

但是，我接下来删除cudaDeviceSynchronize()、重新运行Nsight，然后看到它们现在重叠了：

[Memcpy HtoD]
         [kernelThings]

内核现在在cudaMemcpy完成前 10 微秒启动。

显然，正确的解决方法是使用流cudaMemcpyAsync()：

cudaMemcpyAsync(devInputData, ..., cudaMemcpyHostToDevice, myStream);

但是，我的问题是为什么我在使用时没有看到进程重叠cudaDeviceSynchronize()？简单的答案是我可能对不同版本的 CUDA 或 GPU 不那么幸运吗？

标签： c++cuda

解决方案

不要依赖隐式/默认流和操作及其同步行为的拐杖。对于您的第一个“Hello world”级程序 - 这可能很方便，但正如您自己所注意到的，您必须成为 API 律师或通灵者才能猜测在某些复杂场景中究竟会发生什么。

只需确保您在内核中使用的每个缓冲区都是：

  1. 由同一命令队列上的先前操作填充或


  2. 2.1 如果它是一个输入缓冲区 - 有一个事件对确保在内核启动另一个流之前发生任何影响缓冲区的事情，并且

    2.2 如果它是一个输出缓冲区 - 有一个事件对确保内核在输出缓冲区用于其他地方之前执行

推荐阅读