4

我正在编写一个必须执行块间同步(N 维和其他内存传输操作的总和)的代码。当我增加问题的维度时,结果是错误的。

我正在与__threadfence()第一个维度(N<192)进行同步,这没关系,但是如果我__threadfence()在代码段中插入其他维度,则结果对于更多维度是正确的。

一个threadfence()还不够同步?此外,数据结果在同一块中使用。

在编程指南中,信息表明threadfence等待所有内存空间都准备好(共享和全局)

4

2 回答 2

11

没有很好的方法来执行块之间的同步。您可以采用一种 hacky 方法来等待自旋并占用您的 GPU 内存带宽,或者您可以终止您的内核并启动一个新内核。

__threadfence()不适用于块之间的同步。__threadfence()用于暂停当前线程,直到所有先前对共享内存和全局内存的写入都被其他线程可见。它不会停止也不会影响其他线程的位置!

您可以检查以下问题:

于 2012-07-20T16:12:34.183 回答
1

合作组将允许同一内核中不同块之间的同步。它现在也很容易使用。

#include <cooperative_groups.h>
namespace cg = cooperative_groups;

// and then in your code
cg::grid_group grid = cg::this_grid();
grid.sync(); // All threads in all blocks must run this line

整个内核中每个块中的每个线程都必须运行它grid.sync()。只有在所有线程都运行了下一行之后,才会继续执行下一行。

https://docs.nvidia.com/cuda/cuda-c-programming-guide/index.html#cooperative-groups

于 2021-09-14T06:04:57.743 回答