第02章：HSA-Runtime架构概览

章节概述
本章深入HSA Runtime的内部架构，介绍其三层结构设计、核心组件关系以及初始化流程。通过理解Runtime的整体架构，你将为后续学习具体模块打下坚实基础。

难度级别: 🟢 基础
预计阅读时间: 45分钟
前置知识: 第01章 - 什么是HSA与异构计算

📋 本章学习目标

完成本章学习后，你将能够：

• ✅ 理解HSA Runtime的三层架构设计
• ✅ 掌握主要组件（Agent、Queue、Signal、Memory）的关系
• ✅ 了解Runtime的初始化与销毁流程
• ✅ 熟悉Runtime配置选项和环境变量
• ✅ 知道如何启用调试和追踪功能

2.1 Runtime层次结构

HSA Runtime采用三层架构设计，实现了接口与实现的分离，提供了良好的可扩展性。

2.1.1 三层架构总览

┌─────────────────────────────────────────────────┐ │ 应用程序 (User Application) │ │ (C/C++, Python, HIP, OpenCL等) │ └─────────────────────────────────────────────────┘ ↓ 调用 ┌─────────────────────────────────────────────────┐ │ 第一层：公共API层 (Public API Layer) │ │ ┌───────────────────────────────────────────┐ │ │ │ inc/hsa.h - HSA标准API │ │ │ │ inc/hsa_ext_amd.h - AMD扩展API │ │ │ │ inc/hsa_ext_image.h - Image扩展 │ │ │ │ inc/hsa_ext_finalize.h - Finalizer扩展 │ │ │ └───────────────────────────────────────────┘ │ │ │ │ 特点： │ │ • C语言接口，ABI稳定 │ │ • 遵循HSA规范 │ │ • 跨平台、跨厂商 │ └─────────────────────────────────────────────────┘ ↓ 实现 ┌─────────────────────────────────────────────────┐ │ 第二层：核心实现层 (Core Runtime Layer) │ │ ┌───────────────────────────────────────────┐ │ │ │ core/runtime/runtime.cpp │ │ │ │ - Runtime主控类 │ │ │ │ - Agent管理 │ │ │ │ - 资源协调 │ │ │ └───────────────────────────────────────────┘ │ │ ┌───────────────────────────────────────────┐ │ │ │ 核心组件实现 (C++) │ │ │ │ ├─ core/runtime/amd_gpu_agent.cpp │ │ │ │ ├─ core/runtime/amd_cpu_agent.cpp │ │ │ │ ├─ core/runtime/amd_aql_queue.cpp │ │ │ │ ├─ core/runtime/signal.cpp │ │ │ │ ├─ core/runtime/amd_memory_region.cpp │ │ │ │ └─ core/runtime/amd_topology.cpp │ │ │ └───────────────────────────────────────────┘ │ │ ┌───────────────────────────────────────────┐ │ │ │ Driver抽象 (core/inc/driver.h) │ │ │ │ - 面向对象的驱动接口封装 │ │ │ │ - Driver基类定义 │ │ │ │ - 支持多种驱动类型(KFD/XDNA/Virtio) │ │ │ │ - 统一的硬件操作接口 │ │ │ └───────────────────────────────────────────┘ │ │ │ │ 特点： │ │ • C++实现，面向对象设计 │ │ • 设备无关的抽象 │ │ • 业务逻辑核心 │ └─────────────────────────────────────────────────┘ ↓ 调用 ┌─────────────────────────────────────────────────┐ │ 第三层：驱动抽象层 (Driver Abstraction Layer) │ │ ┌───────────────────────────────────────────┐ │ │ │ Thunk Layer (libhsakmt.so) │ │ │ │ - KFD用户态封装 │ │ │ │ - 系统调用转换 │ │ │ │ - 设备枚举 │ │ │ └───────────────────────────────────────────┘ │ │ ┌───────────────────────────────────────────┐ │ │ │ KFD (Kernel Fusion Driver) │ │ │ │ - amdgpu内核模块 │ │ │ │ - 硬件资源管理 │ │ │ │ - 内存管理 │ │ │ │ - 队列调度 │ │ │ └───────────────────────────────────────────┘ │ │ │ │ 特点： │ │ • 内核态驱动 │ │ • 直接操作硬件 │ │ • 权限管理 │ └─────────────────────────────────────────────────┘ ↓ 控制 ┌─────────────────────────────────────────────────┐ │ GPU硬件 (AMD GPU Hardware) │ │ GCN/RDNA/CDNA架构 │ └─────────────────────────────────────────────────┘ 

2.1.2 公共API层 (inc/hsa.h)

这一层定义了符合HSA规范的C语言接口，是应用程序的直接入口：

