CUDA：来自不同经线但同一个块的2个线程尝试写入相同的SHARED内存位置：危险？

这会导致共享内存不一致吗？CUDA：来自不同经线但同一个块的2个线程尝试写入相同的SHARED内存位置：危险？

我的内核的代码看起来是这样的（伪代码）：

__shared__ uint histogram[32][64]; 

uint threadLane = threadIdx.x % 32; 

for (data){ 
    histogram[threadLane][data]++; 
}

这会不会导致碰撞，因为，在具有64个线程的块，ID为 “x” 和“线程（X + 32） “会经常写入矩阵中的相同位置？

该程序计算给定矩阵的直方图。我有一个相同的CPU程序。由GPU计算的直方图一直比CPU计算的直方图低1/128，我无法弄清楚原因。

来源

2017-07-12 ismarlowe

您能否提供一些更详细的信息，特别是关于'data'与threadIdx以及启动配置有关的内容？编译的东西会更好。 –

这很危险。它导致竞争条件。

如果您无法保证块中的每个线程都具有对共享内存中某个位置的唯一写访问权限，那么您遇到了问题，因为您需要通过同步来解决问题。

看看本文使用SM的直方图计算的正确和有效的方式：http://developer.download.nvidia.com/compute/cuda/1.1-Beta/x86_website/projects/histogram64/doc/histogram.pdf

注意，是很多图书馆的网络，使您可以在计算一行，Thrust for instance直方图。

来源

2017-07-12 17:19:52

谢谢你immensly！这开始让我发疯。我将不得不以不同的方式重做整个事情。 – ismarlowe

看看'cuda-memcheck --tool racecheck'：http://docs.nvidia.com/cuda/cuda-memcheck/index.html#using-racecheck和https://stackoverflow.com/questions/13861017/cuda-racecheck-shared-memory-array-and-cudadevicesynchronize –

有没有办法强制一个warp在另一个接口之前完成某个任务？或者，是否可以指示机器以原子方式访问共享内存？ – ismarlowe

CUDA：来自不同经线但同一个块的2个线程尝试写入相同的SHARED内存位置：危险？

回答

相关问题