GPU加速（一）CUDA C编程及GPU基本知识

前 言

笔记来自深蓝学院《CUDA入门与深度神经网络加速》

补充：线程与线程

• 线程是进程中执行运算(CPU调度)的最小单位。同一类线程共享代码和数据空间；进程是资源分配的最小单位。每个进程都有独立的代码和数据空间。
• 多进程是指操作系统能同时运行多个任务（程序）；多线程是指在同一程序中有多个顺序流在执行。
• 一个线程只能属于一个进程，而一个进程可以有多个线程，但至少有一个线程。
• 资源分配给进程，同一进程的所有线程共享该进程的所有资源；处理机分给线程，即真正在处理机上运行的是线程。
• 线程在执行过程中，需要协作同步。不同进程的线程间要利用消息通信的办法实现同步。线程是指进程内的一个执行单元,也是进程内的可调度实体。

课堂笔记

1. GPU和CPU的基础知识

延迟：指令从发出到最终返回结果的时间。

吞吐量：单位时间处理指令的条数。

CPU特点（以减少延迟为导向）：

• 多级高速缓存结构：经常访问的内容放在低级缓存中，不经常访问的放在高级缓存中。
• 复杂的控制单元：分支预测机制（if else），流水线数据前送。
• 强大运算单元：整型浮点型复杂运算。

GPU特点（以增加指令吞吐为导向）：

• 缓存小：指令访问缓存的次数减少，提升内存访问指令的吞吐。
• 控制简单：没有分支预测和数据转发，使得控制简单的指令提升吞吐。
• 精简运算单元：计算单元多，以长延时流水线的形式提升运算简单指令的吞吐量

1.1 CPU与GPU对比

CPU适合处理复杂的指令（提升单指令的运行速度）；
GPU适合处理多个简单指令（提升指令的执行数量）。

1.2 什么样的问题适合GPU

频繁访问缓存和控制复杂的指令不适合使用GPU，例如频繁访问磁盘的杀毒软件程序。

2. CUDA C编程基本知识

CUDA并行计算的整体步骤：

• 主机端（CPU）申请内存到设备端（GPU）
• 设备端核函数（GPU上运行的函数）实现运算密集和并行特点的指令
• 设备端拷贝到主机端 释放现存和内存
  

2.1 CUDA 内存模型

CUDA中的内存模型根据硬件端和软件端分为几个层次。

2.1.2 硬件端

• 每个线程处理器（SP）都用自己的registers（寄存器）。
• 每个SP都有自己的local memory（局部内存），register和local memory只能被线程自己访问。
• 每个多核处理器（SM）内都有自己的shared memory（共享内存），shared memory 可以被线程块内所有线程访问。
• 一个GPU的所有SM共有一块global memory（全局内存），不同线程块的线程都可使用。

2.1.2 软件端

• 线程处理器（SP）对应线程（thread）
• 多核处理器（SM）对应线程块（thread block）
• 设备端（device）对应线程块组合体（grid）

3. 并行计算向量相加

• 主机端：1.申请显存内存，2.转移拷贝，4.释放。
  1. 申请显存（全局内存）的函数cudaMalloc( )，例如：cudaError_t cudaMalloc (void **devPtr, size_t size) ，两个参数：地址和申请内存大小。
  2. 从设备全局内存中释放对象cudaFree( )，例如cudaError_t cudaFree ( void* devPtr )，指向释放对象的指针。
  3. （同步）内存数据复制传递cudaMemcpy( )，例如
     cudaError_t cudaMemcpy (void *dst, const void *src, size_t count, cudaMemcpyKind kind)，从指针src拷贝到指针dst，count为执行大小，kind目前支持的四种选项：cudaMemcpyHostToDevice、cudaMemcpyDeviceToHost、cudaMemcpyDeviceToDevice、cudaMemcpyDefault。
• 设备端：3.读写线程寄存器，读写GRID的全局内存，读写block中的共享内存。