CUDA单线程每块运行正常，多线程每块执行报错求助

问题概述 #

我是CUDA初学者，正在做基础实验：有一个1D数组arr1，元素从99递减到0（即<99,98,...2,1,0>），核函数逻辑是取索引i处的元素n，将arr1[n]赋值给新数组arr2的索引i处，预期arr2结果为<0,1,2,...,97,98,99>。但该功能仅在线程块大小设为1时正常运行，当用shuffle<<<25,4>>>（25个线程块、每块4个线程）或shuffle<<<25,5>>>调用时，会出现索引错误值。

错误原因 #

核函数中全局线程索引计算逻辑完全错误，这是CUDA入门阶段的典型问题：

• 错误的索引公式：int i = threadIdx.x + blockDim.x + blockIdx.x;，不符合CUDA全局线程索引的计算规则，导致大量线程的起始索引偏移，甚至直接遗漏部分索引（比如线程块大小为4时，索引0-3没有线程处理）。
• 循环步长错误：i += blockDim.x，步长应设置为总线程数（gridDim.x * blockDim.x），否则会出现线程重复处理部分元素、同时遗漏其他元素的情况。

标准的CUDA全局线程索引计算公式为：

int i = threadIdx.x + blockIdx.x * blockDim.x;

循环步长需设置为总线程数，确保每个元素被恰好一个线程处理：

i += gridDim.x * blockDim.x;

修正后的完整代码 #

#include<stdio.h>
#include<stdlib.h>
#include<math.h>
#include<cuda.h>
#include<cuda_runtime.h>

__global__ void shuffle(int *arr1,int *arr2,int n){
    // 修正全局线程索引计算
    int i = threadIdx.x + blockIdx.x * blockDim.x;
    // 步长改为总线程数
    while(i < n){
        arr2[i] = arr1[arr1[i]];
        i += gridDim.x * blockDim.x;
    }
}

int main(){
    int m=10,n=10;
    int num = m*n;
    int size = num*sizeof(int);

    int *arr1,*arr2;
    arr1 = (int*)malloc(size);
    arr2 = (int*)malloc(size);

    // 初始化arr1为元素从99递减到0
    for(int i=0; i<num; i++){
        arr1[i] = 99 - i;
    }

    int *d_arr1,*d_arr2;
    cudaMalloc(&d_arr1,size);
    cudaMalloc(&d_arr2,size);

    cudaMemcpy(d_arr1,arr1,size,cudaMemcpyHostToDevice);

    // 任意合法的线程配置都可正常运行，示例：25块×4线程
    shuffle<<<25,4>>>(d_arr1,d_arr2,num);
    // 添加CUDA运行错误检查，方便调试
    cudaError_t err = cudaGetLastError();
    if(err != cudaSuccess){
        printf("CUDA kernel error: %s\n", cudaGetErrorString(err));
        return 1;
    }

    cudaMemcpy(arr2,d_arr2,size,cudaMemcpyDeviceToHost);

    long error = 0;
    printf("Num: %d\n",num);

    // 验证结果
    for(int i=0;i<num;i++){
        if(arr2[i] != i){
            error+=1;
            printf("索引%d: 预期%d，实际%d\n",i,i,arr2[i]);
        }
        else{
            printf("索引%d: %d\n",i,arr2[i]);
        }
    }

    printf("错误数: %ld\n",error);

    // 释放内存，避免泄漏
    free(arr1);
    free(arr2);
    cudaFree(d_arr1);
    cudaFree(d_arr2);

    return 0;
}

额外说明 #

1. 修正了arr1的初始化逻辑：原代码中for(int i=num-1;i>=0;i--){arr1[i] = i;}实际是让arr1[i]=i（元素从0递增到99），和描述的“从99递减到0”不符，已调整为arr1[i] = 99 - i。
2. 添加了CUDA错误检查：调用核函数后通过cudaGetLastError()捕获运行时错误，快速定位问题。
3. 补充了内存释放操作：规范内存管理，避免内存泄漏。

内容的提问来源于stack exchange，提问作者Manav Karthikeyan