HIP（Heterogeneous-compute Interface for Portability）

3 秒看懂

HIP 是 AMD 為對抗 NVIDIA CUDA 生態鎖定而打造的 GPU 程式設計介面，語法幾乎與 CUDA 一一對應，讓已有 CUDA 程式碼能以最小改動移植到 AMD GPU 上執行。 它是 AMD ROCm 軟體棧的核心基石——沒有 HIP，AMD 的 AI 加速卡就是一堆不會說話的矽。

3 分鐘產業解釋

為什麼需要 HIP？

當前 AI 訓練/推論算力市場被 NVIDIA 佔據約 80%+ 份額（行業估算，口徑各異），其中最深的護城河不是硬體規格，而是 CUDA 軟體生態。全球幾乎所有深度學習架構（PyTorch、TensorFlow、JAX 等）的核心運算元庫、分散式通訊庫、profiling 工具都優先甚至僅支援 CUDA。這造成了極強的 vendor lock-in。

AMD 擁有 MI200/MI300 系列 GPU，在 HBM 容量和部分理論算力指標上具備競爭力，但如果開發者需要從頭重寫程式碼才能在 AMD 硬體上跑，遷移成本高到足以抵消硬體價效比優勢。

HIP 的核心戰略意義：大幅降低 CUDA 程式碼向 AMD GPU 遷移的摩擦成本。

產業定位

┌─────────────────────────────────────────────────┐
│              使用者態 AI 架構                        │
│   PyTorch / TensorFlow / JAX / vLLM / ...       │
├─────────────────────────────────────────────────┤
│         AI 運算元庫 & 分散式通訊                      │
│   MIOpen / RCCL / FlashAttention / ...           │
├─────────────────────────────────────────────────┤
│   ★ HIP Runtime & API（本次主角）                  │
├─────────────────────────────────────────────────┤
│   ROCm 核心驅動 / KFD（Linux Kernel）              │
├─────────────────────────────────────────────────┤
│   AMD GPU 硬體（CDNA 架構：MI300X 等）             │
└─────────────────────────────────────────────────┘

15 分鐘專家深入

HIP 的設計哲學

HIP 的定位不是”另一門新語言”，而是一層 薄抽象層。其設計目標可以歸納為三點：

語法近似 CUDA：HIP 的 API 命名、記憶體模型、Kernel 啟動語法、流/事件機制與 CUDA 保持高度對應關係（開發者社群通常描述為 ~90% 以上 API 可一一對映）。
雙向編譯：同一份 HIP 原始碼可以通過 HIP/ROCm 編譯路徑編譯到 AMD GPU，也可以通過 NVIDIA CUDA 編譯路徑編譯到 NVIDIA GPU（條件是使用 CUDA 相容子集）。
開源開放：HIP 及 ROCm 整體棧以開源方式釋出在 GitHub，降低了企業對閉源驅動的依賴顧慮。

核心元件

元件	功能
hipcc	編譯器驅動指令碼，後端根據目標平台排程到 clang（AMD 路徑）或 nvcc（NVIDIA 路徑）
HIP Runtime	提供裝置管理、記憶體分配（hipMalloc 等）、Kernel 啟動、流/事件同步等執行時 API
HIPIFY 工具鏈	包含 HIPIFY-clang（基於 LLVM AST 的原始碼翻譯器）和 HIPIFY-perl（基於正則的快速轉換指令碼），自動將 `.cu` 檔案轉換為 `.hip` 檔案
hipBLAS / hipFFT / hipSOLVER 等	HIP 級別的數學庫封裝，底層可排程到 ROCm 原生實現或 cuBLAS 等

Kernel 啟動語法對照

這是理解 HIP 最直觀的方式：

// CUDA 寫法
myKernelgridDim, blockDim, sharedMemBytes, stream(args);

// HIP 寫法 —— 幾乎相同
hipLaunchKernelGGL(myKernel, gridDim, blockDim, sharedMemBytes, stream, args);
// 或使用宏等價形式，依版本而定

