0.2 ROCm 软件栈全景：libhsakmt 在 rocm 中的位置

原创已于 2026-05-05 10:09:24 修改 · 592 阅读

9 GEO检测

于 2026-04-27 05:00:00 首次发布

ROCm rocr-libhsakmt库实现剖析专栏收录该内容

71 篇文章 ¥79.90 ¥99.00

订阅专栏

在深入 libhsakmt 的逐模块分析之前，我们需要先建立一个全局视角：ROCm 软件栈的整体架构是什么样的？libhsakmt 究竟处于哪一层？它与上层的 HSA Runtime、下层的 KFD 内核驱动之间是什么关系？本篇将回答这些问题，为后续的深度剖析奠定架构认知基础。

1. ROCm 软件栈分层架构

ROCm（Radeon Open Compute）是 AMD 面向 GPU 通用计算的开源软件平台。其软件栈采用严格的分层设计，从上到下可划分为五个层次：

+---------------------------------------------------+
|            User App / AI Framework / HPC          |
+---------------------------------------------------+
|        Programming Model: HIP / OpenCL / OpenMP   |
+---------------------------------------------------+
|    HSA Runtime (libhsa-runtime64.so) -- ROCr      |
+---------------------------------------------------+
|    Thunk Layer (libhsakmt) -- ROCt                |
+---------------------------------------------------+
|    KFD Kernel Driver (/dev/kfd) -- ROCk           |
+---------------------------------------------------+
|    AMDGPU DRM Driver + GPU Hardware               |
+---------------------------------------------------+

第一层：用户应用。 PyTorch、TensorFlow、各类 HPC 科学计算程序，它们通过 HIP 或 OpenCL 等编程接口使用 GPU。

第二层：编程模型层。 HIP 是 AMD 提供的 GPU 编程接口（类似 CUDA），负责将 hipMalloc、hipLaunchKernel 等高层调用翻译为 HSA Runtime API。OpenCL 运行时也位于这一层。

第三层：HSA Runtime（ROCr）。 AMD 对 HSA（Heterogeneous System Architecture）标准的实现，以 libhsa-runtime64.so 的形式交付。它提供标准的 hsa_*() 系列 C API，负责 GPU 代理（Agent）管理、内存区域抽象、AQL 队列调度、信号同步、代码对象加载等核心功能。

第四层：Thunk 层（ROCt）。 即本专栏的主角 ——libhsakmt。它是用户态与内核态之间的"翻译层"，将 HSA Runtime 的操作请求转换为对 /dev/kfd 设备节点的 ioctl 系统调用。

第五层：KFD 内核驱动（ROCk）。 运行在内核态的 AMD GPU 计算驱动，负责硬件队列的实际创建、GPU 页表的管理、中断处理等底层硬件操作。

2. libhsakmt 的定位：用户态与内核态的桥梁

libhsakmt 的全称是 HSA Kernel Mode Thunk，“Thunk"一词源自编译器术语，意为"转换层"或"适配层”。它的核心职责可以用一句话概括：

将用户态的 GPU 资源管理请求，翻译为内核态 KFD 驱动能理解的 ioctl 调用。

具体来说，当上层 HSA Runtime 需要创建一个 GPU 队列时，它不会直接操作 /dev/kfd，而是调用 libhsakmt 提供的 hsaKmtCreateQueue() 函数。libhsakmt 负责：

将参数组装为 KFD ioctl 所需的数据结构；
通过 ioctl() 系统调用发送给内核驱动；
处理返回值，将内核态的错误码转换为用户态的状态码。

这种设计的好处是：上层 Runtime 不需要关心内核接口的细节变化，libhsakmt 作为中间层承担了版本适配与接口稳定的责任。

熟悉 Linux 图形栈的朋友可以这样理解：libdrm 是图形渲染栈（Mesa / Vulkan）与 DRM 内核驱动之间的用户态桥梁，libhsakmt 在 GPU 计算栈中扮演的角色与之完全对等 —— 它是 HSA Runtime 与 KFD 内核驱动之间的用户态桥梁。两者都通过 ioctl 与各自的内核驱动通信，只是一个面向图形，一个面向计算。

3. libhsakmt 与 HSA Runtime 的构建关系

一个容易产生困惑的问题是：libhsakmt 是一个独立的共享库吗？

答案是：不是。 在当前的构建体系中，libhsakmt 始终被编译为静态库，然后链接进 libhsa-runtime64.so。最终交付给用户的只有一个库文件 libhsa-runtime64.so，libhsakmt 的代码已经嵌入其中。

从构建系统的角度看：

rocr-runtime/
├── libhsakmt/          → 编译为 libhsakmt.a（静态库）
└── runtime/
    └── hsa-runtime/    → 编译为 libhsa-runtime64.so（链接 libhsakmt.a）

这意味着，虽然 libhsakmt 在源码层面是一个独立模块，但在运行时它是 HSA Runtime 的一部分。用户的应用程序只需链接 libhsa-runtime64.so，即可间接使用 libhsakmt 的所有功能。

