CUDA编程模型（上）

作者：cpplinklist，联系方式：412012558@qq.com
本章节参考：CUDA C编程权威指南

编程思想与典型模式

cuda的思想主要在于将CPU视为主机，将GPU视为设备，我们可以把CPU认为是这个任务的完成人，操作系统将任务交给CPU，CPU必须按部就班的完成所有计算任务。但是有些任务确实很机械重复而且数量巨大，那么CPU可以把这部分简单而又机械重复的工作“外包给GPU”，也就是说，他们之间的关系是一种甲方和外包工厂的关系。在CUDA中，每次CPU发送给GPU的一个任务，称为一个kernel（核函数）。

图1 甲方和工厂的关系

那么从这个思想出发，如果一个甲方需要把一个任务外包给工厂，需要几个步骤。
1.工厂仓库腾出空间
2..把原材料从甲方运输到工厂
3..工厂把工作干完
4..把成品从工厂运输回来

图2 外包到工厂步骤

同样的CUDA编程也是这个逻辑
1.在显存上分配空间
2..把数据从CPU内存拷贝到GPU显存
3..调用核函数（kernel函数）
4..把计算完的数据从GPU显存拷贝回CPU内存

图3 CUDA代码逻辑
其中，GPU的显存和主机的内存之间通过PCIE数据线进行通信。

图片说明
图4 GPU与CPU之间的数据通信

那么我们就按照上述的步骤看几个重要的函数。

分配显存

在GPU上进行计算，要有数据，有数据就要分配内存。在C语言中，分配空间用malloc，CUDA中使用cudaMalloc可以在GPU上分配空间。

cudaError_t cudaMalloc(void **devPtr,size_t size) ;

看代码就很容易理解，和C语言的malloc是一样的，只不过这个函数malloc出来的内存是在GPU上的。*devPtr指向被分配的地址（使用的时候将指针地址传入函数参数，具体看下文完整代码申请GPU显存部分）

CPU与GPU之间的数据传输

数据要从CPU传输到GPU上，需要拷贝，C语言中的内存拷贝是memcpy。cuda中的拷贝是cudaMemcpy，它可以从主机拷到设备，也可以从设备拷到主机，当然，主机上某段内存和主机上另一段内存，设备上某段内存和设备上另一段内存也是可以的。

cudaError_t cudaMemcpy(void *dst, const void* src, size_t count,cudaMemcpyKind kind);

和C语言的memcpy是一样的，从src拷贝到dst，count指定数据的大小。
最后一个参数分为四种情况，cuda定义了四种宏。

cudaMemcpyHostToHost    //一段主机上的内存，拷贝到另一段主机上的内存
cudaMemcpyHostToDevice   //一段主机上内存，拷贝到设备上
cudaMemcpyDeviceToHost   //一段设备上的内存，拷贝到主机上
cudaMemcpyDeviceToDevice  //一段设备上的内存，拷贝到另一段设备上的内存

总结

那么好，到现在为止，我们可以认为cuda的代码大概应该长这个样子。（除去kernel函数）

#include<stdio.h>
#include<stdlib.h>
int main()
{
    //h开始的数据代表数据在主机上（Host），d开头的数据代表数据在设备（Device）

    //sum on GPU 计算两个向量的加法，C = A+B

    int nBytes = 1024;
    float *h_A, *h_B, *h_C;
    float *d_A, *d_B, *d_C;
    h_A = (float *)malloc(nBytes);
    h_B = (float *)malloc(nBytes);
    h_C = (float *)malloc(nBytes);

    //申请GPU上的显存
    cudaMalloc((float **)&d_A,nBytes);
    cudaMalloc((float **)&d_B,nBytes);
    cudaMalloc((float **)&d_C,nBytes);

    //拷贝CPU上的数据到GPU
    cudaMemcpy(d_A,h_A,nBytes,cudaMemcpyHostToDevice);
    cudaMemcpy(d_B,h_B,nBytes,cudaMemcpyHostToDevice);

    ****** kernel fuction ******
    ***** add d_A + d_B on GPU *******
    ****** finished *******

    cudaMemcpy(h_C,d_C,nBytes,cudaMemcpyDeviceToHost);

    //we get d_C on GPU
}

以上代码就是一个cuda程序的逻辑了，至于kernel函数怎么写，在下一节讲解。