1.分配CPU内存和GPU全局内存空间

首先使用malloc分配三块内存，分别是存放数据源host_a和host_b以及存放向量相加的结果的host_c。其次采用cudaMalloc分配三块内存，分别是存储数据device_a和device_b以及存放向量相加的结果device_c。cudaMalloc的用法与malloc的用法相似。其具体代码如代码清单4-1所示。

2.初始化内存空间

本文通过采用rand（）函数随机分配数据给host_a和host_b，从而完成内存空间的初始化。其具体代码如代码清单4-2所示。

代码清单4-1 分配内存空间和GPU全局内存空间

代码清单4-2 初始化内存空间

下面来推导线程块号为blockIdx、线程号为threadIdx的线程完成向量计算的数组下标：块内地址为threadIdx.x∗blockDim.x+threadIdx.y，把线程号映射到[0，blockDim.x∗blockDim.y-1]。线程块地址为blockIdx.x∗gridDim.x+blockIdx.y，把块号映射到[0，gridDim.x∗gridDim.y-1]。所以该线程的地址为i=（blockIdx.x∗gridDim.x+blockIdx.y）∗blockDim.x∗blockDim.y+threadIdx.x∗blockDim.x。其具体的向量加法kernel函数定义如代码清单4-5所示。

代码清单4-5 kernel函数

3.将要计算的数据从CPU内存传输到GPU全局内存上

本文通过cudaMemcpy函数将host端的数据源host_a和host_b传到device端的数据源device_a和device_b。cudaMalloc与cudaMemcpy类似，不过cudaMemcpy多出一个参数，指示传输内存的方向。在这里因为是从host内存传输到device内存，所以使用cudaMemcpyHostToDevice。如果是从device内存传输到host内存，则使用cudaMemcpyDeviceToHost。其具体代码如代码清单4-3所示。

代码清单4-3 将要计算的数据从内存上传输到全局内存上

4.执行kernel

首先，需要完成kernel函数配置设置，其具体配置代码如代码清单4-4所示。初始化dimBlock为16∗16∗1的dim3类型，指定线程块的三个维度，这里第三维是1，即退化为16∗16的二维线程块。为了最大化并行，安排每一个线程负责一次向量加法，那么则需要ceil（n/（16∗16））个线程块，设置线程网络也为二维方阵的话，其gridsize取ceil（sqrt（ceil（n/（16∗16））））。同样初始化维度变量dim-Grid，注意Grid只能是二维以下的，第三个维度设置被自动忽略。设置中用了上取整，是为了保证至少有一个线程完成向量每对元素的相加，那么这样设置可能会导致线程数目多于向量长度，因此在Kernel函数中需要让这些线程直接退出，而不是引起数组超界的错误。

代码清单4-4 kernel函数配置

下面来推导线程块号为blockIdx、线程号为threadIdx的线程完成向量计算的数组下标：块内地址为threadIdx.x∗blockDim.x+threadIdx.y，把线程号映射到[0，blockDim.x∗blockDim.y-1]。线程块地址为blockIdx.x∗gridDim.x+blockIdx.y，把块号映射到[0，gridDim.x∗gridDim.y-1]。所以该线程的地址为i=（blockIdx.x∗gridDim.x+blockIdx.y）∗blockDim.x∗blockDim.y+threadIdx.x∗blockDim.x。其具体的向量加法kernel函数定义如代码清单4-5所示。(www.daowen.com)

代码清单4-5 kernel函数

5.将计算结果从GPU全局内存传输到CPU内存

在完成kernel计算后，需要把device端处理结果传回host端上。本文通过cu-daMemcpy函数将device端的处理结果decice_c传到host端口host_c内存上。其次，在使用完device后，我们需要将device的内存释放。其具体代码如代码清单4-6所示。

代码清单4-6 全局内存上的数据传输到内存上

5.将计算结果从GPU全局内存传输到CPU内存

在完成kernel计算后，需要把device端处理结果传回host端上。本文通过cu-daMemcpy函数将device端的处理结果decice_c传到host端口host_c内存上。其次，在使用完device后，我们需要将device的内存释放。其具体代码如代码清单4-6所示。

代码清单4-6 全局内存上的数据传输到内存上

6.结果数据处理

在获取完device端的处理结果decice_c后，可能我们还需进行进一步处理，比如打印处理结果。在完成处理后，我们需要将host的内存释放。其具体代码如代码清单4-7所示。

代码清单4-7 处理传输到内存上的数据

6.结果数据处理

在获取完device端的处理结果decice_c后，可能我们还需进行进一步处理，比如打印处理结果。在完成处理后，我们需要将host的内存释放。其具体代码如代码清单4-7所示。

代码清单4-7 处理传输到内存上的数据