4. HSA Runtime 如何调用 libhsakmt

HSA Runtime 内部通过一个 Driver 抽象层 来访问 libhsakmt。核心类层次如下：

core::Driver（抽象基类）
    ├── AMD::KfdDriver      → 调用 hsaKmt*() 函数（Linux 原生路径）
    ├── AMD::XdnaDriver     → 面向 AI Engine（NPU）的驱动后端
    └── AMD::VirtioDriver   → 面向虚拟化环境的 VirtIO 后端

KfdDriver 是最核心的实现，它将 Runtime 层的操作逐一映射到 libhsakmt 的 API：

HSA Runtime 操作	KfdDriver 调用的 libhsakmt API	功能
发现 GPU 设备	`hsaKmtOpenKFD` + `hsaKmtAcquireSystemProperties`	打开 KFD 并获取系统拓扑
查询节点属性	`hsaKmtGetNodeProperties` / `GetNodeMemoryProperties`	获取 GPU 核心数、显存大小等
创建调度队列	`hsaKmtCreateQueue`	创建硬件命令队列
分配 GPU 内存	`hsaKmtAllocMemory` + `hsaKmtMapMemoryToGPU`	分配并映射显存
同步等待	`hsaKmtCreateEvent` + `hsaKmtWaitOnEvent`	创建硬件事件并等待完成
性能采集	`hsaKmtPmcStartTrace` / `hsaKmtPmcQueryTrace`	启动 / 查询硬件性能计数器

这种 Driver 抽象的设计使得 Runtime 能够支持多种后端：标准 KFD 路径、虚拟化 VirtIO 路径、甚至 AI 引擎（XDNA）路径，而上层代码无需修改。

5. 一次 hipMalloc 的完整调用链路

为了更直观地理解 libhsakmt 的位置，我们追踪一次 hipMalloc() 调用从用户代码到硬件的完整路径：

hipMalloc(&ptr, size)
    |
    v
HIP Runtime: parse args, select target GPU
    |
    v
HSA Runtime (hsa_amd_memory_pool_allocate):
    |  create alloc request, select MemoryRegion
    v
KfdDriver::AllocateMemory():
    |  call libhsakmt hsaKmtAllocMemory()
    v
libhsakmt (memory.c -> fmm.c):
    |  assemble ioctl params, manage Aperture VA space
    |  ioctl(fd, AMDKFD_IOC_ALLOC_MEMORY_OF_GPU, ...)
    v
KFD Kernel Driver:
    |  alloc physical VRAM, setup GPU page table mapping
    v
AMDGPU DRM + GPU Hardware: actual VRAM allocation

从这条链路可以清晰地看到：libhsakmt 是 用户态最后一道关卡，它之下就是内核态的 ioctl 调用。所有 GPU 资源操作的"最后一公里"，都由 libhsakmt 来完成。

6. HSA Runtime 的核心子系统（libhsakmt 之上）

理解 libhsakmt 之上的 HSA Runtime 子系统，有助于把握 libhsakmt 各模块被调用的场景：

Runtime 子系统	职责	依赖的 libhsakmt 模块
Topology（拓扑发现）	枚举 CPU / GPU Agent，构建设备连接图	OpenClose、Topology
Agent（设备代理）	封装 GPU/CPU 节点的属性与能力	Topology
Memory Region（内存区域）	抽象 GPU 显存、系统内存、LDS 等内存池	Memory、FMM
AQL Queue（队列调度）	管理 HSA AQL 格式的命令队列	Queue
Signal（信号系统）	实现硬件/软件信号与 IPC 信号	Event
Blit Engine（拷贝引擎）	SDMA 和 Kernel 两种 GPU 内存拷贝路径	Queue、Memory
Code Loader（代码加载器）	加载 GPU ELF 代码对象，解析符号与重定位	Memory
PC Sampling（采样扩展）	GPU 程序计数器采样	PC Sampling

7. 小结

将本篇的核心信息浓缩为三句话：

libhsakmt 是 ROCm 软件栈中用户态与内核态的分界线 —— 它之上是纯用户态的 C++ 抽象（HSA Runtime），它之下是通过 ioctl 进入内核的系统调用。
libhsakmt 以静态库形式嵌入 libhsa-runtime64.so —— 源码独立，运行时合一。HSA Runtime 通过 Driver 抽象层调用 libhsakmt 的 hsaKmt*() 系列 API。
理解 libhsakmt 就是理解 ROCm GPU 资源管理的本质 —— 拓扑发现、内存分配、队列创建、事件同步、性能采集，所有硬件资源操作最终都汇聚于此。

后续各章将逐一深入 libhsakmt 的每个模块，从 OpenClose 的设备握手开始，到 VirtIO 虚拟化后端收尾，带你走完这条从硬件到用户态的完整技术链路。

标签