【CUDA】C1 GPU基础与CUDA编程

最新推荐文章于 2025-11-23 15:46:28 发布

原创最新推荐文章于 2025-11-23 15:46:28 发布 · 1.1k 阅读

20 ·

CC 4.0 BY-SA版权

文章标签：

#CUDA #GPU #并行计算

部署运行你感兴趣的模型镜像

一、GPU基本知识与CUDA编程

1. CPU和GPU架构

1）CPU架构
●
● 核心特点：
○ 多级缓存结构：采用L1/L2/L3三级缓存，通过空间换时间策略提升存储访问速度
○ 复杂控制单元：
■ 分支预测机制：预判条件分支指令（如if-else/break）的执行路径
■ 流水线前传机制：将高频访问数据提前推送至流水线前端
○ 强大运算单元：支持整型/浮点四则运算及与或非等逻辑运算
● 设计导向：延迟优先
○ 延迟定义：指令发出到返回结果的时间间隔
2）GPU架构
●
● 核心特点：
○ 精简缓存：缓存容量显著小于CPU，减少内存访问次数
○ 简单控制单元：
■ 无分支预测机制
■ 无数据转发机制
○ 海量计算单元：
■ 采用长延时流水线设计
■ 单控制器管理多运算单元（SIMD架构）
● 设计导向：吞吐优先（单位时间指令执行量比CPU高10倍以上）
○ 吞吐量定义：单位时间内处理指令的条数
3）CPU与GPU特点
●
● 适用场景选择标准：
○ CPU优势场景：
■ 控制复杂指令（如含多重条件分支）
■ 计算复杂指令（如混合运算）
■ 典型案例：程序流程控制、操作系统核心
○ GPU优势场景：
■ 计算密集型：数值计算占比 > 内存访问占比
● 反例：网络传输、磁盘清理工具
■ 数据并行：可分解为相同指令的子任务
● 反例：时序任务（后一状态依赖前一状态）
● 形象比喻：
○ CPU如同1个大学生解决复杂问题
○ GPU如同1000个中学生并行处理简单任务

2. GPU编程与CUDA

