c++ - CUDA推力库中的函数是否隐式同步？

Question

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

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

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

谢谢

Answer 1

您的代码至少在两个方面被破坏。

1. 它大概是一个设备指针:

   kernel_1<<<blocks, threads>>>(it);
                                 ^^
   

   不允许使用原始设备指针作为推力算法的参数:

   dt = *(thrust::max_element(it, it + 10));
                              ^^
   

   除非您将该指针包装在 thrust::device_ptr 中，或者明确使用 thrust::device 执行策略作为算法的参数。如果没有任何这些线索，推力将调度主机代码路径(这可能会出现段错误)，如 thrust quick start guide 中所讨论的那样。 .

2. 如果您使用 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 本身将插入同步。

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