編譯流程

             .hip / .cpp 原始檔
                    │
                    ▼
               ┌─────────┐
               │  hipcc   │  (編譯器驅動指令碼)
               └────┬────┘
                    │
          ┌─────────┴──────────┐
          ▼                    ▼
   AMD 目標路徑            NVIDIA 目標路徑
   clang + AMDGPU         nvcc (或 clang +
   LLVM 後端               CUDA 後端)
          │                    │
          ▼                    ▼
   .hsaco (AMD GPU         .cubin / .ptx
    機器碼物件)             (NVIDIA GPU 物件)
          │                    │
          ▼                    ▼
   HIP Runtime 載入      CUDA Runtime 載入
   → AMD GPU 執行        → NVIDIA GPU 執行

技術原理（深入機制層）

記憶體模型

HIP 採用與 CUDA 一致的 統一記憶體層級 概念：

┌──────────────────────────────────────────┐
│               Host (CPU)                 │
│  主機記憶體：常規 malloc / new              │
├──────────────┬───────────────────────────┤
│              │     PCIe / CXL / xGMI     │
│              │     (資料搬運通道)          │
├──────────────┴───────────────────────────┤
│              Device (GPU)                │
│  ┌─────────────────────────────────┐     │
│  │  Global Memory (HBM / GDDR)     │     │
│  │  hipMalloc() 分配               │     │
│  ├─────────────────────────────────┤     │
│  │  Shared Memory (每個 CU/Block)  │     │
│  │  __shared__ 宣告，片上 SRAM      │     │
│  ├─────────────────────────────────┤     │
│  │  Registers (每執行緒私有)          │     │
│  ├─────────────────────────────────┤     │
│  │  Constant / Texture Memory      │     │
│  │  (只讀最佳化路徑)                  │     │
│  └─────────────────────────────────┘     │
└──────────────────────────────────────────┘

關鍵 API 對應關係：

功能	CUDA API	HIP API
裝置記憶體分配	`cudaMalloc`	`hipMalloc`
主機→裝置複製	`cudaMemcpy`	`hipMemcpy`
同步等待	`cudaDeviceSynchronize`	`hipDeviceSynchronize`
流建立	`cudaStreamCreate`	`hipStreamCreate`
事件記錄	`cudaEventRecord`	`hipEventRecord`
裝置查詢	`cudaGetDeviceProperties`	`hipGetDeviceProperties`

Kernel 維度與硬體對映

HIP 中 Kernel 的執行緒組織（grid → block → thread）與 CUDA 完全一致：

Grid（網格）
 ├── Block(0,0)    Block(1,0)    ...  Block(N,0)
 ├── Block(0,1)    Block(1,1)    ...  Block(N,1)
 │    ...
 └── Block(0,M)    ...

每個 Block 包含若干 Thread
 → 對映到 AMD GPU 上的 CU（Compute Unit）
 → 對映到 NVIDIA GPU 上的 SM（Streaming Multiprocessor）

AMD CDNA 架構（如 MI300 系列）上的 CU 內部包含：

向量 ALU（VALU）：執行 32/64 寬 SIMD 運算
矩陣核心（Matrix Core / MFMA）：執行矩陣融合乘加，類似 NVIDIA Tensor Core
標量單元（SALU）：處理控制流和地址計算
共享記憶體/LDS（Local Data Share）：Block 內執行緒共享的片上儲存
暫存器檔案：每執行緒私有

⚠️ 注意：具體 CU 數量、暫存器檔案大小、LDS 容量等硬體引數因具體 GPU 型號而異，參見各產品規格書，此處不列出過時或不確定的數字。

並行程式設計模型

HIP 支援與 CUDA 相同的三種並行層級：

資料並行（Data Parallel）：大量執行緒對不同資料執行相同操作
網格/塊並行（Grid/Block Parallel）：通過 Kernel 啟動配置控制並行度
協作組（Cooperative Groups）：HIP 提供 cooperative_groups 名稱空間，支援 Block 內、跨 Block 級別的執行緒協作同步

HIPIFY 的技術侷限