1）CUDA
●
● 定义: CUDA(Compute Unified Device Architecture)是英伟达2007年推出的GPU编程架构
● 特点:
○ 提供易用编程接口，无需学习复杂着色语言或图形处理原语
○ 专为NVIDIA GPU设计
● 对比OpenCL:
○ OpenCL是2008年发布的异构平台并行编程开放标准
○ 支持设备更广泛(CPU/GPU/DSP/FPGA等)
○ 本课程以CUDA为例讲解GPU编程
2）CUDA编程并行计算整体流程
●
● 三阶段流程:
○ 内存分配与拷贝: 从主机内存(host memory)申请并拷贝到设备内存(device memory)
○ 核函数执行: 在设备端执行并行计算
○ 结果回传: 从设备内存拷贝回主机内存，释放显存
3）CUDA编程术语
● 硬件术语
○
○ 设备端(Device): 指GPU
○ 主机端(Host): 指CPU
○ 核函数(Kernel): GPU上运行的函数
○ 通信机制: 通过内存和显存互相拷贝传递参数
● 内存模型 07:55
○
○ 寄存器(Registers):
■ 每个线程处理器(SP)专用
■ 仅线程自身可访问
○ 局部内存(Local Memory):
■ 每个SP专用
■ 访问速度较慢
○ 共享内存(Shared Memory):
■ 每个多核处理器(SM)内共享
■ 线程块内所有线程可访问
○ 全局内存(Global Memory):
■ 所有SM共享
■ 不同线程块都可访问
● 软件模型
○
○ 线程(Thread): 对应线程处理器(SP)
○ 线程块(Thread Block): 对应多核处理器(SM)
○ 网格(Grid): 对应整个GPU设备
○ 线程块特性
■
■ 独立性: 块内线程计算彼此独立
■ 共享内存: 块内所有线程可访问共享内存
■ 同步机制: 可通过时钟同步块内所有线程
■ 典型应用: 如256线程并行执行向量加法
○ 网格特性
■
■ 组成: 多个线程块的组合体
■ 独立性: 网格内线程块彼此独立
■ 全局内存: 所有线程块可访问全局内存
■ 同步机制: 可同步网格内所有线程块
○ 线程索引计算
■
■ 索引维度: 线程块和线程索引均可为1D/2D/3D
■ 计算公式:
● 全局线程ID = 块ID × 块维度 + 线程ID
● 例如:i=blockIdx.x∗blockDim.x+threadIdx.xi = blockIdx.x * blockDim.x + threadIdx.xi=blockIdx.x∗blockDim.x+threadIdx.x
○ 线程束(Warp)
■
■ 定义: GPU执行的基本单元(32个线程)
■ SIMT架构: 单指令多线程执行模式
■ 调度特点:
● 线程块分配到SM执行
● 每个线程有自己的程序计数器和状态寄存器
■ 编程建议: 线程块大小设为32的倍数
4）内容回顾
● 核心概念:
○ 设备端(GPU)/主机端(CPU)/核函数
● 内存模型:
○ 硬件侧: SP→SM→Device
○ 软件侧: Thread→Thread Block→Grid
● 三大特性:
○ 内部元素独立
○ 共享内存访问
○ 时钟同步机制
● 关键计算:
○ 线程索引定位
○ 线程束约束(32倍数)
5）应用案例
● 例题:向量相加计算
○
○
○ 并行计算特点：向量相加是典型的并行计算案例，满足四个GPU适用条件：
■ 内存访问次数少：仅需简单读取A、B数组元素
■ 控制简单：无复杂分支预测和数据转换
■ 计算简单：仅需加法运算
■ 并行度高：各元素计算相互独立
○ CUDA实现原理：
■ 每个线程处理一对元素的加法
■ 线程索引通过i=threadIdx.x+blockDim.x∗blockIdx.xi = threadIdx.x + blockDim.x * blockIdx.xi=threadIdx.x+blockDim.x∗blockIdx.x计算
■ 需要判断i<ni<ni<n防止越界访问
■
■
○ 核函数定义：
■ 使用__global__修饰符声明
■ 参数为设备端指针AdA_dAd、BdB_dBd、CdC_dCd和向量长度nnn
■ 返回类型必须为void
○ 线程组织：
■ 每个线程块包含256个线程
■ 线程块数量为⌈n/256⌉\lceil n/256 \rceil⌈n/256⌉
■ 使用dim3结构定义网格和块维度
6）CUDA编程流程
●
● 基本流程：
○ 主机端申请显存并拷贝数据到设备端
○ 设备端执行核函数计算
○ 将结果拷贝回主机端并释放显存
● 内存管理函数：
○ cudaMalloc：设备端显存分配
○ cudaMemcpy：主机与设备间数据传输
○ cudaFree：显存释放
○
● cudaMemcpy种类：
○ cudaMemcpyHostToDevice：主机→设备
○ cudaMemcpyDeviceToHost：设备→主机
○ cudaMemcpyDeviceToDevice：设备间拷贝
○ 默认类型（一般不使用）
● 编译组织
○
○ CUDA编译工具：使用NVCC编译器专门处理设备端代码（.cu文件），主机端代码仍用GCC/G++编译（.c/.cpp文件）。
○ 文件组织方式：设备端函数（global/__device__定义）放在.cu文件，主机端代码和头文件放在.h/.c/.cpp文件。
○ 分步编译方法：GPU代码用NVCC生成.o，CPU代码用GCC生成.o，最后链接合并为可执行文件（EXE）。
○ 静态库编译：用NVCC将设备代码编译为静态库（.a/.lib），简化多文件链接过程。
○ 推荐编译策略：采用makefile实现一键式并发编译（CMake/NVCC结合），兼顾效率与工程管理。

二、参考资料

例题1：CPU与GPU相关问题调研
● 流水线加速机制：建议通过计算机体系结构书籍调研流水线机制如何加速指令执行，包括降低延迟和提高效率的具体原理。
● CPU三级缓存特点：需要调研各级缓存（L1/L2/L3）的存储特点及适用场景，了解不同类型数据在不同级别缓存的存放策略。
● GPU架构特点：
○ 控制与计算单元结合：研究GPU中控制单元与计算单元的协同工作机制
○ 线程束概念：深入理解线程束（Warp）在硬件端和软件端的实现原理，明确其作为核函数基本执行单元的原因
GPU并行计算时间分析
● CPU实现特点：
○ 执行时间：10万维向量相加耗时约2×10−42×10^{-4}2×10−4秒
○ 实现方式：使用timeval结构体记录函数调用前后时间差
○ 编译方法：通过g++直接编译生成可执行文件
● GPU实现特点：
○ kernel执行时间：10万维向量相加仅需1.7×10−51.7×10^{-5}1.7×10−5秒，比CPU快约10倍
○ 线程配置：每个线程块256线程，网格大小为⌈n/256⌉\lceil n/256 \rceil⌈n/256⌉
○ 时间瓶颈：显存与内存数据拷贝耗时占比较大（约3×10−43×10^{-4}3×10−4秒）
○ 编译方法：使用nvcc单行命令编译
● 性能优化启示：
○ 计算复杂度影响：简单操作（如向量相加）受限于数据传输，复杂操作（如矩阵相乘）更能体现GPU并行优势
○ 计时策略：应区分kernel执行时间和完整流程时间，后者包含主机-设备数据传输时间
○
● 实验环境：
○ GPU型号：Tesla P40
○ 显存容量：24GB
○ 监控指令：使用nvidia-smi查看GPU状态信息