并行计算
并行计算(Parallel Computing)是通过将任务分解为多个子任务、利用多个处理单元同时执行来加速计算的方法。它是高性能计算(HPC)的理论与实践基础,直接支撑了现代科学模拟、大数据处理和深度学习训练等场景。
核心问题包括:如何将任务有效分解(Amdahl 定律的瓶颈)、如何在处理单元间通信与同步(MPI/OpenMP)、如何利用 GPU 的大规模并行能力(CUDA),以及如何将这些技术应用于分布式深度学习训练(数据并行/模型并行/流水线并行)。本笔记从理论基础到编程模型再到实际应用,系统覆盖并行计算的核心知识。
并行计算理论基础
Amdahl's Law(阿姆达尔定律)
Amdahl's Law 描述了在固定问题规模下,增加处理器数量所能获得的理论最大加速比。
公式:
其中:
- \(S(n)\):使用 \(n\) 个处理器时的加速比(Speedup)
- \(P\):程序中可并行化的比例(\(0 \le P \le 1\))
- \(n\):处理器数量
- \((1 - P)\):程序中必须串行执行的比例
关键洞察
当 \(n \to \infty\) 时,\(S = \frac{1}{1 - P}\)。即使拥有无限多的处理器,加速比也受限于串行部分。例如,若 \(P = 0.95\)(95% 可并行),最大加速比仅为 \(\frac{1}{0.05} = 20\) 倍。
不同 P 和 n 下的加速比:
| 可并行比例 P | n=2 | n=4 | n=8 | n=16 | n=64 | n=256 | n=∞ |
|---|---|---|---|---|---|---|---|
| 0.50 | 1.33 | 1.60 | 1.78 | 1.88 | 1.97 | 1.99 | 2.00 |
| 0.75 | 1.60 | 2.29 | 2.91 | 3.37 | 3.76 | 3.94 | 4.00 |
| 0.90 | 1.82 | 3.08 | 4.71 | 6.40 | 8.77 | 9.61 | 10.00 |
| 0.95 | 1.90 | 3.48 | 5.93 | 9.14 | 15.42 | 18.28 | 20.00 |
| 0.99 | 1.98 | 3.88 | 7.48 | 13.91 | 39.26 | 72.12 | 100.00 |
启示
串行部分是并行加速的终极瓶颈。优化并行程序时,首先应关注如何减少串行部分的比例,而不是一味增加处理器数量。
Gustafson's Law(古斯塔夫森定律)
Gustafson's Law 从另一个角度看待并行加速:固定执行时间,随着处理器数量增加,可以解决更大规模的问题。
公式:
其中:
- \(n\):处理器数量
- \(\alpha\):程序中串行部分所占的比例
Amdahl vs Gustafson:视角不同
- Amdahl's Law:固定问题规模,增加处理器 → Strong Scaling(强扩展)
- Gustafson's Law:固定执行时间,增加问题规模 → Weak Scaling(弱扩展)
| 对比维度 | Amdahl's Law | Gustafson's Law |
|---|---|---|
| 固定量 | 问题规模 | 执行时间 |
| 扩展类型 | Strong Scaling | Weak Scaling |
| 假设 | 问题大小不变,加处理器 | 处理器多了,解更大的问题 |
| 结论 | 加速比有上限 | 加速比可随 n 线性增长 |
| 适用场景 | 实时系统、延迟敏感任务 | 科学模拟、大规模数据处理 |
性能指标
- Speedup(加速比):\(S(n) = \frac{T_1}{T_n}\),其中 \(T_1\) 是串行执行时间,\(T_n\) 是 \(n\) 个处理器的并行执行时间
- Efficiency(效率):\(E(n) = \frac{S(n)}{n}\),理想情况下 \(E = 1\)(即 100%),实际中 \(E < 1\)
- Scalability(可扩展性):当处理器数量增加时,程序能否保持良好的效率
| 指标 | 理想值 | 含义 |
|---|---|---|
| Speedup | \(S(n) = n\) | 线性加速,处理器翻倍则速度翻倍 |
| Efficiency | \(E = 1\) | 每个处理器都被充分利用 |
| Scalability | 效率不随 n 下降 | 系统能高效利用更多资源 |
并行编程模型
共享内存模型(Shared Memory)
特点: 多个线程共享同一地址空间,通过读写共享变量进行通信。
┌─────────────────────────────────────┐
│ 共享内存空间 │
│ ┌───┐ ┌───┐ ┌───┐ ┌───┐ │
│ │ A │ │ B │ │ C │ │ D │ ... │
│ └───┘ └───┘ └───┘ └───┘ │
└────┬───────┬───────┬───────┬───────┘
│ │ │ │
Thread0 Thread1 Thread2 Thread3
OpenMP
OpenMP 是一种基于编译器指令的共享内存并行编程接口,通过 #pragma 指令实现并行化,使用简单。
向量加法示例:
#include <stdio.h>
#include <omp.h>
#define N 1000000
int main() {
double a[N], b[N], c[N];
// 初始化
for (int i = 0; i < N; i++) {
a[i] = i * 1.0;
b[i] = i * 2.0;
}
// OpenMP 并行向量加法
#pragma omp parallel for
for (int i = 0; i < N; i++) {
c[i] = a[i] + b[i];
}
printf("c[0] = %f, c[N-1] = %f\n", c[0], c[N-1]);
return 0;
}
编译方式
gcc -fopenmp vector_add.c -o vector_add
常用 OpenMP 指令:
| 指令 | 功能 |
|---|---|
#pragma omp parallel |
创建并行区域 |
#pragma omp parallel for |
并行化 for 循环 |
#pragma omp critical |
临界区,同一时刻只允许一个线程执行 |
#pragma omp atomic |
原子操作 |
#pragma omp barrier |
同步屏障 |
#pragma omp reduction(+:sum) |
归约操作 |
优缺点:
| 优点 | 缺点 |
|---|---|
| 编程简单,增量式并行化 | 仅限单机(单节点)内 |
| 编译器指令方式,改动小 | 需注意数据竞争(Data Race) |
| 支持 C/C++/Fortran | 可扩展性受限于单机核心数 |
分布式内存模型(Distributed Memory)
特点: 各节点拥有独立内存,节点之间通过消息传递进行通信。
┌──────────┐ 消息传递 ┌──────────┐
│ Node 0 │◄────────────►│ Node 1 │
│ ┌──────┐ │ (网络) │ ┌──────┐ │
│ │ 内存 │ │ │ │ 内存 │ │
│ └──────┘ │ │ └──────┘ │
└──────────┘ └──────────┘
▲ ▲
│ 消息传递 │
▼ ▼
┌──────────┐ ┌──────────┐
│ Node 2 │◄────────────►│ Node 3 │
│ ┌──────┐ │ │ ┌──────┐ │
│ │ 内存 │ │ │ │ 内存 │ │
│ └──────┘ │ │ └──────┘ │
└──────────┘ └──────────┘
MPI(Message Passing Interface)
MPI 是分布式内存并行编程的标准接口,适用于跨节点的大规模并行计算。
基本操作:
| 函数 | 功能 |
|---|---|
MPI_Init |
初始化 MPI 环境 |
MPI_Comm_rank |
获取当前进程编号 |
MPI_Comm_size |
获取总进程数 |
MPI_Send |
发送消息(阻塞) |
MPI_Recv |
接收消息(阻塞) |
MPI_Bcast |
广播(一对多) |
MPI_Reduce |
归约(多对一) |
MPI_Allreduce |
全归约(多对多) |
MPI_Finalize |
结束 MPI 环境 |
MPI 向量加法示例:
#include <stdio.h>
#include <stdlib.h>
#include <mpi.h>
#define N 1000000
int main(int argc, char *argv[]) {
int rank, size;
MPI_Init(&argc, &argv);
MPI_Comm_rank(MPI_COMM_WORLD, &rank);
MPI_Comm_size(MPI_COMM_WORLD, &size);
int local_n = N / size;
double *local_a = malloc(local_n * sizeof(double));
double *local_b = malloc(local_n * sizeof(double));
double *local_c = malloc(local_n * sizeof(double));
// 每个进程初始化自己的数据块
for (int i = 0; i < local_n; i++) {
local_a[i] = (rank * local_n + i) * 1.0;
local_b[i] = (rank * local_n + i) * 2.0;
}
// 每个进程计算自己的部分
for (int i = 0; i < local_n; i++) {
local_c[i] = local_a[i] + local_b[i];
}
if (rank == 0) {
printf("Process 0: local_c[0] = %f\n", local_c[0]);
}
free(local_a); free(local_b); free(local_c);
MPI_Finalize();
return 0;
}
编译与运行
mpicc vector_add_mpi.c -o vector_add_mpi
mpirun -np 4 ./vector_add_mpi
优缺点:
| 优点 | 缺点 |
|---|---|
| 可跨节点,扩展性强 | 编程复杂,需显式管理通信 |
| 适合大规模集群 | 通信开销可能成为瓶颈 |
| 无数据竞争问题 | 调试困难 |
GPU 并行(CUDA)
CUDA 编程模型
CUDA(Compute Unified Device Architecture)是 NVIDIA 提出的 GPU 通用计算编程模型。
层次结构:
Grid(网格)
├── Block (0,0) Block (1,0) Block (2,0)
│ ├── Thread (0,0) ├── Thread (0,0) ├── ...
│ ├── Thread (1,0) ├── Thread (1,0) │
│ ├── Thread (0,1) ├── Thread (0,1) │
│ └── ... └── ... └── ...
├── Block (0,1) Block (1,1) Block (2,1)
│ └── ... └── ... └── ...
- Grid(网格):由多个 Block 组成,对应一次 Kernel 调用
- Block(线程块):由多个 Thread 组成,Block 内线程可协作
- Thread(线程):最小执行单元
核函数(Kernel)
Kernel 是在 GPU 上执行的函数,使用 __global__ 修饰,通过 <<<grid, block>>> 语法启动。
GPU 内存层次
| 内存类型 | 速度 | 大小 | 作用域 | 生命周期 |
|---|---|---|---|---|
| Registers(寄存器) | 最快 | 极小 | 单个 Thread | Thread |
| Shared Memory(共享内存) | 很快 | 较小(~48KB/Block) | Block 内所有 Thread | Block |
| Global Memory(全局内存) | 较慢 | 较大(GB 级) | 所有 Thread | 由主机管理 |
| Constant Memory(常量内存) | 快(有缓存) | 64KB | 所有 Thread(只读) | 由主机管理 |
CUDA 向量加法示例
#include <stdio.h>
#include <cuda_runtime.h>
#define N 1000000
// 核函数:每个线程计算一个元素
__global__ void vector_add(float *a, float *b, float *c, int n) {
int idx = blockIdx.x * blockDim.x + threadIdx.x;
if (idx < n) {
c[idx] = a[idx] + b[idx];
}
}
int main() {
float *h_a, *h_b, *h_c; // 主机(CPU)内存
float *d_a, *d_b, *d_c; // 设备(GPU)内存
size_t bytes = N * sizeof(float);
// 分配主机内存并初始化
h_a = (float*)malloc(bytes);
h_b = (float*)malloc(bytes);
h_c = (float*)malloc(bytes);
for (int i = 0; i < N; i++) {
h_a[i] = i * 1.0f;
h_b[i] = i * 2.0f;
}
// 分配设备内存
cudaMalloc(&d_a, bytes);
cudaMalloc(&d_b, bytes);
cudaMalloc(&d_c, bytes);
// 数据从主机拷贝到设备
cudaMemcpy(d_a, h_a, bytes, cudaMemcpyHostToDevice);
cudaMemcpy(d_b, h_b, bytes, cudaMemcpyHostToDevice);
// 启动核函数
int blockSize = 256;
int gridSize = (N + blockSize - 1) / blockSize;
vector_add<<<gridSize, blockSize>>>(d_a, d_b, d_c, N);
// 结果从设备拷贝回主机
cudaMemcpy(h_c, d_c, bytes, cudaMemcpyDeviceToHost);
printf("c[0] = %f, c[N-1] = %f\n", h_c[0], h_c[N-1]);
// 释放内存
cudaFree(d_a); cudaFree(d_b); cudaFree(d_c);
free(h_a); free(h_b); free(h_c);
return 0;
}
编译方式
nvcc vector_add.cu -o vector_add
GPU vs CPU 适用场景对比
| 对比维度 | CPU | GPU |
|---|---|---|
| 核心数 | 少(几个到几十个) | 多(数千个) |
| 单核性能 | 强(复杂逻辑) | 弱(简单运算) |
| 并行粒度 | 粗粒度 | 细粒度(SIMT) |
| 适合任务 | 逻辑控制、分支密集 | 数据并行、计算密集 |
| 典型应用 | OS 调度、编译、数据库 | 矩阵运算、图像处理、深度学习 |
| 内存带宽 | 较低 | 很高(HBM) |
| 延迟 | 低(单任务快) | 高(启动开销大) |
选择建议
- 任务有大量独立、相同的计算 → 用 GPU
- 任务有复杂分支逻辑和依赖 → 用 CPU
- 最佳实践是 CPU + GPU 异构协同
并行算法设计
常见并行模式
Master-Worker(主从模式)
一个 Master 进程负责任务分配与结果收集,多个 Worker 进程负责实际计算。
┌──────────┐
│ Master │
│ 分配任务 │
└──┬──┬──┬─┘
│ │ │
┌──────┘ │ └──────┐
▼ ▼ ▼
┌────────┐┌────────┐┌────────┐
│Worker 0││Worker 1││Worker 2│
│ 计算 ││ 计算 ││ 计算 │
└───┬────┘└───┬────┘└───┬────┘
│ │ │
└────┐ │ ┌────┘
▼ ▼ ▼
┌──────────┐
│ Master │
│ 收集结果 │
└──────────┘
MapReduce
受函数式编程启发,将计算分为 Map(映射) 和 Reduce(归约) 两个阶段。是大数据处理(如 Hadoop、Spark)的核心模型。
输入数据 → [Map] → 中间键值对 → [Shuffle] → [Reduce] → 输出结果
Pipeline(流水线)
数据依次流过多个处理阶段,每个阶段由不同处理器执行,类似工厂流水线。
数据流 → [Stage 1] → [Stage 2] → [Stage 3] → [Stage 4] → 输出
处理器0 处理器1 处理器2 处理器3
Divide and Conquer(分治)
将问题递归地拆分为子问题,各子问题并行求解,最后合并结果。典型示例:并行归并排序、并行快速排序。
通信模式
Point-to-Point(点对点)
两个进程之间直接发送/接收消息。
Process 0 ──── data ────► Process 1
Collective Communication(集合通信)
涉及一组进程的通信操作,是并行计算中最常用的通信模式。
Broadcast(广播):一个进程将数据发送给所有进程。
P0 [A] ──► P0 [A]
P1 [A]
P2 [A]
P3 [A]
Scatter(散发):一个进程将数据分块发送给各进程。
P0 [A B C D] ──► P0 [A]
P1 [B]
P2 [C]
P3 [D]
Gather(收集):各进程将数据汇集到一个进程。
P0 [A] ──► P0 [A B C D]
P1 [B]
P2 [C]
P3 [D]
Reduce(归约):对各进程的数据执行某种运算(如求和),结果存于一个进程。
P0 [1] ──► P0 [1+2+3+4 = 10]
P1 [2] (op = SUM)
P2 [3]
P3 [4]
AllReduce(全归约):Reduce 后将结果广播给所有进程。等价于 Reduce + Broadcast。
P0 [1] ──► P0 [10]
P1 [2] P1 [10]
P2 [3] P2 [10]
P3 [4] P3 [10]
AllReduce 的重要性
AllReduce 是分布式深度学习训练中最核心的通信操作,用于同步各 GPU 上的梯度。常见实现包括 Ring AllReduce 和 Tree AllReduce。
同步与负载均衡
Barrier 同步
所有线程/进程到达 Barrier 后才能继续执行,确保各并行分支在某一时刻对齐。
Thread 0: ──work── │ ──work──
Thread 1: ──work────│ ──work──
Thread 2: ──work── │ ──work──
Barrier
锁与原子操作
- 互斥锁(Mutex):保护临界区,同一时刻只允许一个线程访问共享资源
- 自旋锁(Spinlock):线程忙等待直到获得锁,适用于锁持有时间短的场景
- 原子操作(Atomic Operation):硬件级别的不可分割操作,如
atomic_add、compare_and_swap
注意事项
- 死锁(Deadlock):两个或多个线程互相等待对方释放锁
- 数据竞争(Data Race):多线程同时读写同一变量,导致结果不确定
- 过度同步会严重降低并行性能,应尽量减少锁的粒度和持有时间
负载均衡
| 策略 | 静态负载均衡 | 动态负载均衡 |
|---|---|---|
| 分配时机 | 程序运行前预先分配 | 运行时根据负载动态调整 |
| 实现复杂度 | 简单 | 较复杂 |
| 适用场景 | 任务计算量均匀、可预测 | 任务计算量不均匀、不可预测 |
| 开销 | 低 | 有调度开销 |
| 典型方法 | 均匀分块、循环分配 | Work Stealing、任务队列 |
负载不均衡的影响
如果各处理器的任务量差异很大,快的处理器必须等待慢的处理器完成,导致整体执行时间由最慢的处理器决定(木桶效应)。这会显著降低并行效率。
并行计算在深度学习中的应用
现代深度学习模型(如 LLM)规模巨大,单个 GPU 无法独立完成训练,分布式并行训练已成为必备技术。
数据并行(Data Parallelism)
每个 GPU 持有完整的模型副本,各自处理不同的数据批次(mini-batch),然后通过 AllReduce 同步梯度。
GPU 0: Model Copy + Data Batch 0 → Gradient 0 ─┐
GPU 1: Model Copy + Data Batch 1 → Gradient 1 ─┤ AllReduce → 平均梯度 → 更新模型
GPU 2: Model Copy + Data Batch 2 → Gradient 2 ─┤
GPU 3: Model Copy + Data Batch 3 → Gradient 3 ─┘
- 优点:实现简单,通信模式固定
- 缺点:每个 GPU 必须能放下完整模型
模型并行(Model Parallelism)
将模型的不同层或不同部分分配到不同 GPU 上。适用于单个 GPU 无法容纳整个模型的情况。
GPU 0: Layer 1-10
GPU 1: Layer 11-20
GPU 2: Layer 21-30
GPU 3: Layer 31-40
- 优点:可训练超大模型
- 缺点:GPU 之间有数据依赖,容易出现 GPU 空闲(气泡问题)
流水线并行(Pipeline Parallelism)
将模型按层分段后,将不同的 micro-batch 以流水线方式送入各段,减少 GPU 空闲时间。
时间 → t0 t1 t2 t3 t4 t5
GPU 0: [MB0] [MB1] [MB2] [MB3]
GPU 1: [MB0] [MB1] [MB2] [MB3]
GPU 2: [MB0] [MB1] [MB2] [MB3]
GPU 3: [MB0] [MB1] [MB2] [MB3]
气泡问题
流水线启动和结束阶段存在 GPU 空闲的"气泡"(Bubble),增加 micro-batch 数量可以减少气泡比例。GPipe 和 PipeDream 是常见的流水线并行框架。
混合并行策略
实际大规模训练中通常组合使用多种并行策略:
| 并行策略 | 切分维度 | 典型框架 |
|---|---|---|
| 数据并行 | 数据(Batch) | PyTorch DDP, Horovod |
| 模型并行(张量并行) | 模型层内(权重矩阵) | Megatron-LM |
| 流水线并行 | 模型层间 | GPipe, PipeDream |
| 混合并行(3D 并行) | 同时切分数据 + 层内 + 层间 | Megatron-LM + DeepSpeed |
常见框架
- PyTorch DDP(DistributedDataParallel):数据并行的标准方案
- DeepSpeed(Microsoft):支持 ZeRO 优化,大幅降低显存占用
- Megatron-LM(NVIDIA):专为超大 Transformer 模型设计的张量并行方案
- FSDP(Fully Sharded Data Parallel):PyTorch 原生的模型分片方案
AllReduce 在分布式训练中的作用
在数据并行训练中,每轮迭代的关键步骤:
- 各 GPU 用各自的数据前向传播,计算 Loss
- 各 GPU 反向传播,得到本地梯度
- 通过 AllReduce 将所有 GPU 的梯度求平均
- 各 GPU 用平均梯度更新模型参数
Ring AllReduce 是最常用的实现方式:将 \(n\) 个 GPU 排列成环形,通过 \(2(n-1)\) 步通信完成全归约,通信量与 GPU 数量无关(仅与数据大小相关),可扩展性好。
GPU 0
╱ ╲
GPU 3 GPU 1
╲ ╱
GPU 2
每个GPU只与相邻GPU通信,经过2(n-1)步完成AllReduce
优化方向
- 梯度压缩:减少通信数据量
- 计算与通信重叠:反向传播计算梯度的同时,已计算完的层可以先开始 AllReduce
- 混合精度训练(FP16/BF16):减少显存占用和通信量