自動翻譯並非萬能，以下場景通常需要人工干預：

✅ 能自動轉換的：
   - 大部分 API 呼叫名替換（cudaXxx → hipXxx）
   - 標頭檔案路徑替換
   - Kernel 啟動語法替換
   - 錯誤碼型別替換（cudaError_t → hipError_t）

❌ 需要人工處理的：
   - NVIDIA 專有庫呼叫（cuDNN → 需替換為 MIOpen；
     NCCL → 需替換為 RCCL；cuBLAS → rocBLAS 等）
   - CUDA 特有的 PTX 內聯彙編
   - 依賴 cuSPARSE、cuSOLVER 等稀疏/求解庫的程式碼
   - Tensor Core 特有的 WMMA/MMA 內聯 PTX
   - 依賴 CUDA Graphs 等較新特性的程式碼（HIP 支援程度逐版本演進）
   - NVLink 特有的 P2P 訪問模式

技術演進史

時間	里程碑	意義
2016	AMD 釋出 ROCm 平台早期版本，HIP 作為核心元件首次亮相	打破 CUDA 壟斷的第一步
2017-2018	ROCm 1.x/2.x 持續迭代，支援 GCN 架構 GPU（如 Vega/Radeon Instinct MI25/MI50/MI60）	初步建立開發者社群
2019	HIPIFY 工具鏈成熟化，開始能較好處理中等規模 CUDA 專案移植	降低了實際工程遷移門檻
2020	ROCm 3.x/4.x，支援 MI100（Arcturus，CDNA1 架構），引入 Matrix Core 程式設計支援	首次在 HPC/AI 領域具備真正競爭力
2022	ROCm 5.x，支援 MI200 系列（CDNA2），引入 MI200 系列多 die 封裝特性支援	進入大型模型訓練競技場
2023	ROCm 5.5/5.6/5.7 + HIP 5.x，支援 MI300 系列（CDNA3），與 PyTorch 官方合作加深	主流 AI 架構正式後端支援
2024	ROCm 6.x 持續迭代，HIP 在 MI300X 上的生態適配成為 AMD AI 戰略核心	全面挑戰 NVIDIA 在大型模型訓練/推論的地位

具體版本號和時間線基於 AMD 公開發布記錄整理，個別迭代週期可能存在數週偏差。

技術路線對比

HIP vs CUDA vs SYCL vs OpenCL

維度	HIP (AMD)	CUDA (NVIDIA)	SYCL (Khronos)	OpenCL (Khronos)
維護方	AMD（開源）	NVIDIA（閉源）	Intel/Codeplay 等（開源實現多）	Khronos（規範）
目標硬體	主要 AMD GPU，亦可編譯到 NVIDIA GPU	NVIDIA GPU 專用	多廠商（CPU/GPU/FPGA）	多廠商（最廣泛）
程式語言	C/C++（類 CUDA 語法）	C/C++（原生語法）	C++（基於 SYCL 規範）	C（子集）
學習曲線	低（對 CUDA 開發者幾乎零成本）	基準（最大社群）	中（需學習 SYCL 模型）	高（API 冗長）
生態成熟度	中等，快速增長中	極高（行業事實標準）	較低，但增長中	低（HPC/AI 領域份額小）
AI 架構支援	PyTorch/TF 通過 ROCm 後端支援	原生最優支援	有限（Intel oneAPI 棧）	極少
Kernel 啟動	`hipLaunchKernelGGL`	`<<<` 語法	`cgh.parallel_for`	`clEnqueueNDRangeKernel`
記憶體管理	`hipMalloc/hipFree`	`cudaMalloc/cudaFree`	`sycl::malloc_device`	`clCreateBuffer`
分散式通訊	RCCL	NCCL	oneCCL（Intel）	無標準方案
典型部署規模	萬卡級開始出現（MI300X 叢集）	十萬卡級（H100/B200 叢集）	小規模	非 HPC 主流

評價總結

