2026年03月专栏上线了本系列在libhsakmt分析的基础上向上扩展到runtime层的实现。每周放送57篇。专栏内容简介本系列文档旨在帮助开发者系统学习 AMD ROCm HSA Runtime异构系统架构运行时涵盖从基础概念到高级特性的完整技术栈。HSA Runtime 是 ROCm 软件栈的核心组件提供了 GPU 设备管理、内存管理、任务调度、信号同步等关键功能。主要模块: Agent、Queue、Signal、Memory、Loader难度分级: 基础 | 中级 | 高级 学习路线图第一部分HSA基础 (1-5章) ↓ 第二部分核心组件 (6-12章) ↓ 第三部分内存管理 (13-17章) ↓ 第四部分任务调度 (18-22章) ↓ 第五部分高级特性 (23-28章)第一部分HSA基础篇 建立异构计算和HSA架构的基础认知01-什么是HSA与异构计算1.1 异构计算的演进历史1.2 HSA规范的诞生背景1.3 HSA的核心设计理念统一内存空间用户态队列平台无关性1.4 HSA与CUDA/OpenCL的对比1.5 ROCm软件栈中的HSA定位02-HSA-Runtime架构概览2.1 Runtime层次结构公共API层 (inc/hsa.h)核心实现层 (core/runtime/)驱动抽象层 (core/driver/)2.2 主要组件关系图2.3 初始化与销毁流程2.4 Runtime配置与环境变量2.5 调试与追踪接口03-HSA编程模型基础3.1 Host与Device概念3.2 执行模型Kernel执行Wavefront与Work-itemWork-group组织3.3 内存模型Global MemoryGroup/Local MemoryPrivate Memory3.4 同步原语SignalBarrierFence3.5 HSA程序生命周期04-开发环境搭建4.1 ROCm安装与配置4.2 HSA Runtime编译4.3 示例程序运行4.4 调试工具介绍rocgdbrocprofHSA_DEBUG环境变量4.5 常见问题排查05-HSA-API快速入门5.1 Runtime初始化hsa_init()hsa_shut_down()5.2 设备发现与查询hsa_iterate_agents()hsa_agent_get_info()5.3 内存分配hsa_memory_allocate()hsa_memory_free()5.4 队列创建与管理hsa_queue_create()hsa_queue_destroy()5.5 完整示例代码解析第二部分核心组件篇 深入理解Runtime的核心数据结构和机制[06-Agent设备抽象]](https://gpuers.blog.csdn.net/article/details/156336927)6.1 Agent概念与类型CPU AgentGPU AgentAIE Agent (AI Engine)6.2 Agent类层次结构core::Agent(基类)├── CpuAgent ├── GpuAgent/GpuAgentInt └── AieAgent6.3 Agent属性查询6.4 设备拓扑与互连6.5 Agent生命周期管理07-GPU-Agent深度解析7.1 GpuAgent初始化流程7.2 GPU硬件能力查询Compute Units (CU)Wavefront大小ISA支持7.3 GPU内存区域管理7.4 DMA引擎 (SDMA/Blit)7.5 GPU特有功能Shader可见缓存Trap HandlerProfiling支持08-CPU-Agent与混合调度8.1 CpuAgent实现特点8.2 CPU内存访问8.3 CPU-GPU协同8.4 NUMA亲和性8.5 Host Queue机制09-Driver驱动抽象层9.1 Driver接口设计9.2 KFD (Kernel Fusion Driver) 集成libhsakmt封装Thunk层实现9.3 设备发现与枚举9.4 内存管理接口9.5 队列管理接口10-Topology拓扑发现10.1 系统拓扑解析10.2 NUMA节点映射10.3 设备互连关系PCIe链路XGMI/Infinity Fabric10.4 Link属性查询10.5 拓扑信息缓存11-ISA指令集架构11.1 ISA抽象与版本管理11.2 GCN/RDNA/CDNA架构差异11.3 ISA能力查询11.4 代码对象加载11.5 向前兼容性处理12-Runtime生命周期管理12.1 初始化序列详解Runtime::Acquire()流程 ├── 加载Driver ├── 枚举Agents ├── 初始化MemoryRegions ├── 建立拓扑关系 └── 注册Extension12.2 延迟初始化 (Lazy Init)12.3 多线程安全保证12.4 资源清理与释放12.5 异常处理机制第三部分内存管理篇 掌握HSA的统一内存模型与分配策略13-Memory-Region内存区域13.1 MemoryRegion类设计13.2 内存区域类型System Memory (Host)Local Memory (VRAM)LDS (Local Data Share)GDS (Global Data Share)13.3 内存属性Fine-grained vs Coarse-grainedCoherency特性13.4 跨设备访问能力13.5 内存池 (Pool) 管理14-内存分配与释放14.1 分配流程14.2 内存对齐要求14.3 大页支持14.4 VRAM管理策略14.5 内存碎片处理15-统一内存与一致性15.1 HSA统一地址空间15.2 Fine-grained内存硬件缓存一致性原子操作支持15.3 Coarse-grained内存显式同步需求Fence与Acquire/Release15.4 内存一致性模型15.5 跨Agent内存访问16-IPC共享内存16.1 进程间内存共享16.2 IPC Handle机制16.3 GPU Direct RDMA16.4 多进程场景16.5 安全性考虑17-DMA数据传输17.1 DMA传输类型Host ↔ DeviceDevice ↔ DevicePeer-to-Peer17.2 SDMA引擎使用17.3 Blit Kernel备用方案17.4 异步拷贝与信号17.5 性能优化技巧第四部分任务调度篇 理解AQL队列与内核调度机制18-Queue队列机制18.1 HSA队列模型18.2 Queue类层次18.3 用户态队列优势18.4 队列创建参数18.5 队列状态管理19-AQL协议详解19.1 AQL (Architected Queueing Language)19.2 AQL数据包格式19.3 Packet类型Kernel DispatchBarrierAgent Dispatch19.4 Doorbell机制19.5 Packet提交流程20-Kernel-Dispatch内核派发20.1 Dispatch Packet结构20.2 工作项组织Grid维度Workgroup维度Wavefront分配20.3 内核参数传递20.4 内存段配置Private ScratchGroup LDS20.5 完整派发示例21-Scratch内存管理21.1 Scratch的作用私有栈空间Spill寄存器21.2 动态Scratch分配21.3 Scratch Cache优化21.4 大Scratch处理21.5 性能影响分析22-Barrier与同步22.1 Barrier PacketAND语义OR语义22.2 依赖信号管理22.3 多队列同步22.4 内存FenceAcquire FenceRelease Fence22.5 同步开销优化第五部分高级特性篇 掌握信号、加载器、异常处理等高级功能23-Signal信号系统23.1 Signal设计理念23.2 Signal类层次23.3 SharedSignal结构23.4 信号值操作原子读写CAS (Compare-And-Swap)算术操作23.5 信号等待策略24-Signal等待机制24.1 Wait条件判断24.2 等待实现策略Active Polling (忙等)MWAITX指令Event驱动 (中断)24.3 超时处理24.4 多信号等待24.5 性能权衡25-Loader代码加载器25.1 HSA代码对象格式ELF结构.note段.text段 (代码).rodata段 (常量)25.2 代码对象加载流程25.3 符号解析与重定位25.4 多设备代码管理25.5 动态加载与卸载26-Trap-Handler异常处理26.1 GPU异常类型Memory FaultIllegal InstructionTrap指令26.2 Trap Handler安装26.3 异常信息收集26.4 调试支持26.5 Wave保存与恢复27-Extension扩展机制27.1 HSA Extension系统27.2 AMD专有扩展27.3 Image扩展27.4 Finalizer扩展27.5 Profile扩展27.6 扩展注册与查询28-性能分析与调试28.1 Profiling支持时间戳查询性能计数器28.2 API Trace调用拦截参数记录28.3 调试环境变量28.4 Core Dump分析28.5 常见问题排查学习路径建议 快速入门路径 (1-2周)目标: 能够编写基本的HSA程序基础理解(第一部分: 01-05章)重点: 01, 02, 05章跳过: 详细的编程模型细节核心组件(第二部分: 06, 09, 12章)理解Agent和Runtime初始化浏览Driver抽象任务调度(第四部分: 18, 19, 20章)掌握Queue和AQL基础学习Kernel派发实践项目编译运行示例代码实现简单的向量加法 系统掌握路径 (3-4周)目标: 深入理解Runtime实现完整基础(第一部分: 全部)核心组件(第二部分: 全部)重点研读GPU Agent实现内存管理(第三部分: 全部)理解统一内存模型任务调度(第四部分: 全部)掌握AQL协议细节高级特性(第五部分: 23, 24, 25章)Signal和Loader机制 问题驱动路径目标: 针对性解决具体问题性能优化: 07→17→21→28章内存问题: 13→14→15→16章调度问题: 18→19→20→22章调试排查: 04→26→28章扩展开发: 09→25→27章代码阅读顺序建议第一阶段公共API理解# 1. HSA公共头文件inc/hsa.h# 核心API定义inc/hsa_ext_amd.h# AMD扩展API# 2. 基础数据结构inc/amd_hsa_signal.h# Signal结构inc/amd_hsa_queue.h# Queue结构第二阶段核心类设计# 3. 核心抽象类core/inc/agent.h# Agent基类core/inc/queue.h# Queue基类core/inc/signal.h# Signal基类core/inc/memory_region.h# MemoryRegion基类# 4. Runtime主控core/inc/runtime.h# Runtime类定义第三阶段具体实现# 5. Agent实现core/runtime/amd_gpu_agent.cpp# GPU Agentcore/runtime/amd_cpu_agent.cpp# CPU Agent# 6. Queue实现core/runtime/amd_aql_queue.cpp# AQL队列core/runtime/queue.cpp# Queue基础# 7. Signal实现core/runtime/signal.cpp# Signal实现core/runtime/default_signal.cpp# 默认信号core/runtime/interrupt_signal.cpp# 中断信号# 8. Runtime初始化core/runtime/runtime.cpp# Runtime实现第四阶段专题深入# 9. 内存管理core/runtime/amd_memory_region.cpp core/runtime/amd_blit_sdma.cpp# 10. 代码加载loader/executable.cpp core/runtime/amd_hsa_loader.cpp# 11. 扩展功能core/runtime/hsa_ext_amd.cpp术语速查表术语全称/解释说明HSAHeterogeneous System Architecture异构系统架构AQLArchitected Queueing LanguageHSA队列协议Agent-HSA中的计算设备抽象Queue-用户态命令队列Signal-同步信号对象Packet-AQL命令包Wavefront-GPU执行的SIMD单元Work-item-单个线程Work-group-线程组KFDKernel Fusion Driver内核驱动Thunk-KFD用户态封装层SDMASystem DMA系统DMA引擎LDSLocal Data Share工作组共享内存GDSGlobal Data Share全局共享数据Scratch-私有栈/Spill空间Doorbell-队列提交通知机制Trap Handler-GPU异常处理器Finalizer-代码生成与优化关键数据结构速查Agent相关// 基类classcore::Agent{uint32_tnode_id_;DeviceType device_type_;hsa_agent_t public_handle_;// 虚函数DmaCopy, DmaCopyAsync, ...};// GPU实现classGpuAgent:publicAgent{uint32_tnode_id_;HsaNodeProperties properties_;std::vectorMemoryRegion*regions_;// SDMA, Blit, ISA, ...};Queue相关// AQL队列classAqlQueue:publicQueue{amd_queue_t amd_queue_;// 硬件队列结构uint64_t*ring_buf_;// 环形缓冲区uint32_tring_size_;Signal*doorbell_signal_;};Signal相关// 共享信号结构structSharedSignal{amd_signal_t amd_signal;// 包含value, kind等Signal*core_signal;uint64_ttimestamps[];};核心API速查初始化hsa_status_thsa_init(void);hsa_status_thsa_shut_down(void);Agent管理hsa_status_thsa_iterate_agents(hsa_status_t(*callback)(hsa_agent_t,void*),void*data);hsa_status_thsa_agent_get_info(hsa_agent_t agent,hsa_agent_info_t attribute,void*value);内存管理hsa_status_thsa_memory_allocate(hsa_region_t region,size_t size,void**ptr);hsa_status_thsa_memory_free(void*ptr);hsa_status_thsa_memory_copy(void*dst,constvoid*src,size_t size);队列管理hsa_status_thsa_queue_create(hsa_agent_t agent,uint32_tsize,hsa_queue_type_t type,void(*callback)(hsa_status_t,hsa_queue_t*,void*),void*data,uint32_tprivate_segment_size,uint32_tgroup_segment_size,hsa_queue_t**queue);hsa_status_thsa_queue_destroy(hsa_queue_t*queue);uint64_thsa_queue_add_write_index_relaxed(consthsa_queue_t*queue,uint64_tvalue);信号管理hsa_status_thsa_signal_create(hsa_signal_value_t initial_value,uint32_tnum_consumers,consthsa_agent_t*consumers,hsa_signal_t*signal);voidhsa_signal_destroy(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_ttimeout_hint,hsa_wait_state_t wait_state_hint);贡献指南本文档系列持续完善中欢迎贡献发现错误: 提交Issue说明问题提供示例: 特别欢迎实际代码案例提出建议: 讨论学习路径优化参考资源官方文档HSA SpecificationROCm DocumentationHSA Runtime API Reference
ROCm rocr-runtime 库详细解析目录
2026年03月专栏上线了本系列在libhsakmt分析的基础上向上扩展到runtime层的实现。每周放送57篇。专栏内容简介本系列文档旨在帮助开发者系统学习 AMD ROCm HSA Runtime异构系统架构运行时涵盖从基础概念到高级特性的完整技术栈。HSA Runtime 是 ROCm 软件栈的核心组件提供了 GPU 设备管理、内存管理、任务调度、信号同步等关键功能。主要模块: Agent、Queue、Signal、Memory、Loader难度分级: 基础 | 中级 | 高级 学习路线图第一部分HSA基础 (1-5章) ↓ 第二部分核心组件 (6-12章) ↓ 第三部分内存管理 (13-17章) ↓ 第四部分任务调度 (18-22章) ↓ 第五部分高级特性 (23-28章)第一部分HSA基础篇 建立异构计算和HSA架构的基础认知01-什么是HSA与异构计算1.1 异构计算的演进历史1.2 HSA规范的诞生背景1.3 HSA的核心设计理念统一内存空间用户态队列平台无关性1.4 HSA与CUDA/OpenCL的对比1.5 ROCm软件栈中的HSA定位02-HSA-Runtime架构概览2.1 Runtime层次结构公共API层 (inc/hsa.h)核心实现层 (core/runtime/)驱动抽象层 (core/driver/)2.2 主要组件关系图2.3 初始化与销毁流程2.4 Runtime配置与环境变量2.5 调试与追踪接口03-HSA编程模型基础3.1 Host与Device概念3.2 执行模型Kernel执行Wavefront与Work-itemWork-group组织3.3 内存模型Global MemoryGroup/Local MemoryPrivate Memory3.4 同步原语SignalBarrierFence3.5 HSA程序生命周期04-开发环境搭建4.1 ROCm安装与配置4.2 HSA Runtime编译4.3 示例程序运行4.4 调试工具介绍rocgdbrocprofHSA_DEBUG环境变量4.5 常见问题排查05-HSA-API快速入门5.1 Runtime初始化hsa_init()hsa_shut_down()5.2 设备发现与查询hsa_iterate_agents()hsa_agent_get_info()5.3 内存分配hsa_memory_allocate()hsa_memory_free()5.4 队列创建与管理hsa_queue_create()hsa_queue_destroy()5.5 完整示例代码解析第二部分核心组件篇 深入理解Runtime的核心数据结构和机制[06-Agent设备抽象]](https://gpuers.blog.csdn.net/article/details/156336927)6.1 Agent概念与类型CPU AgentGPU AgentAIE Agent (AI Engine)6.2 Agent类层次结构core::Agent(基类)├── CpuAgent ├── GpuAgent/GpuAgentInt └── AieAgent6.3 Agent属性查询6.4 设备拓扑与互连6.5 Agent生命周期管理07-GPU-Agent深度解析7.1 GpuAgent初始化流程7.2 GPU硬件能力查询Compute Units (CU)Wavefront大小ISA支持7.3 GPU内存区域管理7.4 DMA引擎 (SDMA/Blit)7.5 GPU特有功能Shader可见缓存Trap HandlerProfiling支持08-CPU-Agent与混合调度8.1 CpuAgent实现特点8.2 CPU内存访问8.3 CPU-GPU协同8.4 NUMA亲和性8.5 Host Queue机制09-Driver驱动抽象层9.1 Driver接口设计9.2 KFD (Kernel Fusion Driver) 集成libhsakmt封装Thunk层实现9.3 设备发现与枚举9.4 内存管理接口9.5 队列管理接口10-Topology拓扑发现10.1 系统拓扑解析10.2 NUMA节点映射10.3 设备互连关系PCIe链路XGMI/Infinity Fabric10.4 Link属性查询10.5 拓扑信息缓存11-ISA指令集架构11.1 ISA抽象与版本管理11.2 GCN/RDNA/CDNA架构差异11.3 ISA能力查询11.4 代码对象加载11.5 向前兼容性处理12-Runtime生命周期管理12.1 初始化序列详解Runtime::Acquire()流程 ├── 加载Driver ├── 枚举Agents ├── 初始化MemoryRegions ├── 建立拓扑关系 └── 注册Extension12.2 延迟初始化 (Lazy Init)12.3 多线程安全保证12.4 资源清理与释放12.5 异常处理机制第三部分内存管理篇 掌握HSA的统一内存模型与分配策略13-Memory-Region内存区域13.1 MemoryRegion类设计13.2 内存区域类型System Memory (Host)Local Memory (VRAM)LDS (Local Data Share)GDS (Global Data Share)13.3 内存属性Fine-grained vs Coarse-grainedCoherency特性13.4 跨设备访问能力13.5 内存池 (Pool) 管理14-内存分配与释放14.1 分配流程14.2 内存对齐要求14.3 大页支持14.4 VRAM管理策略14.5 内存碎片处理15-统一内存与一致性15.1 HSA统一地址空间15.2 Fine-grained内存硬件缓存一致性原子操作支持15.3 Coarse-grained内存显式同步需求Fence与Acquire/Release15.4 内存一致性模型15.5 跨Agent内存访问16-IPC共享内存16.1 进程间内存共享16.2 IPC Handle机制16.3 GPU Direct RDMA16.4 多进程场景16.5 安全性考虑17-DMA数据传输17.1 DMA传输类型Host ↔ DeviceDevice ↔ DevicePeer-to-Peer17.2 SDMA引擎使用17.3 Blit Kernel备用方案17.4 异步拷贝与信号17.5 性能优化技巧第四部分任务调度篇 理解AQL队列与内核调度机制18-Queue队列机制18.1 HSA队列模型18.2 Queue类层次18.3 用户态队列优势18.4 队列创建参数18.5 队列状态管理19-AQL协议详解19.1 AQL (Architected Queueing Language)19.2 AQL数据包格式19.3 Packet类型Kernel DispatchBarrierAgent Dispatch19.4 Doorbell机制19.5 Packet提交流程20-Kernel-Dispatch内核派发20.1 Dispatch Packet结构20.2 工作项组织Grid维度Workgroup维度Wavefront分配20.3 内核参数传递20.4 内存段配置Private ScratchGroup LDS20.5 完整派发示例21-Scratch内存管理21.1 Scratch的作用私有栈空间Spill寄存器21.2 动态Scratch分配21.3 Scratch Cache优化21.4 大Scratch处理21.5 性能影响分析22-Barrier与同步22.1 Barrier PacketAND语义OR语义22.2 依赖信号管理22.3 多队列同步22.4 内存FenceAcquire FenceRelease Fence22.5 同步开销优化第五部分高级特性篇 掌握信号、加载器、异常处理等高级功能23-Signal信号系统23.1 Signal设计理念23.2 Signal类层次23.3 SharedSignal结构23.4 信号值操作原子读写CAS (Compare-And-Swap)算术操作23.5 信号等待策略24-Signal等待机制24.1 Wait条件判断24.2 等待实现策略Active Polling (忙等)MWAITX指令Event驱动 (中断)24.3 超时处理24.4 多信号等待24.5 性能权衡25-Loader代码加载器25.1 HSA代码对象格式ELF结构.note段.text段 (代码).rodata段 (常量)25.2 代码对象加载流程25.3 符号解析与重定位25.4 多设备代码管理25.5 动态加载与卸载26-Trap-Handler异常处理26.1 GPU异常类型Memory FaultIllegal InstructionTrap指令26.2 Trap Handler安装26.3 异常信息收集26.4 调试支持26.5 Wave保存与恢复27-Extension扩展机制27.1 HSA Extension系统27.2 AMD专有扩展27.3 Image扩展27.4 Finalizer扩展27.5 Profile扩展27.6 扩展注册与查询28-性能分析与调试28.1 Profiling支持时间戳查询性能计数器28.2 API Trace调用拦截参数记录28.3 调试环境变量28.4 Core Dump分析28.5 常见问题排查学习路径建议 快速入门路径 (1-2周)目标: 能够编写基本的HSA程序基础理解(第一部分: 01-05章)重点: 01, 02, 05章跳过: 详细的编程模型细节核心组件(第二部分: 06, 09, 12章)理解Agent和Runtime初始化浏览Driver抽象任务调度(第四部分: 18, 19, 20章)掌握Queue和AQL基础学习Kernel派发实践项目编译运行示例代码实现简单的向量加法 系统掌握路径 (3-4周)目标: 深入理解Runtime实现完整基础(第一部分: 全部)核心组件(第二部分: 全部)重点研读GPU Agent实现内存管理(第三部分: 全部)理解统一内存模型任务调度(第四部分: 全部)掌握AQL协议细节高级特性(第五部分: 23, 24, 25章)Signal和Loader机制 问题驱动路径目标: 针对性解决具体问题性能优化: 07→17→21→28章内存问题: 13→14→15→16章调度问题: 18→19→20→22章调试排查: 04→26→28章扩展开发: 09→25→27章代码阅读顺序建议第一阶段公共API理解# 1. HSA公共头文件inc/hsa.h# 核心API定义inc/hsa_ext_amd.h# AMD扩展API# 2. 基础数据结构inc/amd_hsa_signal.h# Signal结构inc/amd_hsa_queue.h# Queue结构第二阶段核心类设计# 3. 核心抽象类core/inc/agent.h# Agent基类core/inc/queue.h# Queue基类core/inc/signal.h# Signal基类core/inc/memory_region.h# MemoryRegion基类# 4. Runtime主控core/inc/runtime.h# Runtime类定义第三阶段具体实现# 5. Agent实现core/runtime/amd_gpu_agent.cpp# GPU Agentcore/runtime/amd_cpu_agent.cpp# CPU Agent# 6. Queue实现core/runtime/amd_aql_queue.cpp# AQL队列core/runtime/queue.cpp# Queue基础# 7. Signal实现core/runtime/signal.cpp# Signal实现core/runtime/default_signal.cpp# 默认信号core/runtime/interrupt_signal.cpp# 中断信号# 8. Runtime初始化core/runtime/runtime.cpp# Runtime实现第四阶段专题深入# 9. 内存管理core/runtime/amd_memory_region.cpp core/runtime/amd_blit_sdma.cpp# 10. 代码加载loader/executable.cpp core/runtime/amd_hsa_loader.cpp# 11. 扩展功能core/runtime/hsa_ext_amd.cpp术语速查表术语全称/解释说明HSAHeterogeneous System Architecture异构系统架构AQLArchitected Queueing LanguageHSA队列协议Agent-HSA中的计算设备抽象Queue-用户态命令队列Signal-同步信号对象Packet-AQL命令包Wavefront-GPU执行的SIMD单元Work-item-单个线程Work-group-线程组KFDKernel Fusion Driver内核驱动Thunk-KFD用户态封装层SDMASystem DMA系统DMA引擎LDSLocal Data Share工作组共享内存GDSGlobal Data Share全局共享数据Scratch-私有栈/Spill空间Doorbell-队列提交通知机制Trap Handler-GPU异常处理器Finalizer-代码生成与优化关键数据结构速查Agent相关// 基类classcore::Agent{uint32_tnode_id_;DeviceType device_type_;hsa_agent_t public_handle_;// 虚函数DmaCopy, DmaCopyAsync, ...};// GPU实现classGpuAgent:publicAgent{uint32_tnode_id_;HsaNodeProperties properties_;std::vectorMemoryRegion*regions_;// SDMA, Blit, ISA, ...};Queue相关// AQL队列classAqlQueue:publicQueue{amd_queue_t amd_queue_;// 硬件队列结构uint64_t*ring_buf_;// 环形缓冲区uint32_tring_size_;Signal*doorbell_signal_;};Signal相关// 共享信号结构structSharedSignal{amd_signal_t amd_signal;// 包含value, kind等Signal*core_signal;uint64_ttimestamps[];};核心API速查初始化hsa_status_thsa_init(void);hsa_status_thsa_shut_down(void);Agent管理hsa_status_thsa_iterate_agents(hsa_status_t(*callback)(hsa_agent_t,void*),void*data);hsa_status_thsa_agent_get_info(hsa_agent_t agent,hsa_agent_info_t attribute,void*value);内存管理hsa_status_thsa_memory_allocate(hsa_region_t region,size_t size,void**ptr);hsa_status_thsa_memory_free(void*ptr);hsa_status_thsa_memory_copy(void*dst,constvoid*src,size_t size);队列管理hsa_status_thsa_queue_create(hsa_agent_t agent,uint32_tsize,hsa_queue_type_t type,void(*callback)(hsa_status_t,hsa_queue_t*,void*),void*data,uint32_tprivate_segment_size,uint32_tgroup_segment_size,hsa_queue_t**queue);hsa_status_thsa_queue_destroy(hsa_queue_t*queue);uint64_thsa_queue_add_write_index_relaxed(consthsa_queue_t*queue,uint64_tvalue);信号管理hsa_status_thsa_signal_create(hsa_signal_value_t initial_value,uint32_tnum_consumers,consthsa_agent_t*consumers,hsa_signal_t*signal);voidhsa_signal_destroy(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_ttimeout_hint,hsa_wait_state_t wait_state_hint);贡献指南本文档系列持续完善中欢迎贡献发现错误: 提交Issue说明问题提供示例: 特别欢迎实际代码案例提出建议: 讨论学习路径优化参考资源官方文档HSA SpecificationROCm DocumentationHSA Runtime API Reference
