CUDA并行计算——2.Introduction to CUDA C

这个系列是对CUDA并行计算课程的复习回顾，主要用作个人复习之用，难免有很多疏漏之处，错误的地方请在评论区指出，谢谢。

教材：Programming Massively Parallel Processors A Hands-On Approach 第3版

课程参照：伊利诺伊大学ECE408

下方是胡文美教授讲授ECE408的配套资源网页：

Home Pagehttp://gputeachingkit.hwu.crhc.illinois.edu/

目录

1.基本CUDA内存模型

2.基本CUDA执行模型

3.例子：向量加法

---

1.基本CUDA内存模型

这里简要介绍一下基本的CUDA内存模型，后续还会继续深入了解。上图展示的就是Host（CPU）端与Device（GPU）端进行数据传输， 在运行CUDA程序时，需要将数据从CPU端传输到GPU端进行运算，数据将会被存储到GPU端的global memory中。

上图是更详细的存储模型展示，但是这幅图缺少了shared memory以及纹理内存，以后再介绍。在上图中，GPU有三个层次：

• grid
• block
• thread

一个grid可以包含多个block，一个block可以包含多个thread。（这句话非常重要！）

其中，同一个grid的不同block中的thread可以共享global memory中的数据， 而同一个block中的thread可以共享shared memory中的数据（一个block有一个shared memory），每个thread自带私有的寄存器。

2.基本CUDA执行模型

上图中 展示的是CUDA的基本执行过程，它是一个异构Host(CPU) +Device(GPU)应用的C语言程序，串行部分使用Host的C语言代码，并行部分使用Device的kernel函数。

3.例子：向量加法

我将从向量加法这个例子来简单阐述一个CUDA程序的构成。下面展示的是传统的C语言写的向量加法： 

而要使用CUDA实现，下方展示的是Host端代码，你可以首先阅读以下的代码：

 作为初学者，阅读上述代码大都会一头雾水，下面我开始讲解这个代码的思路：

 首先介绍一下代码中出现的变量，对A、B两个向量相加，将结果存放到C中。在Host端的变量，前面添加h_，在Device端的变量，前面添加d_。如图所示，代码大致分为三个部分：

1. 为变量A、B、C在Device端分配内存空间，并将A、B从Host端复制到Device端。
2. 启动核函数，即在Device端实现向量加法
3. 将计算好的结果C从Devic端复制到Host端，并且释放Device端的变量

其中，第一步中分配内存空间使用 cudaMalloc()函数，复制变量使用cudaMalloc()函数。第三步中复制变量也使用cudaMalloc()函数，释放Device内存使用cudaFree()函数。关于核函数部分，将会在下一节进行讲解。

大致总结一下，CUDA程序需要三个步骤：

1. 将数据从CPU端传输到GPU端
2. 在GPU端启动核函数进行计算
3. 将GPU端计算的结果传输回CPU端

第二节结束，剩余部分将在之后的小节继续讲解。