相当于 CUDA 中的障碍(CLK_GLOBAL_MEM_FENCE)

Equivalent of barrier(CLK_GLOBAL_MEM_FENCE) in CUDA

在 CUDA 中调用 barrier(CLK_GLOBAL_MEM_FENCE) (OpenCL) 的等效项是什么?

它应该等到块中的所有线程都到达屏障。并且在屏障之前完成的全局内存访问应该对屏障之后的块中的所有线程可见。

__syncthreads() 就足够了,还是 __threadfence() 是全局内存栅栏所必需的,或两者兼而有之?如果是这样,两者的调用顺序应该是什么?

您可能希望阅读有关内存栅栏和执行障碍的 CUDA 文档here

__syncthreads() 既是执行屏障(对于块中的线程)也是共享和全局内存操作的内存栅栏。对于全局内存操作,栅栏强制执行仅针对块中的线程。

这里的fencing有特定的含义: 在栅栏之前发生的内存操作,对于块中的其他线程将具有在栅栏之后发生的操作之前实际发生(即可见)的可见性。请查看已链接的文档以获取对此的详细说明。

如果您还需要(全局内存)块外的防护(即对于网格中的所有线程),那么您还必须使用 __threadfence()。与 __syncthreads() 不同,__threadfence() 不是 任何类型的执行障碍。它只是一个内存栅栏。

调用顺序无关紧要(只要没有中间操作)。栅栏是一个分界点,__threadfence() 的功能是 __syncthreads()

中包含的栅栏功能的超集