CUDA：生態護城河最深，是事實標準，但鎖定 NVIDIA 硬體
HIP：CUDA 生態的最佳”影子”——學 CUDA 語法、抄 CUDA 生態、降低遷移成本；但獨立生態仍弱於 CUDA
SYCL：理想主義的跨平台方案，但 AI 生態支援遠不及前兩者
OpenCL：在 AI/HPC 領域已被邊緣化

上下游

上游（HIP 依賴什麼）

┌─────────────────────────────────────────────────┐
│  LLVM/Clang 開源編譯器基礎設施                     │
│  （AMD 維護 AMDGPU LLVM 後端）                    │
├─────────────────────────────────────────────────┤
│  Linux Kernel（AMDGPU 驅動 + KFD）               │
│  → 提供使用者態到核心態的 GPU 排程通道               │
├─────────────────────────────────────────────────┤
│  AMD GPU 硬體 ISA（CDNA 指令集架構）               │
│  → 編譯目標二進位制的基礎                            │
├─────────────────────────────────────────────────┤
│  行業標準介面：PCIe / CXL / xGMI / UALink        │
│  → GPU 間互聯 & GPU-CPU 互聯                     │
└─────────────────────────────────────────────────┘

下游（誰依賴 HIP）

┌─────────────────────────────────────────────────┐
│  AI 架構層                                       │
│  PyTorch (ROCm 後端) / TensorFlow (ROCm 建置)    │
│  JAX (ROCm 支援) / ONNX Runtime                  │
├─────────────────────────────────────────────────┤
│  運算元庫 & 加速庫                                  │
│  rocBLAS / MIOpen / composable_kernel (CK)       │
│  FlashAttention-ROCm / vLLM-ROCm                 │
│  Triton（OpenAI 編譯器，可編譯到 AMD GPU）        │
├─────────────────────────────────────────────────┤
│  分散式通訊                                      │
│  RCCL（ROCm 集體通訊庫，對標 NCCL）               │
├─────────────────────────────────────────────────┤
│  企業 AI 平台                                    │
│  各大雲端廠商 AMD GPU 例項（Azure ND MI300X 等）    │
│  企業自建 AMD GPU 叢集                            │
└─────────────────────────────────────────────────┘

關鍵指標

評估 HIP/ROCm 生態成熟度的關鍵指標：

指標	說明	當前狀態（定性）
架構覆蓋率	主流 AI 架構的 ROCm 後端適配程度	PyTorch 已官方支援；TensorFlow 支援但更新頻率低於 NVIDIA；JAX 支援進行中
運算元庫完備度	MIOpen/rocBLAS 等對標 cuDNN/cuBLAS 的覆蓋範圍	核心運算元已覆蓋，但長尾運算元和最新研究性運算元仍有差距
HIPIFY 轉換成功率	大型 CUDA 專案一鍵轉換後可編譯執行的比例	因專案而異；簡單專案可達較高成功率，複雜專案（深度使用 CUDA 專有特性）需大量手動修改
社群活躍度	GitHub issue/PR 數量、開發者論壇活躍度	持續增長，但絕對數量級仍遠小於 CUDA 生態
效能對標	HIP/ROCm 路徑 vs CUDA 路徑在相同模型訓練中的吞吐比	因 workload 和版本差異大；MI300X 在部分 LLM 訓練中已接近或達到競爭力水平 [多來源報道，具體數字因 benchmark 配置而異]

供需與市場資料

供給側

AMD GPU 產品線（與 HIP/ROCm 生態直接相關）：
- MI300X：旗艦 AI 訓練/推論加速卡，整合大容量 HBM3（具體容量/位寬參見 AMD 官方規格書）
- MI300A：APU 形態，CPU+GPU 融合封裝
- MI200 系列（MI250X/MI250）：上一代旗艦，仍廣泛部署
產能：依賴台積電先進封裝（CoWoS 類方案，具體合作模式以供應鏈報道為準 [行業報告]）

需求側

大型模型訓練：數千至數萬張 MI300X 的叢集部署已在多家頭部 AI 公司進行 [行業報道]
推論部署：MI300X 的大 HBM 容量優勢使其在長上下文推論場景具備吸引力
HPC：超級計算機（如 El Capitan 等）採用 AMD GPU + ROCm 棧

