线程发散较少的内核

Question

结果的预期值= 8。PIN可以指向这有什么问题吗？结果应该具有8个值，但它打印出1。任何人可以提供帮助吗？

#include <stdio.h>`
#include <assert.h>
//define array size 8
#define ARRAY_SIZE 8
__global__ void vecAddKernel(int * A_d) {
//thread Index
unsigned int t = threadIdx.x;
    for (unsigned int stride = blockDim.x / 2; stride > 0; stride /= 2) {
    __syncthreads();
    if (t < stride)
        A_d[t] += A_d[t + stride];
    }
}
int main(int argc, char * * argv) {
    int A_h[ARRAY_SIZE];
   // initializing all values in A_h array to 1
    for (int i = 0; i < ARRAY_SIZE; i++) {
        A_h[i] = 1;
    }
    int * A_d, result;
   // reserving size array A_d of 8 in cuda
    cudaMalloc((void * * ) & A_d, ARRAY_SIZE * sizeof(int));
    cudaMemcpy(A_d, A_h, ARRAY_SIZE * sizeof(int), cudaMemcpyHostToDevice);
    vecAddKernel << < 1, ARRAY_SIZE / 2 >>> (A_d);
   Copy the first index of A_d to the result.
    cudaMemcpy( &result, &A_d[0], sizeof(int), cudaMemcpyDeviceToHost);
  // outputting the value of result
    printf("Result = %dn", result);
    //freeing the memory
    cudaFree(A_d);
}

Answer 1

我不确定您如何获得Result = 1。

当我编译并运行您的代码时，我会看到Result = 4。这是因为内核内部循环中 stride的初始值应为 blockDim.x，而不是 blockDim.x / 2（循环的第一个迭代应添加由 ARRAY_SIZE / 2隔开的值对，而 blockDim.x已经是 CC_7）。

）。

用unsigned int stride初始化器中的CC_1替换blockDim.x / 2使程序正确。

如果您有兴趣执行这样的阵列减排，则可能需要查看__shfl_down和其他随附的Shuffle函数，其中包括Kepler：https：//devblogs.nvidia.com/parallelforall/faster-paraster-parelled-parellect-parelalle-reductions-reductions-reductions-reductions-reductions-reductions-reductions-kepler/