CUDA 共享内存阵列 - 奇怪的行为

Question

在 CUDA 内核中，我有类似于以下的代码。我试图为每个线程计算一个分子，并将分子累加到块上以计算分母，然后返回比率。但是，CUDA 将 denom 的值设置为由具有最大 threadIdx.x 的块中的线程为 numer 计算的任何值，而不是跨块中所有线程计算的 numer 值的总和。有谁知道发生了什么事吗？

extern __shared__ float s_shared[];

float numer = //calculate numerator

s_shared[threadIdx.x] = numer;
s_shared[blockDim.x] += numer;
__syncthreads();

float denom = s_shared[blockDim.x];
float result = numer/denom;

“结果”应始终介于 0 和 1 之间，并且在整个块中总和应为 1，但对于每个线程（其中 threadIdx.x 为最大值）和不限于其他线程的范围的其他值，它等于 1.0在街区里。

Answer 1

您没有将求和正确同步到 blockDim.x 地点。在添加总和之前，没有线程会等待查看其他线程写入的内容。有点像

• 大家都读零，
• 回家，计算零+数字。
• Everone 将零+数字写入内存位置

我想，高 threadId 会获胜，因为它很有可能最后行动。

为了进行快速求和，您想要做的是对s_shared[threadIdx.x]

• 每个人都写下自己的号码
• 一半线程计算对的总和并将其写入新位置
• 四分之一的线程计算对的总和，并将其写入新位置
• ETC
• 直到你只有一个线程和一个总和

这需要 O(n) 工作量和 O(log n) 时间。