市場規模

全球 AI 加速器市場：2024 年預估約 $500-700 億美元 [多家行業分析機構估算，口徑有差異]
AMD 資料中心 GPU 營收：2023-2024 財年快速爬坡，從 ~$10 億量級向更高規模增長 [AMD 財報]
ROCm/HIP 生態的市場份額：在 AI 加速器軟體棧中仍為個位數百分比，但增長趨勢明顯

⚠️ 上述市場資料為行業估算範圍，不同來源差異較大，投資決策需以各公司財報和權威報告為準。

代表公司與資本對映

層級	代表公司	與 HIP 的關係	公開交易資訊
核心提供方	AMD	HIP/ROCm 開發和維護者	NASDAQ: AMD
編譯器基礎設施	AMD + LLVM 社群	AMD 維護 AMDGPU LLVM 後端	LLVM 為開源專案
架構整合	Meta (PyTorch), Google (JAX/TF)	接受 AMD 貢獻的 ROCm 後端程式碼	NASDAQ: META, NASDAQ: GOOGL
下游雲端使用者	Microsoft Azure, Oracle Cloud	部署 MI300X 例項，間接依賴 ROCm/HIP 棧	NASDAQ: MSFT, NYSE: ORCL
競品生態	NVIDIA	CUDA 生態的擁有者，HIP 的對標物件	NASDAQ: NVDA
替代方案	Intel (oneAPI/SYCL)	跨平台異構計算的另一條路	NASDAQ: INTC

投資邏輯

看多 HIP/ROCm 生態的理由

硬體價效比已具備競爭力：MI300X 在多項 AI workload 中的價效比/效能已進入可比範圍，給企業遷移提供了經濟動力。
大客戶驅動生態建設：頭部 AI 公司（雲端廠商、大型模型公司）為了降低對 NVIDIA 單一供應商的依賴，有強烈動機投入 ROCm 生態適配。
開源策略降低壁壘：ROCm/HIP 全棧開源，企業可以自行 debug 和最佳化，不完全依賴 AMD 響應速度。
Triton 編譯器的機會：OpenAI Triton 作為新一代 ML 編譯器，可編譯到 AMD GPU，若 Triton 成為主流 Kernel 程式設計方式，將大幅削弱 CUDA 的直接程式設計壁壘。

風險與質疑

生態差距仍然巨大：CUDA 生態積累 15+ 年，HIP/ROCm 在運算元庫覆蓋度、除錯工具、profiler、社群知識沉澱等方面的差距不是短期能彌補的。
軟體穩定性擔憂：ROCm 在生產環境中的穩定性和 bug 率仍有改善空間，部分使用者反饋在複雜分散式訓練中偶發問題。
AMD 軟體投入的持續性：AMD 長期是硬體強、軟體弱的公司形象，ROCm/HIP 的投入力度能否持續加碼是關鍵變數。
NVIDIA 反制：NVIDIA 可通過 CUDA 版本迭代、新 API（如 CUDA Graphs、NVSHMEM 等）持續製造新的遷移摩擦點。

常見誤讀糾偏

誤讀 1：“HIP 程式碼可以無縫在 NVIDIA GPU 上執行”

糾偏：HIP 設計上支援雙向編譯，但實際操作中，從 CUDA 到 HIP 的移植並非一鍵完成。HIPIFY 工具能處理大部分機械性的 API 替換，但以下場景仍需大量人工工作：

使用 NVIDIA 專有數學庫（cuDNN、cuBLAS、NCCL 等）的程式碼需替換為對應的 ROCm 庫
內聯 PTX 彙編無法自動轉換
效能調優引數（block size、shared memory 配置等）在 AMD 和 NVIDIA 架構上通常需要不同的最優值
某些 CUDA 版本的新特性在 HIP 中可能尚未實現或實現有差異

因此更準確的說法是：HIP 大幅降低了遷移成本，但並未消除遷移成本。

