NVIDIA CUDA初了解
NVIDIA的CUDA（Compute Unified Device Architecture，计算统一设备架构）已成为利用图形处理器（GPU）进行通用计算（GPGPU）的关键平台。它不仅提供了一个并行计算的编程模型，还包含了一整套支持软件开发的工具链。本文旨在对CUDA的架构、编程模型及其典型工作流程进行梳理和解析。
CUDA编程模型：并行计算的基石
传统的CPU编程多侧重于串行任务或有限的并行度。CUDA则引入了一种大规模并行的编程范式。
那么，CUDA是如何组织并行任务的呢？
在CUDA中，基本的执行单位是线程 (Thread)。众多线程被组织成线程块 (Thread Block)，同一线程块内的线程可以相互协作，例如通过共享内存进行数据交换和同步。多个线程块则构成一个网格 (Grid)。开发者将计算任务分解为可以在大量线程上并行执行的独立或协作部分，这些部分被称为核函数 (Kernel)，由GPU上的众多核心执行。
图示：CUDA软件开发工具包（SDK）的核心组件。绿色部分通常指代与GPU或NVIDIA技术紧密相关的组件，蓝色部分则偏向于CPU或通用组件。
CUDA软件开发工具包 (SDK) 概览
上图展示了CUDA SDK的主要构成。理解这些组件及其交互对于CUDA开发至关重要。
1. 源代码与预置库
开发者通常使用CUDA C/C++进行编程，这是一种对标准C/C++的扩展。
- Integrated CPU + GPU C Source Code: 源代码文件（通常为
.cu文件）可以同时包含在CPU上执行的主机代码 (Host Code)和在GPU上执行的设备代码 (Device Code/Kernels)。 - CUDA Optimized Libraries: NVIDIA提供了针对多种计算密集型任务（如FFT、BLAS、稀疏矩阵运算等）高度优化的库。这些库能够帮助开发者快速实现高性能的GPU加速，而无需从头编写底层核函数。
2. 编译流程
CUDA C/C++代码是如何被转换成可执行指令的？ 这个过程主要由NVIDIA C Compiler (NVCC)驱动。
- NVIDIA C Compiler (NVCC): NVCC是CUDA的核心编译器。它负责解析
.cu文件，将设备代码编译为GPU可执行的指令，同时分离出主机代码。 - NVIDIA Assembly for Computing (PTX): NVCC可以将设备代码编译为PTX（Parallel Thread Execution）代码。PTX是一种为NVIDIA GPU设计的中间汇编语言，具有指令集虚拟化的特性。这意味着为某一版本GPU编译的PTX代码，可以通过CUDA驱动程序在新架构的GPU上即时编译（JIT）为特定机器码，从而实现一定程度的前向兼容性。
- CPU Host Code & Standard C Compiler: NVCC会将分离出的主机代码传递给系统上安装的标准C/C++编译器（如GCC, MSVC, Clang等），由它们完成主机端可执行文件的编译。
3. 驱动、运行时与工具
- CUDA Driver: CUDA驱动程序是操作系统内核与GPU硬件之间的桥梁。它负责管理GPU资源、加载并执行编译好的GPU代码、以及处理CPU与GPU之间的数据传输。值得注意的是，CUDA驱动程序针对计算任务进行了优化，虽然它依赖于底层的图形驱动，但其API不直接提供图形渲染功能。不过，它支持与图形API（如OpenGL, DirectX）的互操作，允许在计算和图形渲染之间高效共享数据。
- Debugger & Profiler: CUDA SDK包含了一系列开发工具，如
cuda-gdb调试器和NVIDIA Nsight套件（Nsight Systems, Nsight Compute等）用于性能分析。这些工具对于代码调试、性能瓶颈定位和优化至关重要。
4. 硬件基础
最终，代码在物理硬件上执行：
- GPU: 执行设备代码（Kernels）的并行处理器。
- CPU: 执行主机代码，控制程序流程和任务调度。
CUDA程序典型工作流程
一个典型的CUDA应用程序执行流程大致如下：
- 初始化与数据准备 (CPU): 主机代码负责初始化环境，分配CPU和GPU内存。
- 数据传输 (CPU to GPU): 将需要GPU处理的数据从CPU内存拷贝到GPU内存。
- Kernel执行配置 (CPU): 主机代码设置Kernel的执行参数，如Grid和Block的维度，以及共享内存大小等。
- Kernel启动 (CPU -> GPU): 主机代码调用Kernel函数，CUDA驱动程序将Kernel加载到GPU上执行。
- 并行计算 (GPU): GPU上的众多核心并行执行Kernel代码。
- 数据传回 (GPU to CPU): Kernel执行完毕后，主机代码将计算结果从GPU内存拷贝回CPU内存。
- 后续处理与资源释放 (CPU): CPU对结果进行进一步处理或输出，并释放已分配的CPU和GPU资源。
CUDA编程中的内存管理
在CUDA编程中，内存管理有哪些需要注意的方面？
与许多高级语言或CPU编程环境不同，CUDA编程通常要求开发者显式管理GPU内存。 这种显式管理为开发者提供了对内存布局和数据移动的精细控制，这对于性能优化非常关键。然而，它也增加了编程的复杂性，需要开发者谨慎处理内存分配、拷贝和释放，以避免内存泄漏或访问错误。
CUDA API的功能覆盖领域
CUDA API提供了一系列广泛的功能，以支持GPU计算的各个方面，具体包括：
| 管理 | 功能 |
|---|---|
| Devicemanagement | 查询系统中可用的CUDA兼容设备数量、获取各设备属性（计算能力、内存大小、核心数量等）、选择特定设备进行计算任务以及管理设备状态 |
| Context management | CUDA上下文类似于CPU进程在GPU上的一个抽象，它封装了特定设备上的资源和状态，如已分配的内存、加载的模块等。API允许创建、管理和销毁上下文，确保不同计算任务之间的隔离。通常情况下，CUDA运行时API会为每个设备隐式管理一个主上下文。 |
| Memory management | 这是CUDA编程的核心部分之一。API提供了在GPU设备内存（全局内存、常量内存、纹理内存等）以及主机（CPU）内存与设备内存之间进行显式分配、释放和数据传输的功能。例如cudaMalloc(), cudaMemcpy(), cudaFree()等函数。 |
| Code module management | 涉及将在GPU上执行的已编译代码（通常是PTX中间代码或cubin二进制对象）的加载、卸载和管理。这允许开发者动态地将计算核心（kernels）加载到GPU。 |
| Execution control | 这部分API负责配置和启动GPU上的计算核心（kernels）。包括设置kernel的执行维度（Grid和Block的大小）、指定传递给kernel的参数、同步主机与设备的执行（例如等待kernel完成）等。 |
| Texture reference management | 纹理内存在硬件层面支持一些特殊的数据访问模式，如缓存优化、插值和边界寻址模式。虽然源于图形处理，但在通用计算中，纹理内存对于某些具有特定访存模式的应用（如图像处理、物理模拟中的场数据访问）可以提供性能优势。API允许绑定内存区域到纹理参考，并配置其属性。 |
| Interoperability with OpenGL and Direct3D | 为了支持计算与图形渲染的结合，CUDA API提供了与主流图形API（OpenGL和Direct3D）共享资源的机制。例如，可以将图形API创建的顶点缓冲区或纹理直接映射到CUDA的地址空间进行计算，计算结果也可以直接更新到这些图形资源中，避免了昂贵的内存拷贝。 |
CUDA API的两个主要层次
CUDA平台提供了两种不同抽象层次的API，以满足不同开发的需求：
-
A low-level API called the CUDA driver API (底层API：CUDA驱动API) : 驱动API是更接近硬件的接口。它提供了对CUDA设备和上下文更细致、更直接的控制。使用驱动API，开发者需要显式管理更多的细节，例如上下文的创建与销毁、模块的加载等。这种API通常用于需要高度控制或进行库开发的场景。它的函数通常带有
cu前缀（例如cuInit(),cuDeviceGet(),cuModuleLoad()）。 -
A higher-level API called the C runtime for CUDA that is implemented on top of the CUDA driver API (高层API：CUDA C运行时API) : 运行时API是在驱动API之上构建的一个抽象层。它为开发者提供了更便捷、更易用的接口，封装了许多驱动API的复杂操作。例如，上下文管理在运行时API中通常是隐式的。这是大多数应用程序开发者首选的API，因为它简化了编程模型，提高了开发效率。其函数通常带有
cuda前缀（例如cudaGetDeviceCount(),cudaMalloc(),cudaMemcpy()）。实际上，许多运行时API的调用最终会转换为一个或多个驱动API的调用。
Runtime API
如果你想用CUDA写出点“真东西”，而不是停留在 hello_cuda_world，那Runtime API就是你绕不开的话题。
🎯 首先，目标明确：我们要写“真家伙”！
Goal: writing real applications.”
这里的“real applications”可不是随便说说。想想看，实际的GPU应用会面临各种复杂的运行时环境：
- 硬件多样性：用户的显卡型号五花八门。
- CUDA能力各异：不同的卡支持的CUDA Compute Capability不同，这意味着某些特性可能可用，也可能不可用。
- GPU数量不定：单GPU还是多GPU系统？
- 还有很多其他配置相关的常量...
如果所有这些都得靠程序员用一堆 #define 和条件编译来手动处理，那代码的可维护性和可移植性简直是灾难。
这时候，Runtime API 就站出来了：“Runtime API helps with most of this.” 它提供了一套标准的接口，帮助我们查询设备属性、管理设备上下文、控制执行流等等，极大地简化了开发。
♻️ 资源的生命周期管理：初始化与清理
任何一个严肃的程序库，都离不开资源的初始化与清理。CUDA 也不例外。“As with every program, we need to be able to initialize/clean up CUDA when we need/don't need it anymore.”
道理很简单：“Don't waste resources when CUDA isn't needed!” 毕竟GPU资源宝贵，用完及时释放是个好习惯，也是专业程序员的基本素养。
✨ “智能”的自动初始化
有意思的地方来了：“Any Runtime API call will automatically initialize the CUDA runtime.”
比如，你调用任何一个以 cudaSomeFunc() 形式（通常是 cudaMalloc、cudaMemcpy 等）开头的函数，CUDA运行时环境就会被“悄悄地”初始化好。
你可能会问：这种“自动挡”设计有什么讲究吗？
确实，这种设计主要是为了方便开发者。你不需要在程序开头显式地写一段 cudaInit() 之类的代码（实际上也没有这个函数给Runtime API用户）。第一次调用相关的CUDA API时，系统会自动为你搞定底层的初始化工作，包括选择默认设备、创建上下文等。这降低了入门门槛，也减少了遗漏初始化的风险。
可以把它理解为一种“惰性初始化”（Lazy Initialization）的体现，即“按需初始化”。
🧹 资源的显式与自动回收
那么资源释放呢？
“cudaThreadExit() explicitly frees up cuda resources. Called automatically on CPU thread exit.”
这句话包含了两个关键信息：
- 显式释放：你可以通过调用
cudaThreadExit()函数，主动告诉CUDA：“嘿，我这个CPU线程不再需要之前关联的CUDA资源了，你可以回收它们了。” - 自动释放：即使你“忘记”或者因为某些原因没有显式调用
cudaThreadExit()，当当前CPU线程结束时，CUDA运行时系统也会自动帮你完成这个线程所占用的CUDA资源的清理工作。
这里就有个小细节了：既然CPU线程退出时会自动调用，那我还需要手动调用 cudaThreadExit() 吗？
答案是：看情况，但通常推荐在适当的时候显式调用。
- 依赖自动调用：如果你的CUDA操作局限在某个CPU线程的生命周期内，并且在该线程结束时释放资源是可接受的，那么确实可以依赖自动调用。这在一些简单的、生命周期短暂的线程中可能是没问题的。
- 显式调用的好处：
- 更精细的控制：如果你希望在线程退出前提早释放与该线程关联的CUDA资源（比如在一个主循环中，某个阶段不再需要GPU，但线程还要继续执行其他任务），那么显式调用
cudaThreadExit()就很有必要了。它可以让你的程序更精细地控制资源使用，避免不必要的占用。 - 确定性：显式调用可以让你更清楚地知道资源在何时被释放，这对于调试和性能分析可能更有利。
- 避免潜在问题：虽然有自动回收机制，但在复杂的应用中，过度依赖自动机制有时可能会掩盖一些资源管理上的逻辑错误。
- 更精细的控制：如果你希望在线程退出前提早释放与该线程关联的CUDA资源（比如在一个主循环中，某个阶段不再需要GPU，但线程还要继续执行其他任务），那么显式调用
所以，我的建议是：养成在不再需要CUDA资源时显式调用 cudaThreadExit() 的好习惯，尤其是在那些生命周期较长或者资源敏感的线程中。这会让你的代码更加健壮和可控。当然，对于一些临时创建的、很快就会结束的线程，依赖自动回收也是可以接受的。
关注杜子源源，AI、HPC新知，源源不断，带你冲向技术最前锋！
