CUDA中动态定义全局设备数组时cudaMemcpyToSymbol操作报错

你的代码存在两个核心错误，导致cudaMemcpyToSymbol执行失败：

1. 对__device__指针的本质理解错误 #

__device__ int *divi;定义的是设备端的指针变量，而非设备数组本身。你直接用cudaMemcpyToSymbol把主机端的data数组拷贝给这个指针变量，相当于试图把10个int的数据塞进一个仅占4/8字节的指针里，内存大小完全不匹配，必然触发拷贝失败。

正确逻辑应该是：

• 先在设备端分配一块能容纳k个int的内存空间
• 让divi指针指向这块设备内存
• 再把主机端的data数组拷贝到divi指向的设备内存中

2. cudaMemcpyToSymbol参数使用不规范 #

在C环境下，cudaMemcpyToSymbol的第一个参数应直接传入__device__变量名（无需加引号），引号形式是C语言兼容写法，虽部分编译器兼容，但不符合C规范，也容易引发隐性问题。

修正后的代码示例 #

#include <cuda_runtime.h>
#include <iostream>

const int k = 10;
__device__ int *divi; // 设备端指针，用于指向设备数组

void func()
{
    int data[k] = {1,2,3}; // 前三个值为1、2、3，其余为0
    int *d_data;

    // 步骤1：在设备端分配内存
    cudaMalloc(&d_data, k * sizeof(int));
    // 步骤2：让__device__指针divi指向这块设备内存
    cudaMemcpyToSymbol(divi, &d_data, sizeof(int*), 0, cudaMemcpyHostToDevice);
    // 步骤3：把主机端data数组拷贝到divi指向的设备内存
    cudaMemcpy(d_data, data, k * sizeof(int), cudaMemcpyHostToDevice);

    // 错误检查（建议每步CUDA操作都添加）
    cudaError_t err = cudaGetLastError();
    if (err != cudaSuccess) {
        std::cerr << "CUDA error: " << cudaGetErrorString(err) << std::endl;
    }
}

__global__ void kernel()
{
    int idx = threadIdx.x;
    if (idx < k) {
        // 示例：访问divi数组并打印值
        printf("divi[%d] = %d\n", idx, divi[idx]);
    }
}

int main()
{
    func();
    kernel<<<1,128>>>();
    cudaDeviceSynchronize(); // 等待核函数执行完成，确保输出能被捕获

    // 释放设备内存
    int *d_data;
    cudaMemcpyFromSymbol(&d_data, divi, sizeof(int*), 0, cudaMemcpyDeviceToHost);
    cudaFree(d_data);

    return 0;
}

额外注意事项 #

• 所有CUDA操作都应添加错误检查，便于快速定位问题
• 主机端的变长数组int data[k]是C99特性，部分编译器可能不支持，建议改为std::array<int, k> data = {1,2,3};或动态分配内存，提升兼容性
• 核函数执行后需调用cudaDeviceSynchronize()等待执行完成，否则主机端可能提前退出，无法捕获核函数的输出内容

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