CUDA推力库中的函数是否隐式同步？答案

【问题标题】：Are functions in CUDA thrust library synchronized implicitly?CUDA推力库中的函数是否隐式同步？
【发布时间】：2016-04-20 15:16:02
【问题描述】：

我遇到了一些问题在使用推力库中的函数时，我不确定是否应该在它之前手动添加 cudaDeviceSynchronize。例如，

double dt = 0;
kernel_1<<<blocks, threads>>>(it);
dt = *(thrust::max_element(it, it + 10));
printf("%f\n", dt);

由于 kernel_1 是非阻塞的，主机将执行下一条语句。问题是我不确定推力::max_element 是否阻塞。如果它被阻塞，那么它工作得很好；否则，主机会跳过它并执行“printf”语句吗？

谢谢

【问题讨论】：

标签： c++ cuda thrust hpc

【解决方案1】：

您的代码至少有两种方式被破坏。

it 大概是一个设备指针：
```
kernel_1<<<blocks, threads>>>(it);
                              ^^
```
不允许使用原始设备指针作为推力算法的参数：
```
dt = *(thrust::max_element(it, it + 10));
                           ^^
```
除非您将该指针包装在 thrust::device_ptr 中，否则将 thrust::device 执行策略显式用作算法的参数。如果没有任何这些线索，thrust 将分派主机代码路径（可能会出现段错误），如thrust quick start guide 中所述。
如果您使用thrust::device_ptr 或thrust::device 修复了上述项目，则thrust::max_element will return an iterator 的类型与传递给它的迭代器一致。如果您传递thrust::device_ptr，它将返回thrust::device_ptr。如果您将thrust::device 与原始指针一起使用，它将返回一个原始指针。无论哪种情况，在主机代码中取消引用都是非法的：
```
dt = *(thrust::max_element(it, it + 10));
     ^
```
再次，我希望这样的用法会出现故障。

关于异步，可以安全地假设所有返回存储在堆栈变量中的标量的推力算法都是同步的。这意味着 CPU 线程将不会继续执行推力调用，直到堆栈变量被填充了正确的值

一般而言，关于 GPU 活动，除非您使用流，否则所有 GPU 活动都会发送到同一个（默认）流。这意味着所有 CUDA 活动都将按顺序执行，并且在前面的 CUDA 活动完成之前，给定的 CUDA 操作不会开始。因此，即使您的内核启动是异步的，并且 CPU 线程将继续进行 thrust::max_element 调用，但从该调用产生的任何 CUDA 活动都不会开始执行，直到之前的内核启动完成。因此，kernel_1 对 it 引用的数据所做的任何更改都应该在 thrust::max_element 中的任何 CUDA 处理开始之前完成并完全有效。正如我们所见，thrust::max_element 本身会插入同步。

因此，一旦您修复了代码中的缺陷，就不需要在任何地方插入额外的同步。

【讨论】：

感谢您的详尽解释。

【解决方案2】：

这个函数好像不是异步的。

这两个页面都解释了 max_element() 的行为，并且没有将其显式为异步，所以我认为它是阻塞的：

由于它使用迭代器来处理所有元素并找到最大值，我不能认为它是异步的。

您仍然可以使用 cudaDeviceSynchronize 进行实际尝试，但不要忘记在您的设备上设置 the corresponding flag。

【讨论】：