cuda 全局device变量的用法

device全局变量的使用主要用到了两个函数：

cudaMemcpyToSymbol（）

cudaMemcpyFromSymbol（）

__constant__ float constData[256];
float data[256];
cudaMemcpyToSymbol(constData, data, sizeof(data));
cudaMemcpyFromSymbol(data, constData, sizeof(data));

__device__ float devData;
float value = 3.14f;
cudaMemcpyToSymbol(devData, &value, sizeof(float));

__device__ float* devPointer;
float* ptr;
cudaMalloc(&ptr, 256 * sizeof(float));
cudaMemcpyToSymbol(devPointer, &ptr, sizeof(ptr));

假设下面代码保存在var.cu

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


using namespace std;


__device__ int d_data;


__global__ void changeData()
{
    d_data = 567;
}


int main()
{
    cudaError_t err;
    int h_data = 123;
    cudaSetDevice(0);
    cout<<"before kernel, h_data = "<<h_data<<endl;
    changeData<<<1,1>>>();
    err = cudaMemcpyFromSymbol((void*)&h_data, d_data, sizeof(int));
    if(err != cudaSuccess)
    {
        cout<<"from symbol error!"<<endl;
        exit(0);
    }
    cout<<h_data<<endl;
    cudaDeviceReset();
    return 0;
}

编译文件：nvcc var.cu -o var -arch=sm_35

运行可执行文件: ./var

before kernel, h_data =123

567