// inc/hsa.h - 主要API分类// 1. Runtime生命周期hsa_status_thsa_init(void);hsa_status_thsa_shut_down(void);// 2. 系统与Agent信息hsa_status_thsa_system_get_info(hsa_system_info_t attribute,void* value);hsa_status_thsa_iterate_agents(hsa_status_t(*callback)(hsa_agent_t agent,void* data),void* data);hsa_status_thsa_agent_get_info(hsa_agent_t agent,hsa_agent_info_t attribute,void* value);// 3. 内存管理hsa_status_thsa_memory_allocate(hsa_region_t region,size_t size,void** ptr);hsa_status_thsa_memory_free(void* ptr);// 4. 队列操作hsa_status_thsa_queue_create(hsa_agent_t agent,uint32_t size,hsa_queue_type_t type,void(*callback)(hsa_status_t,hsa_queue_t*,void*),void* data,uint32_t private_segment_size,uint32_t group_segment_size,hsa_queue_t** queue);// 5. 信号同步hsa_status_thsa_signal_create(hsa_signal_value_t initial_value,uint32_t num_consumers,consthsa_agent_t* consumers,hsa_signal_t* signal);hsa_signal_value_thsa_signal_wait_scacquire(hsa_signal_t signal,hsa_signal_condition_t condition,hsa_signal_value_t compare_value,uint64_t timeout_hint,hsa_wait_state_t wait_state_hint);

API设计原则：

回调遍历: 用于枚举资源

