CUDA设备端数组初始化代码运行终端挂起问题求助

核心问题分析 #

你的代码运行挂起/异常主要由以下几个严重问题导致：

• 未初始化的共享内存变量引发未定义行为
  __shared__ int temp; 没有初始化，其值为随机垃圾值。同时，block内所有线程会同时读写这个共享变量，引发严重的数据竞争，导致GPU执行逻辑混乱，甚至硬件级别的执行挂起。

• 线程网格大小计算错误+无边界检查
  ARRAY_SIZE=100，BLOCK_SIZE=32，dimgrid=ARRAY_SIZE/BLOCK_SIZE 得到的是3，总线程数仅为3*32=96，无法覆盖数组最后4个元素（索引96-99）。且kernel中没有判断 x < size，当线程索引超出数组范围时，会触发越界内存访问，直接导致程序异常。

• 错误的初始化逻辑
  试图用共享变量来维护全局计数是完全错误的方向——每个线程应该独立计算自己负责的元素值，不需要依赖共享变量同步（此处场景根本不需要共享内存）。

修正后的代码 #

#include <stdio.h>
#include <stdlib.h>

#define ARRAY_SIZE 100
#define BLOCK_SIZE 32

__global__ void cu_kernel(int *a_d, int size)
{
    int x = blockIdx.x * blockDim.x + threadIdx.x;
    // 边界检查：确保线程索引不超出数组范围
    if (x < size) {
        // 直接计算1-8循环的元素值：(索引%8)+1
        a_d[x] = (x % 8) + 1;
    }
}

int main(int argc, char *argv[])
{
    int *a_d, a_h[ARRAY_SIZE];

    // 仅分配需要的设备内存（b_d、c_d、sum_d当前逻辑用不到，可暂时移除）
    cudaMalloc((void**)&a_d, sizeof(int) * ARRAY_SIZE);

    // 正确计算网格大小：向上取整，确保覆盖所有元素
    dim3 dimblock(BLOCK_SIZE);
    dim3 dimgrid((ARRAY_SIZE + BLOCK_SIZE - 1) / BLOCK_SIZE);

    // 调用kernel
    cu_kernel<<<dimgrid, dimblock>>>(a_d, ARRAY_SIZE);
    // 检查kernel执行错误（这一步非常重要，能快速定位GPU端问题）
    cudaError_t kernel_err = cudaGetLastError();
    if (kernel_err != cudaSuccess) {
        fprintf(stderr, "Kernel execution failed: %s\n", cudaGetErrorString(kernel_err));
        exit(1);
    }
    // 等待GPU任务完成（确保kernel执行完毕再拷贝数据）
    cudaDeviceSynchronize();

    // 拷贝结果到主机端
    cudaError_t copy_err = cudaMemcpy(a_h, a_d, sizeof(int) * ARRAY_SIZE, cudaMemcpyDeviceToHost);
    if (copy_err != cudaSuccess) {
        fprintf(stderr, "cudaMemcpy failed: %s\n", cudaGetErrorString(copy_err));
        exit(1);
    }

    // 打印结果
    printf("Final state of the array:\n");
    for (int i = 0; i < ARRAY_SIZE; i++) {
        printf("%d ", a_h[i]);
    }
    printf("\n");

    // 释放设备内存
    cudaFree(a_d);
    return 0;
}

关键修改说明 #

1. 移除无用变量：删除了当前逻辑不需要的b_d、c_d、sum_d，简化代码。
2. 正确的线程网格计算：用(ARRAY_SIZE + BLOCK_SIZE - 1) / BLOCK_SIZE向上取整，确保总线程数覆盖所有数组元素。
3. 添加边界检查：在kernel中判断x < size，避免越界访问。
4. 简化初始化逻辑：直接通过(x % 8) +1生成1-8循环的序列，每个线程独立计算，无数据竞争。
5. 添加错误检查：调用cudaGetLastError()检查kernel执行错误，cudaDeviceSynchronize()确保GPU任务完成后再拷贝数据，方便快速定位问题。
6. 释放设备内存：添加cudaFree(a_d)避免内存泄漏。

后续点积实现提示 #

如果要实现点积，你可以：

• 首先在kernel中计算两个数组对应元素的乘积，存入临时数组（或直接在共享内存中累加）。
• 使用归约算法对乘积数组求和，最终将结果返回主机端。归约是CUDA中常用的并行求和方法，能高效利用GPU的并行能力。

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