在CUDA线程块中可能出现负指数吗？

Question

我有一个非常简单的一维CUDA内核进行包含和，也就是说，如果我们有一个输入的一维数组

x_0，x_1，x_2，.，xn-1

输出将是

x_0，x_0+x_1，x_0+x_1+x_2，.，x0+x1+.xn-1.

下面所示的内核实际上并没有完全完成这个任务，另一方面它在每个块内完成了它的工作。无论如何，我的问题不是如何完全实现包含和，但我认为在线程计算过程中可能会出现负索引错误。

__global__ void parallel_scan_inefficient(float* input, float* output){
// num_threads and max_i are globalled defined  
__shared__ float temp[num_threads];

int i = blockIdx.x*blockDim.x+threadIdx.x;//global index

if (i<max_i)
{
    temp[threadIdx.x]=input[i];
}

for (unsigned int stride=1;stride<=threadIdx.x; stride*=2)
{
    __syncthreads();
    temp[threadIdx.x]+=temp[threadIdx.x-stride];
}

output[i]=temp[threadIdx.x];

}

这段程序实际上是来自Hwu&Kirk的教科书“大规模并行处理器编程”，第9章，第203页。

所以正如您在for-循环中看到的那样

for (unsigned int stride=1;stride<=threadIdx.x; stride*=2)
{
    __syncthreads();
    temp[threadIdx.x]+=temp[threadIdx.x-stride];
}

因为每个块的“线程Idx.x”从0开始，但是“see”从1开始。例如，我们不会看到块中第一个元素的temp-1吗？同样，在一次迭代之后，“see”变成2，我们将看到threadIdx.x=0的temp-2？

这对我来说不太合理，尽管CUDA编译器没有报告任何错误--我为这个内核运行了cuda-memcheck，它仍然很好。结果也是正确的(当然，对于每个块都是正确的，正如我所说的，这个内核只完成了部分包含的和)。

我想我可能犯了一个非常愚蠢的错误，但我就是找不到。任何光都会很感激的。非常感谢。

Answer 1

如果你有这样的代码：

for (unsigned int stride=1;stride<=threadIdx.x; stride*=2)
{
    __syncthreads();
    temp[threadIdx.x]+=temp[threadIdx.x-stride];
}

然后，对于线程Idx.x == 0的线程，for循环将被完全跳过。尝试在main中运行以下代码：

for (unsigned int stride=1;stride<=0; stride*=2)
{
    cout << "I am running" << endl;
}

你会发现控制台里什么都没有。