hsa_iterate_agents([](hsa_agent_t agent,void* data){// 处理每个agentreturn HSA_STATUS_SUCCESS;// 继续遍历}, user_data);

状态码返回: 统一的错误处理

typedefenum{ HSA_STATUS_SUCCESS =0, HSA_STATUS_ERROR_INVALID_ARGUMENT =1, HSA_STATUS_ERROR_OUT_OF_RESOURCES =3,// ... 更多状态码}hsa_status_t;

Opaque Handle: 使用句柄（handle）隐藏内部实现

typedefstructhsa_agent_s{uint64_t handle;}hsa_agent_t;typedefstructhsa_queue_s*hsa_queue_t;

2.1.3 核心实现层 (core/runtime/)

这一层是Runtime的业务逻辑核心，使用C++实现面向对象设计：

核心类体系：

// core/inc/runtime.hnamespace core {classRuntime{public:// 单例模式static Runtime* runtime_singleton_;staticboolAcquire();staticboolRelease();// Agent管理const std::vector<core::Agent*>&agents()const; std::vector<Agent*>&gpu_agents(){return gpu_agents_;} std::vector<Agent*>&cpu_agents(){return cpu_agents_;}// 内存区域const std::vector<core::MemoryRegion*>&regions()const;// 驱动管理 - 支持多种驱动后端 std::vector<std::unique_ptr<Driver>>&AgentDrivers(){return agent_drivers_;} Driver*GetDriverByType(DriverType drv_type);// 系统拓扑voidRegisterAgent(Agent* agent);private:// 驱动接口数组 - 可同时支持多种类型的驱动// 例如：KFD(用于GPU)、XDNA(用于AIE)等 std::vector<std::unique_ptr<Driver>> agent_drivers_;// 按设备类型分类的Agent列表 std::vector<Agent*> gpu_agents_;// GPU agents std::vector<Agent*> cpu_agents_;// CPU agents std::vector<Agent*> aie_agents_;// AIE agents (使用XDNA驱动)// 所有Agent的统一索引（按node_id） std::map<uint32_t, Agent*> agents_by_node_;};// core/inc/agent.hclassAgent:publicChecked<0x3D82E96E87AB41B1>{public:enumDeviceType{ kAmdCpuDevice, kAmdGpuDevice, kAmdDspDevice };virtual hsa_status_t QueueCreate(/* ... */)=0;virtual hsa_status_t DmaCopy(/* ... */)=0;uint32_tnode_id()const{return node_id_;} DeviceType device_type()const{return device_type_;}protected:uint32_t node_id_; DeviceType device_type_; Driver* driver_;// 每个Agent持有对应驱动的引用};}// namespace core

设计特点：

• Runtime使用驱动数组而非单一驱动，支持异构系统中的多种设备
• 每个Driver实例负责一类设备（如KFD负责GPU，XDNA负责AIE）
• Agent创建时从Runtime获取对应类型的Driver引用

Driver类的角色：
core::Driver 类是核心实现层对底层驱动的封装，它：

• 提供统一的C++接口给上层组件（Runtime、Agent等）
• 支持多种驱动后端（KFD、XDNA、KFD_VIRTIO）
• 封装了所有与内核驱动交互的细节

// core/inc/driver.hnamespace core {enumclassDriverType{ XDNA =0, KFD,#ifdefHSAKMT_VIRTIO_ENABLED KFD_VIRTIO,#endif NUM_DRIVER_TYPES };classDriver{public:// 初始化与关闭virtual hsa_status_t Init()=0;virtual hsa_status_t ShutDown()=0;// 系统信息查询virtual hsa_status_t GetSystemProperties(HsaSystemProperties& sys_props)const=0;virtual hsa_status_t GetNodeProperties(HsaNodeProperties& node_props,uint32_t node_id)const=0;// 内存操作virtual hsa_status_t AllocateMemory(const MemoryRegion& mem_region,/*...*/)=0;virtual hsa_status_t FreeMemory(void* mem, size_t size)=0;// 队列操作virtual hsa_status_t CreateQueue(uint32_t node_id, HSA_QUEUE_TYPE type,/*...*/)const=0;virtual hsa_status_t DestroyQueue(HSA_QUEUEID queue_id)const=0;// 时钟与配置virtual hsa_status_t GetClockCounters(uint32_t node_id, HsaClockCounters* clock_counter)const=0;protected:const DriverType kernel_driver_type_;const std::string devnode_name_;// 如 "/dev/kfd"int fd_ =-1;// 设备文件描述符};}// namespace core

设计模式：

• 工厂模式: Runtime根据硬件类型创建Agent和Driver
• 单例模式: Runtime全局唯一实例
• 策略模式: 不同类型的Queue、Signal、Driver实现
• 观察者模式: Signal等待机制

2.1.4 Driver的角色与定位

在介绍驱动抽象层之前，我们需要理解Driver在Runtime中的角色。

Driver在Runtime中的组织方式

从上面的Runtime类定义可以看到，Runtime维护了一个 agent_drivers_ 数组：

classRuntime{private:// 驱动接口数组 - 支持同时管理多种驱动类型 std::vector<std::unique_ptr<Driver>> agent_drivers_;};

为什么是数组而不是单个Driver？

现代异构系统中可能同时存在多种类型的计算设备：

• GPU设备 → 使用KFD驱动 (DriverType::KFD)
• AIE设备 → 使用XDNA驱动 (DriverType::XDNA)
• 虚拟化环境 → 使用Virtio驱动 (DriverType::KFD_VIRTIO)

Runtime需要同时管理这些不同类型的驱动，因此使用数组来存储。

Driver的获取与使用

Runtime提供方法根据驱动类型获取对应的Driver实例：

// 根据驱动类型获取Driver Driver*Runtime::GetDriverByType(DriverType drv_type){auto is_driver_type =[drv_type](const std::unique_ptr<Driver>& d){return d->kernel_driver_type_ == drv_type;};auto driver = std::find_if(agent_drivers_.begin(), agent_drivers_.end(), is_driver_type);return(driver != agent_drivers_.end())? driver->get():nullptr;}

使用场景：

• Agent创建时获取对应类型的Driver
• Runtime初始化时根据系统配置创建相应Driver
• 支持多种驱动共存于同一系统

2.1.5 驱动抽象层详解 (Driver + libhsakmt + KFD)

现在让我们深入了解Driver的具体设计。

Driver类接口设计 (core::Driver)

Driver类提供了统一的硬件操作接口，所有具体驱动都继承此基类：

// core/inc/driver.h - Driver抽象基类namespace core {enumclassDriverType{ XDNA =0,// 用于AIE设备 KFD,// 用于AMD GPU (通过/dev/kfd) KFD_VIRTIO,// 用于虚拟化GPU NUM_DRIVER_TYPES };classDriver{public:Driver(DriverType type, std::string devnode_name):kernel_driver_type_(type),devnode_name_(devnode_name){}// 生命周期管理virtual hsa_status_t Init()=0;virtual hsa_status_t Open()=0;// 打开设备节点virtual hsa_status_t Close()=0;// 关闭设备// 系统拓扑查询virtual hsa_status_t GetSystemProperties(...)const=0;virtual hsa_status_t GetNodeProperties(...)const=0;// 硬件资源管理virtual hsa_status_t AllocateMemory(...)=0;virtual hsa_status_t CreateQueue(...)const=0;virtual hsa_status_t GetClockCounters(...)const=0;// ... 更多接口// 驱动类型标识 DriverType driver_type()const{return kernel_driver_type_;}protected:const DriverType kernel_driver_type_;// 驱动类型const std::string devnode_name_;// 设备节点路径int fd_ =-1;// 设备文件描述符};}// namespace core

Driver的核心职责：

1. 🔧 接口统一：为不同硬件提供统一的C++接口
2. 🎯 类型隔离：通过DriverType区分不同的驱动后端
3. 🔄 参数转换：将HSA类型转换为底层驱动所需的类型
4. 📊 生命周期：管理设备连接、资源分配与释放

Driver的初始化与Agent的关联

了解了Driver的接口设计后，让我们看看它如何在Runtime中被创建和使用：

// Runtime初始化时创建Driver实例boolRuntime::Acquire(){if(ref_count_ ==0){// 1. 根据系统配置创建所需的Driver// 例如：检测到GPU则创建KFD Driver，检测到AIE则创建XDNA Driver// 创建KFD Driver用于GPUif(HasGpuDevices()){auto kfd_driver = std::make_unique<KfdDriver>( DriverType::KFD,"/dev/kfd"); kfd_driver->Init(); kfd_driver->Open(); agent_drivers_.push_back(std::move(kfd_driver));}// 创建XDNA Driver用于AIE（如果存在）if(HasAieDevices()){auto xdna_driver = std::make_unique<XdnaDriver>( DriverType::XDNA,"/dev/accel/accel0"); xdna_driver->Init(); xdna_driver->Open(); agent_drivers_.push_back(std::move(xdna_driver));}// 2. 枚举设备并创建Agent// 每个Agent获取其对应类型的Driver引用for(auto& driver : agent_drivers_){CreateAgentsForDriver(driver.get());}} ref_count_++;returntrue;}// 为特定Driver创建AgentvoidRuntime::CreateAgentsForDriver(Driver* driver){ HsaSystemProperties sys_props; driver->GetSystemProperties(sys_props);for(uint32_t node =0; node < sys_props.NumNodes; node++){ HsaNodeProperties node_props; driver->GetNodeProperties(node_props, node); Agent* agent =nullptr;if(IsGpuNode(node_props)){// GPU Agent - 传入KFD Driver引用 agent =newGpuAgent(node, node_props, driver); gpu_agents_.push_back(agent);}elseif(IsAieNode(node_props)){// AIE Agent - 传入XDNA Driver引用  agent =newAieAgent(node, node_props, driver); aie_agents_.push_back(agent);}if(agent){ agents_by_node_[node]= agent;}}}

Agent如何使用Driver：

// core/runtime/amd_gpu_agent.cppnamespace amd {classGpuAgent:public core::Agent{private: core::Driver* driver_;// 持有对应Driver的引用public:GpuAgent(uint32_t node_id,const HsaNodeProperties& props, core::Driver* driver):node_id_(node_id),driver_(driver){// Agent不拥有Driver，只持有引用// Driver的生命周期由Runtime管理}// 所有硬件操作都通过driver_完成 hsa_status_t QueueCreate(uint32_t size, hsa_queue_t** queue)override{void* ring_buffer; HSAuint64 queue_id;// 调用Driver接口，Driver负责与底层通信return driver_->CreateQueue( node_id_, size,&queue_id,&ring_buffer);} hsa_status_t AllocateMemory(size_t size,void** ptr)override{return driver_->AllocateMemory( node_id_, size, HSA_MEM_HEAP_TYPE_FB_PUBLIC, ptr);}};}// namespace amd

关键设计要点：

• ✅ Runtime持有Driver的所有权（unique_ptr）
• ✅ Agent只持有Driver的引用（裸指针）
• ✅ 多个Agent可以共享同一个Driver实例
• ✅ Driver的生命周期由Runtime统一管理

💡 详细内容：Driver的完整接口定义、KFD交互机制、实战案例等详见 第09章 - Driver驱动抽象层

libhsakmt层与KFD

libhsakmt（原来也叫做Thunk）层是用户态的C接口库，封装了与KFD内核驱动的交互：

// 示例：Thunk提供的接口 HSAKMT_STATUS HSAKMTAPI hsaKmtOpenKFD(void); HSAKMT_STATUS HSAKMTAPI hsaKmtGetNodeProperties( HSAuint32 NodeId, HsaNodeProperties* NodeProperties); HSAKMT_STATUS HSAKMTAPI hsaKmtAllocMemory( HSAuint32 PreferredNode, HSAuint64 SizeInBytes, HsaMemFlags MemFlags,void** MemoryAddress); HSAKMT_STATUS HSAKMTAPI hsaKmtCreateQueue( HSAuint32 NodeId, HSA_QUEUE_TYPE Type, HSAuint32 QueuePercentage, HSA_QUEUE_PRIORITY Priority,void* QueueAddress, HSAuint64 QueueSizeInBytes, HsaEvent* Event, HsaQueueResource* QueueResource);

KFD (Kernel Fusion Driver): 内核态驱动

• 位于Linux内核的 drivers/gpu/drm/amd/amdkfd/
• 通过ioctl与用户态通信
• 管理GPU资源、内存、队列

三者关系：

Runtime/Agent (C++) ↓ 调用 core::Driver (C++ 抽象) ↓ 调用 Thunk Layer (libhsakmt.so, C接口) ↓ ioctl系统调用 KFD (内核驱动) ↓ 操作 GPU硬件 

2.2 主要组件关系图

2.2.1 核心组件交互

┌────────────────────────────────────────────────────────────┐ │ Runtime (单例) │ │ ┌──────────────────────────────────────────────────────┐ │ │ │ • 全局状态管理 │ │ │ │ • Agent分类管理 (GPU/CPU/AIE) │ │ │ │ • MemoryRegion池 │ │ │ │ • Extension管理 │ │ │ └──────────────────────────────────────────────────────┘ │ │ │ │ │ │持有（unique_ptr） │ │ ↓ │ │ ┌──────────────────────────────────────┐ │ │ │ agent_drivers_: vector<unique_ptr> │ │ │ │ ├─ KfdDriver (for GPU) │ │ │ │ ├─ XdnaDriver (for AIE) │ │ │ │ └─ VirtioDriver (for VM) │ │ │ └──────────────────────────────────────┘ │ │ │ │ │ │ │共享引用 │管理 │ │ ↓ ↓ │ │ ┌─────────────┐ ┌─────────────┐ │ │ │ GPU Agents │ │ AIE Agents │ │ │ └─────────────┘ └─────────────┘ │ │ │ │ │ │ │通过各自的Driver │通过各自的Driver │ │ ↓ ↓ │ │ ┌─────────────┐ ┌─────────────┐ │ │ │Thunk (KFD) │ │Thunk (XDNA) │ │ │ └─────────────┘ └─────────────┘ │ └────────────────────────────────────────────────────────────┘ │ │ │ioctl │ioctl ↓ ↓ /dev/kfd /dev/accel/accel0 ┌──────────────────────────────────────┐ │ Agent (抽象基类) │ │ ┌────────────────────────────────┐ │ │ │ 成员: Driver* driver_ │ │ ← 持有Driver引用 │ │ (不拥有，由Runtime管理) │ │ │ └────────────────────────────────┘ │ │ │ │ ├─ GpuAgent (使用KfdDriver) │ │ │ - driver_->CreateQueue() │ │ │ - driver_->AllocateMemory() │ │ │ - VRAM/LDS管理 │ │ │ │ │ ├─ CpuAgent (无需驱动) │ │ │ - System Memory │ │ │ - Host Queue │ │ │ │ │ └─ AieAgent (使用XdnaDriver) │ │ - driver_->CreateQueue() │ │ - AIE特定操作 │ └──────────────────────────────────────┘ │ │拥有 ↓ ┌──────────────────────────────────────┐ │ MemoryRegion (内存区域) │ │ ├─ System Region (Fine/Coarse) │ │ ├─ Local Region (VRAM) │ │ ├─ LDS Region │ │ └─ Group Memory │ └──────────────────────────────────────┘ │ │分配 ↓ ┌──────────────────────────────────────┐ │ Memory Allocation │ │ (用户的内存指针) │ └──────────────────────────────────────┘ ┌──────────────────────────────────────┐ │ Queue (队列抽象) │ │ ├─ AqlQueue (GPU硬件队列) │ │ ├─ HostQueue (CPU软件队列) │ │ └─ InterceptQueue (调试/Trace) │ └──────────────────────────────────────┘ │ │包含 ↓ ┌──────────────────────────────────────┐ │ AQL Packet Ring Buffer │ │ [Dispatch][Barrier][Dispatch]... │ └──────────────────────────────────────┘ │ │关联 ↓ ┌──────────────────────────────────────┐ │ Signal (同步信号) │ │ ├─ DefaultSignal (共享内存) │ │ ├─ InterruptSignal (事件驱动) │ │ └─ IpcSignal (跨进程) │ └──────────────────────────────────────┘ 

2.2.2 对象生命周期

应用程序启动 ↓ hsa_init() ↓ Runtime::Acquire() ├─ 创建Driver实例 │ ├─ 根据平台选择DriverType (KFD/XDNA) │ └─ Driver::Open() 打开 /dev/kfd 或其他设备节点 │ ├─ 加载Thunk库 (通过Driver) ├─ 枚举系统节点 (通过Driver接口) │ ↓ │ 对每个节点创建Agent │ ├─ CPU Agent → System Memory Region │ └─ GPU Agent → VRAM Region + LDS Region │ (Agent持有Driver引用，所有硬件操作通过Driver) │ ├─ 建立拓扑关系 │ (Agent间的互连、NUMA亲和性) │ └─ 注册Extensions (AMD扩展、Image扩展等) 应用程序运行 ↓ hsa_queue_create() ↓ Agent::QueueCreate() ├─ 分配Ring Buffer内存 (通过Driver::AllocateMemory) ├─ 向KFD注册队列 (通过Driver::CreateQueue) └─ 创建Doorbell Signal 应用程序派发任务 ↓ 写入AQL Packet到Queue ↓ Doorbell通知GPU ↓ GPU执行Kernel ↓ Signal同步等待 应用程序退出 ↓ hsa_shut_down() ↓ Runtime::Release() ├─ 销毁所有Queue (通过Driver::DestroyQueue) ├─ 释放所有Memory (通过Driver::FreeMemory) ├─ 注销Agents ├─ Driver::Close() 关闭设备 └─ 销毁Driver实例 

2.2.3 典型调用链示例

以内存分配为例，展示各层的交互及Driver的接入：

应用程序调用 ↓ hsa_memory_allocate(region, size, &ptr) ↓ [公共API层 - hsa.h] 验证参数，查找MemoryRegion对象 ↓ [核心实现层 - MemoryRegion] MemoryRegion::Allocate(size, &ptr) ↓ [核心实现层 - Agent] Agent::AllocateMemory(size, &ptr) ├─ Agent持有 driver_ 成员 └─ 通过driver_调用底层 ↓ [核心实现层 - Driver] ◄─── Driver接入点 Driver::AllocateMemory(node_id, size, flags, &ptr) ├─ 参数转换：HSA类型 → Thunk类型 ├─ 准备HsaMemFlags结构 └─ 调用Thunk接口 ↓ [Thunk层 - libhsakmt.so] hsaKmtAllocMemory(NodeId, Size, Flags, &Address) ├─ 准备ioctl参数 └─ 执行系统调用 ↓ ioctl(fd, AMDKFD_IOC_ALLOC_MEMORY_OF_GPU, &args) ↓ [KFD内核驱动] kfd_ioctl_alloc_memory_of_gpu() ├─ 权限检查 ├─ 内存分配 └─ GPU地址映射 ↓ GPU内存分配完成，返回地址 

Driver接入的关键点：

硬件操作通过Driver：

// Agent需要操作硬件时 hsa_status_t Agent::DoHardwareOperation(){return driver_->SomeHardwareOp(node_id_,...);}

Agent持有引用：

classAgent{protected: core::Driver* driver_;// 不拥有所有权，只持有引用};

创建时接入：

// Runtime初始化时创建Driver driver_ =CreateDriver();// 根据平台选择KfdDriver/XdnaDriver// 创建Agent时传入Driver引用 agent =newGpuAgent(node_id, props, driver_);

各层职责总结：

• 公共API层：标准接口，参数校验
• 核心实现层：业务逻辑，资源管理
• Driver层：接口适配，参数转换，硬件抽象
• Thunk层：系统调用封装
• KFD驱动：硬件操作，权限控制

2.3 初始化与销毁流程

2.3.1 初始化详细流程

// 应用程序调用 hsa_status_t status =hsa_init();// ↓ 内部实现 (core/runtime/runtime.cpp)boolRuntime::Acquire(){ ScopedAcquire<KernelMutex>lock(&runtime_lock_);if(ref_count_ ==0){// 第一次初始化// 1. 检测并创建所需的Driver实例// 根据系统硬件配置，可能创建多个不同类型的Driver// 1a. 检测GPU设备，创建KFD Driverif(DetectGpuDevices()){auto kfd_driver = std::make_unique<KfdDriver>( DriverType::KFD,"/dev/kfd");if(kfd_driver->Init()== HSA_STATUS_SUCCESS && kfd_driver->Open()== HSA_STATUS_SUCCESS){ agent_drivers_.push_back(std::move(kfd_driver));}}// 1b. 检测AIE设备，创建XDNA Driver（如果存在）if(DetectAieDevices()){auto xdna_driver = std::make_unique<XdnaDriver>( DriverType::XDNA,"/dev/accel/accel0");if(xdna_driver->Init()== HSA_STATUS_SUCCESS && xdna_driver->Open()== HSA_STATUS_SUCCESS){ agent_drivers_.push_back(std::move(xdna_driver));}}// 2. 为每个Driver枚举并创建对应的Agentfor(auto& driver : agent_drivers_){ HsaSystemProperties sys_props; driver->GetSystemProperties(sys_props);// 枚举此驱动管理的所有节点for(uint32_t node =0; node < sys_props.NumNodes; node++){ HsaNodeProperties node_props; driver->GetNodeProperties(node_props, node); Agent* agent =nullptr;// 根据节点类型创建相应的Agentif(node_props.NumCPUCores >0){// CPU节点（不需要驱动） agent =new amd::CpuAgent(node, node_props,nullptr); cpu_agents_.push_back(agent);}elseif(node_props.NumFComputeCores >0){// GPU节点 - 传入KFD Driver引用 agent =new amd::GpuAgent(node, node_props, driver.get()); gpu_agents_.push_back(agent);}elseif(IsAieNode(node_props)){// AIE节点 - 传入XDNA Driver引用 agent =new amd::AieAgent(node, node_props, driver.get()); aie_agents_.push_back(agent);}if(agent){// 将Agent添加到统一索引 agents_by_node_[node]= agent;}}}// 3. 为每个Agent初始化MemoryRegion// Agent内部会通过driver_查询内存属性for(auto&[node_id, agent]: agents_by_node_){ agent->InitRegions();}// 4. 建立Agent间的拓扑关系DiscoverTopology();// 5. 加载ExtensionsLoadExtensions();// 6. 初始化工具库（Trace/Profile）LoadTools();} ref_count_++;returntrue;}

初始化流程图：

hsa_init() ↓ Runtime::Acquire() ↓ ┌────────────────────────────────────┐ │ 1. 检测硬件并创建Driver实例 │ │ ┌──────────────────────────┐ │ │ │ 检测GPU → KfdDriver │ │ → new KfdDriver("/dev/kfd") │ │ Init() + Open() │ │ agent_drivers_.push_back() │ └──────────────────────────┘ │ │ ┌──────────────────────────┐ │ │ │ 检测AIE → XdnaDriver │ │ → new XdnaDriver("/dev/accel/accel0") │ │ Init() + Open() │ │ agent_drivers_.push_back() │ └──────────────────────────┘ │ └────────────────────────────────────┘ ↓ ┌────────────────────────────────────┐ │ 2. 为每个Driver枚举设备节点 │ │ 对 agent_drivers_ 中每个driver │ │ ↓ │ │ driver->GetSystemProperties() │ │ driver->GetNodeProperties() │ └────────────────────────────────────┘ ↓ ┌────────────────────────────────────┐ │ 3. 创建Agent并关联Driver │ │ ┌─────────────────────────────┐ │ │ │ CpuAgent(node, props, nullptr) │ → cpu_agents_ │ │ （CPU不需要驱动） │ │ └─────────────────────────────┘ │ │ ┌─────────────────────────────┐ │ │ │ GpuAgent(node, props, kfd_driver) │ → gpu_agents_ │ │ （传入KFD Driver引用） │ │ └─────────────────────────────┘ │ │ ┌─────────────────────────────┐ │ │ │ AieAgent(node, props, xdna_driver) │ → aie_agents_ │ │ （传入XDNA Driver引用） │ │ └─────────────────────────────┘ │ │ 所有Agent统一索引到 │ │ agents_by_node_[node] │ └────────────────────────────────────┘ ↓ ┌────────────────────────────────────┐ │ 4. 初始化MemoryRegion │ │ agent->InitRegions() │ │ (Agent通过driver_查询内存) │ └────────────────────────────────────┘ ↓ ┌────────────────────────────────────┐ │ 5. 建立拓扑关系 │ │ - Agent间连接 │ │ - NUMA亲和性 │ └────────────────────────────────────┘ ↓ ┌────────────────────────────────────┐ │ 6. 加载Extensions │ │ - AMD扩展 │ │ - Image扩展 │ └────────────────────────────────────┘ ↓ ┌────────────────────────────────────┐ │ 7. 初始化工具 │ │ - API Trace │ │ - Profiler │ └────────────────────────────────────┘ ↓ 返回 HSA_STATUS_SUCCESS 关键设计： • agent_drivers_是vector，支持多个Driver共存 • 每个Agent持有对应类型Driver的引用 • Driver由Runtime统一管理生命周期 • 支持异构系统（GPU+AIE同时存在） 

2.3.2 销毁流程

// 应用程序调用hsa_shut_down();// ↓ 内部实现boolRuntime::Release(){ ScopedAcquire<KernelMutex>lock(&runtime_lock_);if(ref_count_ ==1){// 最后一次引用// 1. 卸载工具库UnloadTools();// 2. 卸载ExtensionsUnloadExtensions();// 3. 销毁所有Agent（自动清理Queue、Memory）for(auto agent : agents_){delete agent;} agents_.clear(); gpu_agents_.clear(); cpu_agents_.clear();// 4. 关闭KFD设备hsaKmtCloseKFD();// 5. 卸载Thunk库Unload();} ref_count_--;returntrue;}

2.3.3 引用计数机制

HSA Runtime使用引用计数支持多次初始化/销毁：

// 场景：多个库都依赖HSA Runtime// 库A初始化hsa_init();// ref_count: 0→1, 实际初始化// 库B初始化hsa_init();// ref_count: 1→2, 无操作// 库A清理hsa_shut_down();// ref_count: 2→1, 无操作// 库B清理hsa_shut_down();// ref_count: 1→0, 实际清理

2.4 Runtime配置与环境变量

HSA Runtime提供丰富的环境变量来控制行为和调试。

2.4.1 常用环境变量

2.4.2 调试相关变量

# 1. API追踪exportHSA_TOOLS_LIB=/opt/rocm/lib/libhsa-runtime-tools64.so # 输出所有HSA API调用和参数# 2. 队列调试exportHSA_QUEUE_DEBUG=1# 打印队列创建/销毁信息# 3. 内存调试exportHSA_MEMORY_DEBUG=1# 追踪内存分配/释放# 4. Signal调试exportHSA_SIGNAL_DEBUG=1# 打印Signal等待/触发# 5. 详细日志exportHSA_ENABLE_DEBUG=1exportHSA_DEBUG_LEVEL=7# 0-7，数字越大越详细

2.4.3 性能优化变量

# 1. Scratch内存预分配exportHSA_SCRATCH_MEM=512# 预分配512MB Scratch，避免运行时分配# 2. 队列大小exportHSA_QUEUE_SIZE=8192# 更大的队列可以容纳更多任务# 3. 禁用一致性检查（生产环境）exportHSA_CHECK_FLAT_SCRATCH=0exportHSA_ENABLE_DEBUG=0# 4. SDMA引擎优化exportHSA_ENABLE_SDMA=1# 优先使用硬件DMA传输

2.4.4 配置文件

部分配置可以通过配置文件 /etc/hsa/hsa.conf 设置：

# /etc/hsa/hsa.conf 示例 [runtime] # 默认队列大小 queue_size = 4096 # Scratch内存 scratch_mem_mb = 256 # 是否启用SDMA enable_sdma = true [debug] # 调试级别 0-7 level = 0 # API追踪 api_trace = false [memory] # 大页支持 use_huge_pages = true # SVM保护页 svm_guard_pages = false 

2.5 调试与追踪接口

2.5.1 API追踪工具

HSA Runtime提供了工具拦截机制，允许外部库拦截所有API调用：

// HSA定义的工具接口typedefstructhsa_api_table_s{hsa_status_t(*hsa_init_fn)(void);hsa_status_t(*hsa_shut_down_fn)(void);hsa_status_t(*hsa_agent_get_info_fn)(hsa_agent_t, hsa_agent_info_t,void*);// ... 所有HSA API的函数指针} hsa_api_table_t;// 工具库导出的函数extern"C"boolOnLoad(void* table,// Runtime的原始函数表uint64_t runtime_version,uint64_t failed_tool_count,constchar*const* failed_tool_names){// 保存原始函数 original_table =*(hsa_api_table_t*)table;// 替换为拦截函数 hsa_api_table_t* api_table =(hsa_api_table_t*)table; api_table->hsa_init_fn = InterceptHsaInit; api_table->hsa_agent_get_info_fn = InterceptHsaAgentGetInfo;// ...returntrue;}// 拦截函数示例 hsa_status_t InterceptHsaInit(void){printf("[Trace] hsa_init() called\n");auto start = std::chrono::high_resolution_clock::now(); hsa_status_t status = original_table.hsa_init_fn();auto end = std::chrono::high_resolution_clock::now();printf("[Trace] hsa_init() returned %d, took %lld us\n", status, std::chrono::duration_cast<std::chrono::microseconds>(end - start).count());return status;}

使用工具库：

# 加载追踪工具exportHSA_TOOLS_LIB=/opt/rocm/lib/libhsa-runtime-tools64.so # 运行应用 ./my_hsa_app # 输出示例：# [Trace] hsa_init() called# [Trace] hsa_init() returned 0, took 1234 us# [Trace] hsa_iterate_agents() called# [Trace] Agent 0: CPU, 16 cores# [Trace] Agent 1: GPU, 60 CUs# ...

2.5.2 内部Trace点

Runtime内部使用宏定义追踪点：

// core/inc/runtime.h#ifdefHSA_TRACE#defineTRACE_RUNTIME(fmt,...)\fprintf(stderr,"[HSA] "fmt "\n",##__VA_ARGS__)#else#defineTRACE_RUNTIME(...)#endif// 使用示例TRACE_RUNTIME("Agent %u initialized: %s", agent->node_id(), agent->device_type()== Agent::kAmdGpuDevice ?"GPU":"CPU");

2.5.3 Profiling接口

// HSA提供的时间戳查询 hsa_status_t hsa_system_get_info( HSA_SYSTEM_INFO_TIMESTAMP,// 获取硬件时间戳&timestamp);// 高精度时间戳频率hsa_system_get_info( HSA_SYSTEM_INFO_TIMESTAMP_FREQUENCY,&frequency);// 计算时间double seconds =(end_timestamp - start_timestamp)/(double)frequency;

2.5.4 AMD扩展调试功能

// AMD扩展：设置异常策略hsa_amd_signal_async_handler( signal, HSA_SIGNAL_CONDITION_NE,0, exception_handler,NULL);// AMD扩展：内存追踪hsa_amd_memory_pool_get_info( pool, HSA_AMD_MEMORY_POOL_INFO_SIZE,&total_size);hsa_amd_memory_pool_get_info( pool, HSA_AMD_MEMORY_POOL_INFO_ALLOC_MAX_SIZE,&max_alloc);

🎯 本章总结

核心要点回顾

1. 三层架构设计
   • 公共API层：C接口，符合HSA规范
   • 核心实现层：C++实现，面向对象设计
   • 驱动抽象层：Thunk + KFD，硬件访问
2. 主要组件关系
   • Runtime管理所有Agent和MemoryRegion
   • Agent拥有Queue和内存资源
   • Queue通过Signal实现同步
   • 清晰的生命周期管理
3. 初始化流程
   • 加载Thunk库 → 枚举设备 → 创建Agent
   • 初始化内存区域 → 建立拓扑 → 加载扩展
   • 支持引用计数，多次初始化
4. 配置与调试
   • 丰富的环境变量控制行为
   • 工具拦截机制实现API追踪
   • 内置Trace点和Profiling支持

关键数据结构

// Runtime核心类classRuntime{static Runtime* runtime_singleton_; std::vector<Agent*> agents_; std::vector<MemoryRegion*> regions_; core::Driver* driver_;// 驱动实例int ref_count_;};// Driver抽象（位于核心实现层）classDriver{ DriverType kernel_driver_type_;// KFD/XDNA/VIRTIO std::string devnode_name_;// 设备节点路径int fd_;// 设备文件描述符// 提供所有硬件操作接口};// Agent抽象classAgent{uint32_t node_id_; DeviceType device_type_; std::vector<MemoryRegion*> regions_; core::Driver* driver_;// 持有Driver引用};

思考题

1. 为什么HSA Runtime使用三层架构，而不是直接在API层操作硬件？
   • 可移植性：API层与硬件解耦，同一套API支持不同厂商
   • 可维护性：业务逻辑(核心层)与硬件操作(驱动层)分离
   • 稳定性：API层ABI稳定，内部实现可以优化而不影响应用
   • 安全性：驱动层在内核态，提供权限控制和资源隔离
2. Runtime为什么使用引用计数而不是简单的初始化/销毁？
   • 多库场景：应用可能链接多个都依赖HSA的库（如HIP、OpenCL）
   • 避免冲突：各库独立调用init/shutdown不会互相干扰
   • 资源共享：多个组件可以共用同一个Runtime实例
   • 线程安全：引用计数配合锁保证多线程安全
3. 工具拦截机制如何实现零开销的可选追踪？
   • 按需加载：只有设置HSA_TOOLS_LIB时才加载工具
   • 函数指针替换：拦截层只是简单的函数调用，无额外开销
   • 编译优化：Release版本可以移除所有Trace宏
   • 惰性激活：工具库可以根据条件动态开关追踪
4. 为什么需要Driver类，而不是直接调用Thunk接口？
   • 面向对象封装：Driver提供C++接口，比C接口更易用和安全
   • 多驱动支持：通过虚函数支持KFD、XDNA、Virtio等不同后端
   • 资源管理：Driver类管理设备文件描述符生命周期
   • 接口统一：上层组件（Runtime、Agent）无需关心底层驱动差异
   • 错误处理：统一的异常转换和错误码映射
   • 测试友好：可以Mock Driver类进行单元测试

🔗 导航

• 上一章: 01-什么是HSA与异构计算
• 下一章: 第03章：HSA编程模型基础
• 返回目录: 00-HSA-Runtime学习文档目录

📚 延伸阅读

ROCm rocr-libhsakmt分析系列0：HSA(异构系统架构)驱动概览

相关主题

• 设计模式：《Design Patterns》(GoF)
• 系统架构：《Software Architecture Patterns》
• 插件系统：《Plugin Architecture》
• 驱动开发：Linux Device Drivers (3rd Edition)