誤讀 2：“HIP 只是一個翻譯層，效能必然打折”

糾偏：HIP 編譯到 AMD GPU 時，走的是 原生編譯路徑（LLVM/Clang → AMDGPU ISA），並非先翻譯成 CUDA 再編譯。Kernel 在 AMD GPU 上是以原生機器碼執行的。效能上限取決於：

編譯器後端對 AMDGPU ISA 的最佳化質量
Kernel 本身對 AMD CU 架構的適配程度（如 wavefront 大小差異：AMD 傳統 GCN/CDNA 為 64 寬 wave，NVIDIA 為 32 寬 warp）
記憶體訪問模式是否對齊 AMD 的 HBM 子通道架構

因此 HIP 不是模擬層，而是原生程式設計介面，效能潛力主要取決於軟體最佳化程度和編譯器成熟度。

誤讀 3：“ROCm 生態等同於 HIP”

糾偏：ROCm 是一個完整的軟體平台，HIP 是其中的程式設計介面層。ROCm 還包含：

ROCm 核心驅動（與 Linux 核心互動）
數學庫：rocBLAS、rocFFT、rocSOLVER、rocRAND 等
AI 運算元庫：MIOpen（對標 cuDNN）、composable_kernel
通訊庫：RCCL（對標 NCCL）
除錯/分析工具：rocprof、rocdbg 等
系統管理：rocm-smi（裝置監控）

只關注 HIP 而忽略整個 ROCm 棧的完善程度，會嚴重低估生態建設的工程量。

學習路徑

入門（1-2 天）

前提：有 C/C++ 基礎，最好有 CUDA 程式設計經驗
閱讀 AMD ROCm 官方文件中的 HIP Programming Guide
安裝 ROCm 開發環境（建議 Ubuntu + AMD 官方 deb 倉庫）
執行 HIP 官方示例（hip-examples 或 rocm-examples 倉庫），理解基本的 Kernel 啟動和記憶體管理

進階（1-2 周）

使用 HIPIFY 工具將一個小型 CUDA 專案轉換為 HIP，手動修復轉換失敗的部分
學習 MIOpen（如果做深度學習）或 rocBLAS（如果做線性代數）的使用
理解 AMD CDNA 架構的執行模型：wavefront、CU 架構、Matrix Core 程式設計
使用 rocprof 進行效能分析

高階（1-3 月）

閱讀 composable_kernel 原始碼，理解 AMD 風格的高效能 Kernel 編寫範式
參與 ROCm GitHub 社群，閱讀/提交 issue 和 PR
嘗試在多 GPU 環境下使用 RCCL 進行集合通訊
研究 HIP 在 MI300 系列多 die 架構上的 NUMA 感知程式設計

一句話總結

HIP 是 AMD 破解 NVIDIA CUDA 生態壟斷的關鍵軟體武器——它通過提供一套與 CUDA 幾乎一一對應的開源程式設計介面，大幅降低了 AI/HPC 應用從 NVIDIA GPU 遷移到 AMD GPU 的工程成本，是 ROCm 生態的基石，也是 AMD 資料中心 GPU 業務能否真正放量的核心軟體變數。

延伸閱讀與來源

來源	說明
AMD ROCm 官方文件	HIP API 參考、程式設計指南、安裝指南（最權威的官方來源）
AMD ROCm GitHub	HIP 執行時、ROCm 各元件原始碼
AMD GPUOpen 部落格	技術深度文章、效能最佳化指南
AMD 財報	資料中心 GPU 營收、ROCm 生態投入規模
PyTorch ROCm 文件	PyTorch 對 ROCm 後端的支援矩陣
LLVM AMDGPU 後端文件	理解 HIP 編譯器底層機制
各第三方 AI 評測報告	效能對比資料（具體數字因 benchmark 配置和 ROCm 版本差異大，建議交叉驗證）

免責宣告：本頁市場資料和行業估算來自多渠道綜合，不同來源口徑有差異，不構成投資建議。技術規格以各廠商官方文件為準。