CUDA简介

pqin

已于 2023-12-21 14:14:29 修改

阅读量1.8k

点赞数 18

文章标签： c++ c语言

于 2023-12-21 13:55:42 首次发布

本文链接：https://blog.csdn.net/weixin_43308899/article/details/135128349

版权

说在前面：本文主要介绍通过CUDA进行gpu编程的一些前置知识,包括GPU硬件结构和软件层面的一些内容。

一. 什么是CUDA？

CUDA（Compute Unified Device Architecture）是由NVIDIA开发的一个并行计算平台和应用编程接口（API）模型。它允许软件开发者和软件工程师利用NVIDIA GPU（图形处理单元）进行通用的计算处理。简而言之，CUDA使得开发者能够使用NVIDIA的GPU进行高效的数值计算，从而加速计算密集型任务。以下是CUDA的一些关键特点和用途：

并行处理：CUDA提供了一种方法，使得开发者能夠在GPU上进行大规模的并行计算。由于GPU设计用于处理复杂的图形和图像处理任务，它们非常适合于执行大量的并行操作。
编程模型：CUDA提供了一组扩展的C、C++和Fortran语言的编程接口，使得开发者能夠在GPU上编写程序。这些程序通常包括在CPU上运行的主机代码和在GPU上运行的设备代码。
应用领域：CUDA广泛用于科学计算、工程、机器学习、深度学习、数据分析和图形渲染等领域。通过在GPU上运行计算密集型任务，CUDA可以显著加速应用程序的执行。
内存管理：CUDA提供了对GPU内存的直接访问和管理，允许开发者控制数据在主机和GPU之间的传输。
架构优化：CUDA平台设计用于充分利用NVIDIA GPU的架构，包括它们的多核处理能力和高速内存访问。
生态系统：随着CUDA的普及，围绕CUDA开发了一个庞大的生态系统，包括各种库、框架和工具，如cuBLAS、cuDNN、NVIDIA Nsight等，它们进一步简化了在GPU上进行高性能计算的过程。

总的来说，CUDA是一个强大的平台，它开启了利用GPU进行高效计算的可能性，对于需要大规模并行处理的应用来说尤其有价值。

二. GPU的内存结构

全局内存（Global Memory）:
- 也被称为设备内存，它是GPU上的主要内存资源。
- 用于存储大量的数据，如纹理、顶点数据、帧缓冲等。
- 由GPU和CPU共享，允许数据在主机（CPU）和设备（GPU）之间传输。
共享内存（Shared Memory）:
- 一种更快的内存类型，可由同一个线程块（Block）内的所有线程共享。
- 适用于需要线程间通信或数据重用的场景。
寄存器（Registers）:
- 每个线程都有自己的寄存器集，用于快速访问其私有变量。
- 访问速度非常快，但每个线程可用的寄存器数量有限。
常量内存（Constant Memory）:
- 一种特殊的小容量内存，用于存储所有线程都可读取的常量数据。
- 由专门的高速缓存支持，适用于频繁访问的不变数据。
纹理内存（Texture Memory）:
- 专门用于图形处理的内存，拥有特殊的缓存机制。
- 适用于图形处理中的纹理映射和滤波操作。
本地内存（Local Memory）:
- 当寄存器资源耗尽时，一些局部变量可能会被放置在本地内存中。
- 本质上是全局内存的一部分，因此访问速度较慢
L1缓存（一级缓存）:
- L1缓存通常与GPU中的每个流处理器（Streaming Multiprocessors, SMs）或计算单元（Compute Units, CUs）紧密相关联。
- 它是最快的缓存类型，提供极低的访问延迟。
- L1缓存通常用于存储线程块（Thread Block）内部频繁访问的数据。
- 在很多GPU架构中，L1缓存和共享内存（Shared Memory）可能共享相同的物理内存空间，但具体取决于特定的GPU架构。
L2缓存（二级缓存）:
- L2缓存是位于GPU核心和全局内存之间的一个更大但稍慢的缓存层级。
- 它是所有SMs或CUs共享的，用于减少对全局内存的访问次数和延迟。
- L2缓存对于提高内存访问效率和减少带宽使用非常重要，特别是在处理大量数据时。

