GPU 高性能编程 CUDA : CUDA C 并行编程

基于 GPU 的矢量求和：

#define N = 10;

int main(){
    int a[N],b[N],c[N];
    int *dev_a,*dev_b,*dev_c;
    //GPU上分配内存
    cudaMalloc((void**)&dev_a,N*sizeof(int));
    cudaMalloc((void**)&dev_b,N*sizeof(int));
    cudaMalloc((void**)&dev_c,N*sizeof(int));
    //在CPU上为数组 a 和 b 赋值
    for(int i=0;i<N;i++){
        a[i]=i;
        b[i]=i*i;
    }
    //将数据传入设备
    cudaMemcpy(dev_a,a,N*sizeof(int),cudaMemcpyHostToDevice);
    cudaMemcpy(dev_b,b,N*sizeof(int),cudaMemcpyHostToDevice);
    //调用核函数
    add<<<N,1>>>(dev_a,dev_b,dev_c);
    //将数组 c 复制到主机上
    cudaMemcpy(c,dev_c,N*sizeof(int),cudaMemcpyDeviceToHost);
    //释放GPU上分配的内存
    cudaFree(dev_a);
    cudaFree(dev_b);
    cudaFree(dev_c);
    return 0;
}

_global_ void add(int *a,int *b,int *c){
    int tid=blockIdx.x;
    if(tid<N){
        c[tid]=a[tid]+b[tid];
    }
}

bolckIdx 是一个内置变量，变量中包含的值就是当前执行设备代码的线程块的索引

cudaMemcpy() 接受四个参数，第一个是指针，指向数据复制操作的目的地址，第二个参数是指向要复制的源数据对象，第三个参数是指定要复制的数据的大小，第四个参数指向复制中所涉及的存储器的类型

基于 GPU 的 Julia 集计算：

int main(){
    CPUBitmap bitmap(DIM,DIM);
    unsigned char *dev_bitmap;
    cudaMalloc((void**)&dev_bitmap,bitmap.image_size());
    dim3 grid(DIM,DIM);
    kernel<<<grid,1>>>(dev_bitmap);
    cudaMemcpy(bitmap.get_ptr(),dev_bitmap,bitmap.image_size(),cudaMemcpyDeviceToHust);
    cudeFree(dev_bitmap);
    return 0;
}

程序指定了多个并行线程块来执行函数 kernel , 类型 grid3 是 CUDA 头文件中定义的辅助类型来封装多维数组，类型 dim3 表示一个 3 维数组，可以用于指定启动的线程块的数量，CUDA 运行时希望得到一个三维的 dim3 值，但是当前并不支持三维的线程格，最后一维的大小为 1.

_global_ void kernel(unsigned char *ptr){
    int x=blockIdx.x;
    int y=blockIdx.y;
    int offset = x+y*gridDim.x;
    int juliaValue = julia(x,y);
    ptr[offset*4+0]=255*juliaValue;
    ptr[offset*4+1]=0;
    ptr[offset*4+2]=0;
    ptr[offset*4+3]=255;
}

gridDim 是内置变量，对所有的线程块来说，gridDim是一个常数，用来保存线程格每一维的大小，在这个例子中为（DIM,DIM）

_device_ int julia(int x,int y){
    //和cpp的代码一样,只是加了一个修饰符，表明代码在设备上运行
    return 0;
}