AMD GPU編程——基於ROCm/HIP的架構、工具鏈和開發流程
過敏意、孫世軒、陳鑫宇、孫軼凡
商品描述
本書系統介紹了基於 AMD ROCm 平臺的 HIP 加速計算技術，圍繞 GPU 編程在異構計算中的核心作用，展開對 HIP 編程語言、GPU 硬件架構、性能優化方法、多 GPU 協同計算策略及 ROCm 生態組件的全面講解。內容設計結合了當前實際需求，從語言基礎出發，逐步引導讀者掌握復雜程序結構與高效的並行計算模式。 書中配有大量可運行的代碼示例，涵蓋矩陣運算、圖像處理、聚類分析、圖遍歷等典型任務，突出實踐導向，強調理論與應用的結合。除了基礎語法與程序結構，還深入講解了 HIP 中的內存管理、線程調度、核函數優化、異步執行機制、調試工具鏈等關鍵主題，並結合 ROCm 提供的多種高性能庫（如 rocBLAS、rocFFT、rocRAND）及性能分析工具，構建起完整的開發與優化路徑。 針對當前多樣化的計算平臺需求，書中也探討了多 GPU 系統下的點對點通信與集合通信方法，介紹了 ROCm 與機器學習框架（PyTorch、TensorFlow）的集成方式，以及在數據中心環境中結合容器與作業調度系統的實際部署方案。 本書適合高校計算機、軟件工程、電子信息等專業本科生和研究生使用，也可供從事高性能計算、GPU 應用開發與平臺遷移的科研人員和工程師參考。
作者簡介
過敏意，現任上海交通大學講席教授，長期從事並行計算與編譯器、雲計算、高性能計算、高能效計算等方向的研究。曾主持國家自然科學基金傑青項目、973 計劃項目、863 計劃等項目。他是中組部國家級人才計劃獲得者，歐洲科學院外籍院士，國務院政府特殊津貼專家，IEEE Fellow，中國計算機學會會士、常務理事，擔任教育部創新團隊學術帶頭人，973 計劃首席科學家。
目錄大綱
目 錄
第 1章 介紹 ………………………………………………………………………
1.1 並行編程 ……………………………………………………………………………… 2
1.2 GPU …………………………………………………………………………………… 3
1.3 ROCm ………………………………………………………………………………… 5
1.4 HIP框架 ……………………………………………………………………………… 6
1.5 本書內容 ……………………………………………………………………………… 6
第 2章 開始使用 HIP編程 ………………………………………………………
2.1 在 HIP中實現“Hello World” ……………………………………………………… 8
2.2 使用 HIP進行數據處理——Vector Add 示例 ……………………………………… 9
2.2.1 並行執行的機會……………………………………………………………… 9
2.2.2 組織線程…………………………………………………………………… 10
2.2.3 使用 HIP API 進行數據傳輸 ……………………………………………… 12
2.2.4 錯誤和正確性檢查………………………………………………………… 13
2.2.5 綜合示例…………………………………………………………………… 14
2.3 總結 ………………………………………………………………………………… 16
第 3章 HIP內核編程………………………………………………………………
3.1 在 HIP內核中調用函數 …………………………………………………………… 18
3.1.1 HIP中的 __global__ 函數 ……………………………………………… 18
3.1.2 HIP中的 __device_ _ 函數 ……………………………………………… 19
3.1.3 HIP中的 __host__ 函數 ………………………………………………… 20
3.1.4 結合 __host__ 和 _ _device__ 函數 ……………………………………… 20
3.2 在 HIP內核中使用模板 …………………………………………………………… 21
3.3 在 HIP中使用結構體 ……………………………………………………………… 22
3.4 總結 ………………………………………………………………………………… 26
第 4章 HIP運行時 API ……………………………………………………………
4.1 HIP內存管理 ……………………………………………………………………… 27
4.1.1 固定內存…………………………………………………………………… 27
4.1.2 統一內存…………………………………………………………………… 29
4.2 HIP流 ……………………………………………………………………………… 30
4.2.1 流的基礎知識……………………………………………………………… 31
目 錄
4.2.2 基於流的關鍵 API ………………………………………………………… 31
4.2.3 默認流與非默認流………………………………………………………… 32
4.2.4 並發內核…………………………………………………………………… 33
4.2.5 重疊計算與通信…………………………………………………………… 33
4.3 HIP事件 …………………………………………………………………………… 35
4.3.1 創建 HIP事件 …………………………………………………………… 35
4.3.2 記錄 HIP事件 …………………………………………………………… 35
4.3.3 使用 HIP事件計算耗時 ………………………………………………… 36
4.3.4 使用 HIP事件協調操作 ………………………………………………… 36
4.3.5 使用 HIP事件釋放內存 ………………………………………………… 36
4.3.6 創建具有特定標誌的事件………………………………………………… 37
4.3.7 確保兩種處理單元之間的一致性………………………………………… 37
4.3.8 使用 HIP事件的向量加法 ……………………………………………… 37
4.4 總結 ………………………………………………………………………………… 39
第 5章 GPU編程模式 ……………………………………………………………
5.1 二維內核 …………………………………………………………………………… 40
5.2 模板計算 …………………………………………………………………………… 44
5.3 多內核示例——廣度優先搜索 …………………………………………………… 47
5.4 CPU-GPU 協同計算——KMeans ………………………………………………… 49
5.5 原子操作——直方圖 ……………………………………………………………… 52
5.6 總結 ………………………………………………………………………………… 54
第 6章 AMD GPU內部原理………………………………………………………
6.1 AMD GPU ………………………………………………………………………… 55
6.2 整體架構 …………………………………………………………………………… 57
6.3 命令處理器和 DMA引擎 ………………………………………………………… 58
6.4 工作組分發 ………………………………………………………………………… 58
6.5 指令調度器 ………………………………………………………………………… 60
6.6 SIMD 單元 ………………………………………………………………………… 61
6.7 線程分歧 …………………………………………………………………………… 62
6.8 內存合並 …………………………………………………………………………… 64
目 錄
6.9 內存層次結構 ……………………………………………………………………… 65
6.10 AMD RDNA GPUs ……………………………………………………………… 66
6.11 總結 ……………………………………………………………………………… 67
第 7章 HIP工具……………………………………………………………………
7.1 ROCmInfo ………………………………………………………………………… 68
7.2 ROCm SMI ………………………………………………………………………… 71
7.3 ROCm 調試器 ……………………………………………………………………… 73
7.4 ROCm 分析器 ……………………………………………………………………… 76
7.4.1 ROCTracer ………………………………………………………………… 76
7.4.2 ROCProfiler ……………………………………………………………… 78
7.5 ROCm Profiler V2 ………………………………………………………………… 80
7.5.1 應用程序追蹤……………………………………………………………… 80
7.5.2 內核分析…………………………………………………………………… 81
7.5.3 ROCSys …………………………………………………………………… 84
7.6 使用 Hipify 將 CUDA程序移植到 HIP…………………………………………… 86
7.6.1 Hipify 工具 ………………………………………………………………… 86
7.6.2 一般 Hipify 指南 ………………………………………………………… 90
7.6.3 矩陣轉置的 Hipify ………………………………………………………… 91
7.6.4 常見問題及解決方案……………………………………………………… 93
7.7 總結 ………………………………………………………………………………… 94
第 8章 HIP性能優化………………………………………………………………
8.1 高度並行的工作負載——圖像伽馬校正 ………………………………………… 95
8.2 固定大小的內核——圖像伽馬校正 ……………………………………………… 98
8.3 歸約——數組求和 ………………………………………………………………… 100
8.4 分塊與復用——矩陣乘法 ………………………………………………………… 103
8.5 分塊與合並——矩陣轉置 ………………………………………………………… 107
8.6 總結 ………………………………………………………………………………… 109
第 9章 ROCm庫 ………………………………………………………………
9.1 rocBLAS …………………………………………………………………………… 110
9.1.1 使用 rocBLAS …………………………………………………………… 111
目 錄
9.1.2 rocBLAS 函數 …………………………………………………………… 113
9.1.3 異步執行…………………………………………………………………… 114
9.1.4 MI100 上的 rocBLAS …………………………………………………… 115
9.1.5 從傳統 BLAS 庫遷移 …………………………………………………… 115
9.2 rocSPARSE ………………………………………………………………………… 116
9.2.1 稀疏數據表示……………………………………………………………… 116
9.2.2 rocSPARSE 函數 ………………………………………………………… 118
9.3 rocFFT ……………………………………………………………………………… 120
9.3.1 rocFFT工作流程 ………………………………………………………… 121
9.3.2 FFT執行計劃……………………………………………………………… 122
9.4 rocRAND …………………………………………………………………………… 123
9.5 總結 ………………………………………………………………………………… 126
第 10章 多 GPU編程 …………………………………………………………
10.1 HIP設備 API ……………………………………………………………………… 128
10.2 基於流的多 GPU 編程 …………………………………………………………… 130
10.3 基於線程的多 GPU 編程 ………………………………………………………… 132
10.4 基於 MPI 的多 GPU 編程………………………………………………………… 134
10.5 GPU—GPU 通信 ………………………………………………………………… 137
10.6 RCCL ……………………………………………………………………………… 141
10.6.1 廣播 ……………………………………………………………………… 141
10.6.2 AllReduce ………………………………………………………………… 143
10.7 總結 ……………………………………………………………………………… 145
第 11章 使用 ROCm進行機器學習 …………………………………………
11.1 ROCm上的 PyTorch ……………………………………………………………… 146
11.1.1 安裝 PyTorch …………………………………………………………… 147
11.1.2 測試 PyTorch安裝 ……………………………………………………… 147
11.1.3 使用 Inception v3模型進行圖像分類…………………………………… 148
11.2 ROCm上的 TensorFlow ………………………………………………………… 150
11.2.1 安裝 TensorFlow ………………………………………………………… 150
11.2.2 測試 TensorFlow安裝 …………………………………………………… 151
目 錄
11.2.3 使用 TensorFlow進行訓練 ……………………………………………… 151
11.3 總結 ……………………………………………………………………………… 152
第 12章 數據中心中的 ROCm…………………………………………………
12.1 容器化的 ROCm ………………………………………………………………… 153
12.2 使用 Kubernetes管理 ROCm容器 ……………………………………………… 154
12.3 使用 SLURM管理 ROCm節點 ………………………………………………… 156
12.3.1 SLURM交互模式 ……………………………………………………… 157
12.3.2 SLURM批處理提交模式 ……………………………………………… 157
12.4 總結 ……………………………………………………………………………… 158
第 13章 第三方工具 ……………………………………………………………
13.1 PAPI ……………………………………………………………………………… 159
13.1.1 PAPI工具和測試 ………………………………………………………… 160
13.1.2 PAPI對 AMD GPU的支持 ……………………………………………… 163
13.1.3 預設事件與計數器分析工具包 ………………………………………… 164
13.2 Score-P和 Vampir ………………………………………………………………… 165
13.2.1 使用 Score-P進行跟蹤 ………………………………………………… 166
13.2.2 Score-P的使用 …………………………………………………………… 166
13.2.3 對 Quicksilver應用程序的性能分析 …………………………………… 167
13.3 Trace Compass和 Theia ………………………………………………………… 169
13.4 TAU ……………………………………………………………………………… 174
13.4.1 使用 TAU分析 HIP程序性能 ………………………………………… 174
13.4.2 使用 TAU跟蹤 HIP程序 ……………………………………………… 175
13.4.3 使用 APEX測量 HIP程序 ……………………………………………… 178
13.4.4 TAU總結 ………………………………………………………………… 180
13.5 TotalView調試器 ………………………………………………………………… 180 13.6 HPCToolkit………………………………………………………………………… 184
13.6.1 HPCToolkit的工作流程 ………………………………………………… 185
13.6.2 使用 HPCToolkit分析 PIConGPU ……………………………………… 186
13.6.3 收集和分析性能數據和跟蹤文件 ……………………………………… 186
13.6.4 使用硬件計數器進行測量 ……………………………………………… 188
目 錄
13.7 使用 Linaro Forge進行調試和性能分析………………………………………… 190
13.7.1 Linaro DDT ……………………………………………………………… 190
13.7.2 Linaro MAP ……………………………………………………………… 191
13.7.3 Linaro性能報告 ………………………………………………………… 191
13.7.4 使用 Linaro DDT進行 GPU調試 ……………………………………… 192
13.7.5 使用 Linaro MAP分析 GPU性能 ……………………………………… 195
13.7.6 GPU性能報告 …………………………………………………………… 197 13.8 E4S——極端規模科學軟件棧 …………………………………………………… 198
第 14章 HIP編程實戰：大模型自註意力機制與 FlashAttention …………
14.1 自註意力機制簡介 ……………………………………………………………… 200
14.2 FlashAttention核心思想 ………………………………………………………… 201
14.3 HIP實現 FlashAttention ………………………………………………………… 202
附錄 A ROCm安裝 ……………………………………………………………
A.1 先決條件…………………………………………………………………………… 216
A.2 理解 ROCm 軟件包 ……………………………………………………………… 217
A.3 安裝………………………………………………………………………………… 218
A.3.1 安裝腳本方法 …………………………………………………………… 218
A.3.2 包管理器方法 …………………………………………………………… 219
A.3.3 驗證安裝過程 …………………………………………………………… 220
A.4 升級 ROCm………………………………………………………………………… 220
A.5 卸載 ROCm………………………………………………………………………… 221
附錄 B CDNA匯編………………………………………………………………
B.1 使用 CDNA匯編代碼 …………………………………………………………… 222
B.1.1 檢索 HIP內核二進制文件 ……………………………………………… 222
B.1.2 反匯編 CDNA二進制文件 ……………………………………………… 223
B.2 CDNA寄存器 ……………………………………………………………………… 223
B.3 指令類型 …………………………………………………………………………… 224
B.4 內存訪問指令 ……………………………………………………………………… 224
B.5 示例：移位復制 …………………………………………………………………… 225
B.6 示例：分支 ………………………………………………………………………… 226
目 錄
B.7 比較 CDNA2 和 CDNA3 ………………………………………………………… 228
B.8 總結 ………………………………………………………………………………… 229
附錄 C OmniTools ………………………………………………………………
C.1 Omnitrace ………………………………………………………………………… 230
C.1.1 Omnitrace 配置文件 ……………………………………………………… 230
C.1.2 收集跟蹤數據 …………………………………………………………… 231
C.1.3 輸出和可視化 …………………………………………………………… 232
C.2 Omniperf …………………………………………………………………………… 233
C.2.1 使用 Omniperf 分析程序性能 …………………………………………… 233
C.2.2 使用 CLI 進行分析 ……………………………………………………… 234
C.2.3 使用基於 Web 的 GUI 進行分析 ………………………………………… 236
C.2.4 使用 Grafana 進行分析 …………………………………………………… 236
C.3 總結 ………………………………………………………………………………… 238
參考文獻……………………………………………………………………………