三. CUDA中的线程组织结构

线程（Thread）:
- 最基本的执行单元。
- 每个线程执行相同的代码，但可以通过其唯一的索引进行不同的操作。
线程块（Thread Block）:
- 一组可以协同工作的线程。
- 块内的所有线程可以通过共享内存快速共享数据，并能够通过同步指令协调执行。
- CUDA中每个线程块的线程数有上限，这个上限取决于CUDA版本和使用的GPU架构。
网格（Grid）:
- 多个线程块形成一个网格。
- 网格代表了整个CUDA内核的执行实例。
- 网格可以是一维、二维或三维的，这允许灵活地组织线程以适应不同类型的计算问题。

四. gridDim, blockDim, blockIdx, threadIdx

这里我们通过代码来说明这四者的含义：

__global__ void vec_add(FLOAT *x, FLOAT *y, FLOAT *z, int N)
{
    /* 2D grid */
    int idx = (blockDim.x * (blockIdx.x + blockIdx.y * gridDim.x) + threadIdx.x);
    /* 1D grid */
    // int idx = blockDim.x * blockIdx.x + threadIdx.x;
    if (idx < N) z[idx] = y[idx] + x[idx];
}

这是一个cuda核函数，__global__用于申明这是一个核函数，在gpu上执行，这个函数用于实现将两个向量相加并将结果存储在第三个向量中。每个核函数都会对应一个grid，gridDim.x是指grid在x方向上有多少个block，blockDim.x是每个block在x方向上有多少个thread，blockIdx.x指的是当前block在grid中的x方向上的id，threadIdx是指当前thread在当前block的x方向上的id，这里需要说明网格grid和线程块都有xyz三个维度，具体需要用户自己定义。

五. 什么是warp？

在NVIDIA的CUDA架构中，warp是基本的执行单元，是GPU进行资源调度的基本单位。一个warp包含了一定数量的线程（在当前NVIDIA架构中通常是32个线程），这些线程同时执行相同的指令，但是可能对不同的数据进行操作。

六. 什么是SM?

"SM" 指的是 "流式多处理器"（Streaming Multiprocessor）。SM是NVIDIA CUDA架构GPU的关键组成部分，负责执行计算任务。以下是SM的一些重要特点：

并行处理单元：每个SM包含多个核心，这些核心可以并行处理计算任务。在NVIDIA的GPU中，这些核心通常被称为CUDA核心。
任务分配：GPU中的SM负责处理分配给它的线程块（thread blocks）。线程块是由多个线程组成的，这些线程可以执行相同或不同的指令，并处理不同的数据。
资源配置：每个SM拥有一定数量的资源，如寄存器、共享内存、L1缓存等。这些资源被分配给它处理的线程块。
与Warp的关系：在NVIDIA的架构中，一个SM可以同时处理多个warp。warp是基本的执行单元，包含固定数量的线程（比如32个），这些线程以SIMD（单指令多数据）方式执行指令。
性能优化：对SM的理解对于优化GPU性能非常重要。有效地利用SM的资源和最大化其处理能力是高效GPU编程的关键。
不同架构的差异：不同代的NVIDIA GPU在SM的设计和数量上有所不同，这影响了它们的性能和计算能力。

这里多说一点，不同的block会被优先分配到不同的SM中，原因在于在同一个SM里的不同warp并不是严格并行的，分配到不同的SM可以最大化并行度。

小知识

计算密集型任务和I/O密集型的区别：

计算密集型：
- 这类任务的特点是需要大量的计算，即 CPU 时间。
- 它们通常涉及到复杂的数学运算，如数据分析、图像处理、科学计算等。
- 在这种情况下，CPU 的性能通常是性能瓶颈。
I/O 密集型：
- 这类任务的特点是需要大量的数据读写操作，即对存储设备或网络的输入/输出操作。
- 它们可能涉及到数据库操作、文件处理、网络通信等。
- 在这种情况下，存储设备的读写速度或网络带宽通常是性能瓶颈。