AMD GPU架构入门指南：从硬件原理到CUDA编程实践

1次阅读

共计 2367 个字符，预计需要花费 6 分钟才能阅读完成。

AMD GPU 架构主要分为面向图形渲染的 RDNA（Radeon DNA）和面向高性能计算的 CDNA（Compute DNA）两类。理解二者的设计差异是进行高效 GPU 编程的基础。

RDNA 架构（如 RX 6000 系列）
采用工作组处理器 (WGP) 作为基本计算单元
每个 WGP 包含 2 个着色器引擎(SE)
专为高时钟频率和低延迟优化
CDNA 架构（如 Instinct MI 系列）
使用计算单元 (CU) 作为基础模块
强化矩阵运算能力（Matrix Cores）
支持更高带宽的 Infinity Fabric 互联

RDNA3 调度示意图：[SE] -> [SA] -> [CU]
  │       │
  ├─ SIMD32 ── Wave32
  └─ SIMD64 ── Wave64

CDNA2 调度示意图：[CU] -> [SIMD16 x 4]
  │
  ├─ Matrix Core
  └─ Scalar Unit

添加 ROCm 仓库

wget -qO - https://repo.radeon.com/rocm/rocm.gpg.key | sudo apt-key add -
echo 'deb [arch=amd64] https://repo.radeon.com/rocm/apt/5.7 jammy main' | sudo tee /etc/apt/sources.list.d/rocm.list

安装基础组件

sudo apt update
sudo apt install rocm-hip-sdk rocblas rccl

常见错误及解决方案：

Error: Required kernel module not loaded

sudo apt install linux-headers-$(uname -r)
sudo dkms install -m amdgpu -v $(modinfo amdgpu | grep version | awk '{print $2}')

PCIe BAR sizing 问题
在 GRUB 配置中添加：

amdgpu.cik_support=1 amdgpu.si_support=1 pci=realloc=off

__global__ void matmul_hip(float *A, float *B, float *C, int M, int N, int K) {
    int row = blockIdx.y * blockDim.y + threadIdx.y;
    int col = blockIdx.x * blockDim.x + threadIdx.x;

    if (row < M && col < N) {
        float sum = 0.0f;
        for (int i = 0; i < K; ++i) {sum += A[row * K + i] * B[i * N + col];
        }
        C[row * N + col] = sum;
    }
}

__global__ void matmul_shared(float *A, float *B, float *C, int M, int N, int K) {__shared__ float sA[BLOCK_SIZE][BLOCK_SIZE];
    __shared__ float sB[BLOCK_SIZE][BLOCK_SIZE];

    // 每个线程加载一个元素到共享内存
    sA[threadIdx.y][threadIdx.x] = A[...];
    sB[threadIdx.y][threadIdx.x] = B[...];
    __syncthreads();

    // Bank conflict 检测：确保同一 bank 不同地址访问
    #pragma unroll
    for (int k = 0; k < BLOCK_SIZE; ++k) {sum += sA[threadIdx.y][k] * sB[k][threadIdx.x];
    }
}

配置 (x,y)	计算吞吐量 (TFLOPS)
(8,8)	12.4
(16,16)	14.7
(32,32)	15.2

#include <hip/hip_fp16.h>

__global__ void wmma_example(half *a, half *b, float *c) {__builtin_amdgcn_wmma_f32_16x16x16_f16_w32(...);
}

避免分支发散：

// 错误写法
if (threadIdx.x % 2) {// 路径 A} else {// 路径 B}

// 正确写法
uint lane = threadIdx.x % warpSize;
value = shuffle(value, lane ^ 0x1);

使用向量化加载：

float4 vec = *reinterpret_cast<float4*>(ptr);

ROCm 版本	内核版本要求	支持的 GPU 型号
5.7	≥5.15	MI200, RX 7900 XTX
5.6	≥5.11	MI100, RX 6900 XT

FROM rocm/dev-ubuntu-22.04:5.7

# 必须挂载的设备文件
VOLUME /dev/kfd /dev/dri

# 权限配置
RUN groupadd -r render && usermod -aG render user

实时监控：

rocm-smi --showuse --showpower --showtemp

性能分析：
```
rocprof --stats ./your_app
```

本文从 AMD GPU 硬件架构分析入手，详细演示了从开发环境搭建到核心算法优化的完整流程。通过 HIP 编程实例展示了与传统 CUDA 开发的异同点，重点强调了 RDNA/CDNA 架构特有的优化方法。建议开发者结合 rocprof 工具进行细粒度性能分析，并根据实际硬件特性调整 workgroup 配置。

正文完