SDK Release Note(v2.1)

更新时间:
复制为 MD 格式

1. 亮点

针对真武810E SDK V2.1 主要兼容支持了 CUDA13.0 APIs,Triton3.5.x 等,用户可以根据自己的需求选择升级使用。

2. 版本概述

T-Head SAIL SDK 是平头哥独立自主开发的 AI 软件栈,拥有自主可控的软件知识产权,T-Head SAIL 软件生态设计上由应用层、接口层和 SDK 层组成,具备统一的编程接口,支持平头哥自研软件生态。用户既可以基于 T-Head SAIL APIs 开发应用, 也可以很方便地兼容基于 CUDA APIs 开发的应用程序。

本文重点介绍 PPU SDK 软件栈的主要功能和软件生态,如果您已经拥有 PPU 硬件,想快速体验的话,请参照 SDK快速入门,我们提供了多种快速上手体验方式供您选择。

2.1 支持硬件平台

产品

支持状态

真武810E

2.2 软件栈介绍

image.png

2.3 核心组件

组件名称

概要说明

Firmware

PPU 固件

KMD

PPU 内核驱动

UMD / HGGC

PPU 用户态驱动和运行时

Compiler

PPU 编译器工具链

Acompute

PPU 计算加速库

Acext

PPU 量化加速库

PCCL

PPU 通信加速库

PPU SMI

PPU 设备管理工具

PPU DCGM

PPU 在线监控工具

Asight Systems

PPU 性能分析工具

Asight Compute

PPU 性能分析工具

PPU GDB

PPU 调试工具

PPU MemCheck

PPU Sanitizer工具

PPU hgobjdump

PPU Device Binary工具

Holmes Inference Engine

PPU Holmes 推理加速引擎
Holmes-LLM

CUDA SDK Wrapper

CUDA API 兼容库

2.4 主要功能

2.4.1 支持产品

  • 真武810E 覆盖训练、推理场景。

2.4.2 Firmware

  • 支持二进制包安装(rpm 或 deb),详情请参见Firmware安装指南

  • 支持动态电源管理功能,缺省模式下PPU固件会根据实时的工作负载、温度和功耗等信息动态调节核心频率(200MHZ~最大工作频率)和工作电压。

  • 支持 PPU 锁频功能,详细用法请参考设备管理工具PPU-SMI (v2.1)

  • 支持固件安全签名和固件双备份功能,确保固件内容的安全性和可靠性。

  • 支持电源芯片固件升级功能。

  • 支持 BMC 带外管理功能,包括设备状态监控和固件带外升级等。

2.4.3 内核驱动

  • 支持二进制包安装(rpmdeb)和runfile包安装两种方式,用户可根据需要自由选择,详情请参见KMD安装指南

  • 支持内核驱动和PPU SDK解耦,内核驱动和PPU SDK在前后三个版本间可兼容使用。

  • 支持单机单卡、单机16卡(ICN互联)、多机多卡(ICN互联)和多机多卡(GDR)多种灵活的机器形态,使用GDR功能之前需确保系统已安装alixpu-peermem内核模块(随PPU内核驱动一起发布)。

  • 支持PPU故障上报和故障处理,PPU故障码详细介绍请参考PPU XID定义,ECC故障详细处理流程请参考PPU ECC处理流程(v2.1)

  • 支持auto-reset功能,驱动在检测到 kill overtime 或 cp invalid cmd 错误之后会自动进行 PPU 设备复位,无需用户手动复位即可恢复正常。该功能默认打开,用户可通过ppu-smi查询状态或关闭该功能。

  • 支持MPS(Multi Pipe Service)功能,详细使用说明可参考MPS使用指南(v2.1)

  • 支持 MPS 模式下小模型和大模型任务的潮汐场景使用。

  • 支持nvml GPM 指标采集功能,包括:sm 利用率,sm occupancy,tensor core利用率,显存带宽利用率,L1/L2 cache命中率,pcie 读写速率和 icn link 读写速率。

  • 支持整卡直通虚拟化功能,用户可以在宿主机上解绑PPU驱动直通进虚拟机内使用,此时需要注意不同VM之间的ICN隔离,详细用法请参考整卡直通虚拟化ICN隔离指南

  • 支持MIG (Multi Instance GPU)多实例功能,PPU最多支持8个实例,详细用法请参考MIG使用指南(v2.1)容器隔离使用指南,需注意使用MIG功能时ICN互联功能不可使用。

  • 支持 SRIOV 虚拟化功能,PPU最多支持8VF (Virtual Function),详细用法请参考虚拟化使用指南(v2.1),需注意SRIOVICN互联同时开启时,只允许一个VF拥有ICN互联能力。

  • 支持 SRIOV虚拟化模式下宿主机驱动的热升级功能,即可以在虚拟机内业务运行的同时对宿主机PPU驱动进行更新。

  • 支持16卡真武810E设备按照模组编号排序以获取更好的互联性能,详情可参考PPU多卡ID映射

  • 支持通过ppu-smi工具查询Xid关键错误信息功能。

  • 支持ppu-smi显示容器内的pid而不是hostpid。

  • 支持Linux系统Hibernate&Resume电源管理特性。

  • 支持CRIU (Checkpoint/Restore In Userspace) PPU状态保存/恢复功能。

2.4.4 用户态驱动和运行时

  • 兼容绝大多数cuda runtime api (cudaXXX) 和cuda driver api (cuXXX)

  • 支持发生xid 896错误时输出更多错误日志方便定位问题。

  • 支持stream memory opertaion v2版本API。

  • 支持graph managementedge data相关API。

  • 支持graph managementbatch memory op node相关API。

  • 支持texture基本功能。

  • 支持cuCtxCreate_v3 api。

  • 支持在发生异常之后正确输出device code printf的内容。

  • 支持HGGC_EXCLUSIVE_STREAMS环境变量指定stream group映射到不同的硬件队列。

  • 支持cudaLaunchKernelExC接口。

  • 支持通过asys工具采集stream对应的hardware channel信息。

  • 支持AD扩展API接口。

  • 支持cudaGetDriverEntryPointByVersion接口。

  • 支持cudaLaunchAttributeLaunchCompletionEvent特性。

  • 支持cuda 13.0。

  • 支持cuda graph conditional node特性。

  • 支持texture reference接口。

  • virtual memory management接口增加host memory类型支持。

  • 支持CRIU (Checkpoint/Restore In Userspace) PPU状态保存/恢复功能,需搭配v2.1版本内核驱动使用,详细用法可参考CRIU用户指南

2.4.5 编译器

  • 全面支持多代芯片产品的编译,可以通过option来指定具体架构,默认情况下为全系支持的混合fatbin产物。详细介绍请参考编译器使用指南

  • 基于clang/llvm的编译框架,实现面向PPU架构的、host/device混合编程风格的C/C扩展语言编译器,完整兼容cuda c/c的编程语言规范

  • 提供丰富的编译功能模块,方便开发者通过API的调用方式,灵活构建编译流程,方便集成JIT编译的能力

  • 实现inline ptx指令的支持,其中包括部分支持的dense mma指令、sparse mma指令、texture相关指令; 同时针对ppu tensor core扩展了ptx指令的定义

  • 实现CUDA device API的支持

  • 支持system level reserved shared memory特性

  • gcc host compiler的版本支持范围在[5.5 - 15.0],clang host compiler的支持范围在[clang 9 - clang 21]。

  • 支持triton 2.3.x、3.0.x、3.1.x、3.2.x、3.3.x、3.4.x、3.5.x。

  • 扩展triton.lang的语法,引入对AIU的支持;支持tilelang 0.1.6、0.1.7、0.1.7.post3。

  • Device ELF binary格式默认打开split section选项,减少运行时module load开销

  • 丰富的开发库和二进制工具:hgobjdump、memcheck、ppu-gdb、sanitizer library、hgprune、hgfatbin、hglink & hgJitlink,可以让用户更加方便地开发和调试

    • 运行时JIT编译库hgrtc,详细请参考HGRTC API(v2.1) 最新更新。

      • 支持对编译产物的caching能力,减少JIT编译过程时间

      • 支持Precompiled Header(PCH),减少JIT编译过程时间

    • 发布了PPU 调试工具PPU-GDB,允许在同一个应用程序中同时调试GPUCPU代码,详细介绍请参考GDB使用指南(v2.1)

    • 发布了PPU Memcheck,是一组用于功能性正确检查的工具套件。该套件中包含了一系列的检查工具,包括memcheck、initcheck、synccheck、racecheck,详细介绍请参考Memcheck使用指南 (v2.1)

    • 发布了PPU Binary 工具hgobjdump,用于提取binary中的device相关信息,详细介绍请参考hgobjdump使用指南(v2.1)

    • 发布了PPU Prune 工具hgprune,用于提取binary中的device相关信息,详细介绍请参考hgprune使用指南(v2.1)

    • 发布了PPU Fabin开发库hgfatbin,支持运行时对fabin文件的各类操作。最新更新:

      • 支持使用zstd作为fatbin默认压缩算法。

      • 支持通过--compress-mode调整压缩等级,默认启用快速压缩等级兼顾压缩效率和加载性能。

    • 发布了PPU Link 工具hglink,用于对PPU device codelink操作,详细介绍请参考hglink&hgJitLink使用指南(v2.1)

    • 发布了PPU JitLink开发库hgjitlink,支持运行时对PPU device codelink操作,详细介绍请参考hglink&hgJitLink使用指南(v2.1)

2.4.6 加速库

2.4.6.1 计算加速库
  • 闭源计算库支持:acdnn、acblas、acfft、acsolver、acrand、acsparse

    • acdnn支持算子:

      Conv

      BatchNorm

      Pooling

      Softmax

      Activation

      CTCLoss

      Dropout

      LRN

      LSTM

      GRU

      MultiHeadAttn

      Tensor Ops

      SpartialTransform

      Backend fusion

    • 新增发布 cudnn9 独立安装包。

    • acblas支持算子:

      Level1 系列 Op

      Gemv

      Gemm

      Matmul + epilogue

      MatrixTransform

      trsm

      getrfBatched

      getrsBatched

      geqrfBatched

      gelsBatched

      geam

    • 新增 SetVector/GetVector/SetMatrix/GetMatrix 系列 API 支持。

    • acfft支持:R2C/C2R/C2C/D2Z/Z2Z + FFT/iFFT 变换。

    • acsolver支持:矩阵LU分解/求解,cholesky分解/求解,QR分解,SVD分解,特征值分解。

    • acrand支持:

      • 伪随机生成器XORWOW、MRG32K3A、PHILOX4_32_10。

      • 数据分布:Default/Uniform/Normal/LogNormal。

    • acsparse支持:

      • 支持大部分generic API及部分legacy API。

      • cusparse 13.0中表明为[DEPRECATED] 的API未支持。

    • acext量化库支持

      • 支持 A16W8/A16W4以及PerChannel/GroupWise的各种Kernel变种。

      • 支持A8W8以及PerChannel/PerToken的各种Kernel变种。

      • 支持WeightonlyBatchedGemv对小batchsize的加速kernel。

      • 支持以下类型 MoE:FP16/BF16,a8w8 PerChannel/PerToken。

      • 支持A8W4-Int8 PerToken/PerChannel GroupGemm Kernel。

    • 开源项目支持:

      • PPU FlashMLA 1v5

      • PPU FlashAttention 2.4.2/2.5.6/2.5.7/2.7.2/2.7.4.post1/2.8.2

      • PPU cutlass2.5

      • PPU cutlass3.3/3.4/3.6

      • PPU Xformers 0.0.22/0.0.25/0.0.27/0.0.29.post1/0.0.30

      • PPU FasterTransformer

      • PPU 扩展库 acext

        • 目前主要支持高性能 a16w8、a16w4、a8w8/a8w4 量化、group-gemm、MoE、batch gemv 等实现。

        • 提供 PPU_SDK library 和开源仓库两种方式,供各框架和应用集成。

      • PPU DeepGemm:

        • 支持 dense gemm、group gemm (contiguous、masked、nopad接口)。

        • 支持 FP16、A8W8 format。

        • 支持 PerChanel/PerToken 量化。

2.4.6.2 互联加速库
重要

由于对外云产品使用的网卡为EIC,对客界面可使用的加速库为ACCL-P(替代PCCL)和ACCLEP-P(替代DeepEP)。

  • PCCL:

    • 兼容支持绝大多数 nccl api (ncclXXX) 和环境变量: PCCL: NCCL APIs及环境变量支持情况(v2.1)

    • 支持 AllReduce、AllGather、ReduceScatter、Broadcast、Reduce、Send、Recv 等典型互联算子。

    • 支持单机内多卡通过 ICN、PCIE、ShareMemory 通信。

    • 支持多机之间通过 RDMA (GDR & non GDR) 、Socket、ICN 通信。

    • 支持真武810E芯片构成的多卡服务器类型。

    • 优化了部分配置下 native 算子的性能。

    • 完善 symmetric memory 功能支持并优化性能。

    • 完善 'CE collective' 功能支持与优化性能。

    • 进一步增强 hang debug 相关功能,详细介绍请参考PCCL: 多卡异常问题分析指南(v2.1)

    • pccl tools:

  • sailSHMEM:

    • 兼容支持所有 nvshmem pt2pt 相关 api。

    • 支持单机内与多机节点间的 PPU 卡 pt2pt 操作像 thread/warp/block 级别的 put/get bw、 latency 与 signal 及 wait 等同步操作。

    • 支持 icn link p2p 传输与 rdma ibrc 与 ibgda 网络传输。

    • 支持真武810E芯片构成的多卡服务器类型。

    • 完成基于 nvshmem v3.4.5-0 的 rebase。

  • DeepEP:

    • 支持真武810E 芯片构成的多卡服务器类型。

    • 兼容支持 DeepEP 绝大多数 python API 用法,详情请参考DeepEP 通信库使用指南(v2.1)

    • 支持单机内与节点间的 dispatch 与 combine intranode、internode 与 internode low latency 三种类型 kernel 实现。

    • 支持通过 icn link 及节点间的 rdma ibgda 或 ibrc 等传输方式进行通信。

    • 支持 low latency dispatch kernel 传输的 int8 量化。

    • low latency dispatch 与 combine 算子性能在部分配置下做了优化。

    • 新功能:

      • 支持 internode dispatch cuda graph 功能。

      • 支持 combine single batch overlap 功能。

      • 支持更多 MoE 模型配置。

2.4.7 Video/Image 硬件加速

兼容 Nvidia Video Codec SDK,包括cuvid decode, nvenc(up to v13.0)和nvjpeg(up to v13.0);兼容NPP(up to v13.0)接口。基于此,可以直接支持ffmpeg(up to v8.0), OpenCV4(up to 4.13),DALI(up to v1.53),PyNvVideoCodec, PyAV,TorchCodec(up to v0.9.1),HgDeepStream 等上层框架的Codec硬件加速能力。

Video Decode

  • Support Nvidia cuvid decoder。

  • Codecs:

    • HEVC (H.265) - ITU-T Rec. H.265 (04/2013), ISO/IEC 23008-2

      • Main Profile, Level 5.1, High Tier

      • Main10 Profile, Level 5.1, High Tier

      • Main Still Profile

    • VP9 - vp9-bitstream-specification-v0.6-20160331-draft

      • Profile 0, 8-bit

      • Profile 2, 10-bit

    • AVC (H.264) - ITU-T Rec. H.264 (03/2010) / ISO/IEC 14496-10

      • Main Profile, levels 1 - 5.2

      • High Profile, levels 1 - 5.2

      • High 10 Profile, levels 1 - 5.2

      • Baseline Profile, levels 1 - 5.2

    • AV1 Bitstream & Decoding Process Specification Version 1.0.0 with Errata 1

      • Main Profile, Level 5.1

    • AVS2

  • Resolution Up to 8192x8192

Video Encode

  • Support nvenc

  • Codecs:

    • AVC(H.264):Spec Version 12:ISO/IEC 14496-10 / ITU-T Rec. H.264 (03/2010)

      • Baseline Profile, levels 1 – 5.2

      • Main Profile, levels 1 - 5.2

      • High Profile, levels 1 – 5.2

      • High 10 Profile, levels 1 - 5.2

    • HEVC(H265):ITU-T Rec. H.265 (04/2013), ISO/IEC 23008-2

      • Main Profile, Level 5.1, High Tier

      • Main10 profile, Level 5.1, High Tier

      • Main Still Profile

    • AV1 Bitstream Specification Version 1.0.0 with Errata 1

      • Main Profile, Level 5.1

  • Resolution up to 4K support

  • Support input RGB format (converted to YUV420 via inlinePP)

  • Support crop, scale, rotate with inlinePP

Jpeg

  • Support nvjpeg decoder & encoder

  • Resolution up to 32Kx32K

  • Support RGB format input and output with inlinePP

  • Support crop, scale, rotate with inlinePP

Image Process

  • Support Nvidia 2D Image NPP

Performance

真武810E

FHD 160 streams

FHD 32 streams

UHD 960FPS

2.4.8 软件工具

2.4.8.1 PPU-SMI

为了满足云计算大规模集群监控需求,我们发布了如下 PPU 管理和监控工具和库文件,以便集成到客户集群运维监控系统中。

PPU-SMI(PPU System Management Interface)是一个基于HGML(HanGuang Management Library)的命令行工具,用于辅助用户管理和查看PPU设备。

通过PPU-SMI命令行工具,用户可以:

  • 修改设备配置 / 特性开关

  • 查询指定设备运行参数和特性使能状态

  • 收集运行数据 / 特定事件,导出至表格供后续分析

  • 分析各个应用程序的设备资源使用情况

  • 查询多个PPU设备的拓扑信息

详细介绍请参考 设备管理工具PPU-SMI (v2.1)

  • ppu-smi v2.1 新功能主要包含:

    • 增加支持指定设备组件复位的描述

    • 增加查询PCI class code、addressing mode、Fabric信息、Compute Capability、EGM能力的描述

    • 增加drain子命令支持discoverremove选项的描述

2.4.8.2 PPU DCGM

PPU DCGM是一套用于在集群和数据中心环境中管理和监控PPU的工具,基于开源的DCGM。包括如下组件:

  • dcgmi命令行工具

  • nv-hostengine命令行工具

  • dcgm共享库

  • dcgm-exporter工具

详细介绍请参考管理监控工具DCGM(v2.1)

  • DCGM v2.1新功能主要包括:

    • 更新diag usage information

    • 更新Field Id支持状态

    • 更新hgdcgm-exporter执行结果示例

2.4.8.3 Asight Systems

发布了 PPU 性能分析工具 Asight Systems 和 Asight Compue(类似 Nvidia Nsight Systems/ Compute),可以支持用户进行单机、多机训练、推理等场景的性能分析;

Asight Systems 是一款低开销的系统级的性能分析工具,用来采集系统各种事件,CPU 和 PPU 的活动,API 执行时间以及相关调用栈,NVTX,CPU/PPU activity关联关系等,在 Timeline View 上统一的可视化呈现出来。 通过 imeline View,开发人员可以方便分析 CPU/PPU 的负载和关联关系,找到性能瓶颈,确保 CPU 和 PPU 能够协调的工作,确保最大的并行度。Asight支持统计系统方便对报告进行后处理。详细介绍请参考程序性能分析套件Asight Systems (v2.1)

Asight Systems v2.1新功能主要包含:

  • asys stats子命令支持PPU operator type summary统计规则,支持分析算子类型的占比,支持列出热点算子。

  • asys stats子命令支持PPU operator unit test规则,支持批量生成pytorch算子测试用例代码(实验特性)。

  • asys stats子命令支持PPU Operator BreakdownPPU Operator Kernel Breakdown统计规则,支持对于各层级的算子输出性能分解结果。

  • asys stats子命令支持PPU Operator Breakdown ComparePPU Operator Kernel Breakdown Compare比较规则以比较性能分解结果。

  • asys stats子命令支持PPU operator trace跟踪导出功能,支持导出每算子关联的PPU活动跟踪。

  • asys stats子命令支持PPU kernel summary compare对比规则,支持比较两个报告的PPU kernel汇总结果。

  • asys stats子命令支持HGTX PPU projection summary compare对比规则,支持比较两个报告的HGTXPPU投影汇总结果。

  • asys stats子命令支持PPU operator summary compare对比规则,支持比较两个报告的算子汇总结果。

  • asys stats子命令支持PPU operator type summary compare对比规则,支持比较两个报告的算子类型汇总结果。

  • asys stats子命令支持PCCL desynchronization summary compare对比规则,支持比较两个报告的PCCL通信不同步率。

  • asys stats子命令支持PPU time utilization compare对比规则,支持比较两个报告的PPU时间利用率差异。

  • asys stats子命令PCCL Desynchronization Summary支持P90统计,支持显示大部分PCCL通信不同步率优于此门限调整。

  • asys stats子命令HGTX关联PPU活动相关统计规则支持range-includerange-exclude选项,支持指定HGTX过滤黑名单和白名单。

  • asys stats子命令PPU算子相关统计规则支持order-by选项,支持指定PPU算子统计结果排序方式。

  • asys stats子命令指定统计报告类型时,支持对逗号,和冒号:使用\逃脱符。

  • asys stats子命令在没有指定统计报告类型和输出类型时,默认打印简要的统计结果到终端,并输出详细的统计结果到文件。

  • asys profile新增支持--stats选项,支持在生成报告后打印简要的统计结果到终端,并输出详细的统计结果到文件。

  • asys支持采集pytorch算子跟踪,支持通过HGTX标记pytorch算子的执行时间,支持采集pytorch算子参数信息。

  • asys profile支持通过--format选项指定xlsx输出格式。

  • asys statsanalysis子命令column格式输出样式,使输出结果更加易读,调整asys stats默认生效的统计规则。

  • PPU operator summary统计功能支持识别更多算子,支持自定义算子识别规则。

  • 新增asys compare子命令比较两个报告文件,支持显示精简的比较结果,并输出详细比较数据到csv文件。

  • Timeline View: 时间线tooltip支持切换显示完整时间线名字。

  • Timeline View :为选中的时间线添加更显著的高亮效果和边框颜色,增强选中状态的可视性。

  • Timeline View: 支持按统一高度比例尺显示PPU metricsCPU metrics时间线。

  • Timeline View: 当HGTX行的层数较多时,确保标题名字居中显示。

  • Timeline View中更新range filter时,自动更新Heap Memory View的内容。

  • Events View默认以list模式显示,并提升了list模式下各项操作的性能

  • Funcition View: 支持选择目标线程,选择后仅对该线程进行分析。

  • Asight Systems新增比较视图(Comparision View),支持选择任意两个已打开的.asysrep报告,对其算子、kernel等性能进行对比。

2.4.8.4 Asight Compute

Asight Compute是一款kernel性能分析工具,通过采集PPU硬件 perf counter,组合成为一系列性能指标,我们称为metrics。GUI通过各种维度,把这些metrics呈现出来, 帮助用户深入分析和优化kernel。详细介绍请参考Kernel分析器Asight Compute (v2.1)

Asight Compute v2.1新功能主要包含:

  • acu命令行参数,--csv参数支持与--page联动,输出格式受--page参数控制。

  • acu命令行新增--csv-file <文件> 参数,支持将csv数据直接输出到文件中。

  • acu新增统计系统功能,系统通过基于Python脚本的内置规则或者自定义规则,汇总、计算性能数据,并将结果导出到文件或直接显示在终端。

  • 在使用--kernel-name--kernel-id 选项时如果遇到 "No Kernels" 错误,acu工具会自动打印出当前程序中所有可用的 kernel 名称

  • 增强Details Pagetooltip显示,集成tab与表格等内容,丰富信息展示样式。

  • 提升baseline对比的用户体验,current kernelbaseline时不再参与对比。

2.5 支持的操作系统

类别

操作系统

架构

GCC

Ubuntu

Ubuntu 24.04 LTS

x86_64

13.3.0

Ubuntu 22.04 LTS

11.4.0

Ubuntu 20.04 LTS

9.5.0

Ubuntu 18.04 LTS

7.5.0

2.6 版本兼容性说明

2.6.1 KMD兼容性

  • V2.1版本SDK向前兼容V1.6.xV2.0.x版本的KMD。

  • V2.1版本KMD向前兼容V1.7.xV2.0.x版本的SDK。

  • V2.1版本SDK推荐搭配V2.1版本KMD使用,以获得最全的功能和最佳性能。

2.6.2 Firmware兼容性

  • 真武810E 需要使用 V1.2.1以上的 Firmware 版本。

2.6.3 SDK兼容性

编译器具备在编译过程中产生多代Device Binary的能力,device Fatbin的文件格式版本随之迭代,在兼容性方面存在一些变化。

SDK V2.1 同 SDK V2.0、V1.7、V1.6、V1.5之间,保持在编程API接口的兼容性。但在Device二进制格式和库文件的兼容性方面:

  • 后向兼容

    在旧版本SDK(V2.0、V1.7、V1.6、V1.5)上编译的产物能够在SDK V2.1的环境中正常执行;

  • 前向不兼容

    SDK V2.1上编译的产物不能确保在旧版本SDK(V2.0、V1.7、V1.6、V1.5)的环境中正常执行;

SDK V2.1 同 SDK V1.4及之前的版本不兼容。

3. CUDA 生态兼容

3.1 简介

在 PPU 平台上开发应用程序,用户既可以基于 PPU SDK API 开发应用程序,也可以使用 CUDA 语言编写应用程序,经过 PPU 编译器重新编译后在 PPU 上运行,下图展示了同样的 CUDA 应用程序分别在 PPU 和 GPU 上编译、运行的差异。

**CUDA 生态兼容方法:**通过代码自动生成技术生成 CUDA SDK Wrapper 来自动兼容 CUDA 不同版本的 APIs,从而使得用户的 CUDA 程序经过 PPU 编译器重新编译后即可在 PPU 上运行。

通过 CUDA 生态系统的兼容方案,PPU 与 GPU 在源代码级别上是编译兼容的,但 PPU 二进制与 GPU 二进制不兼容。

image.png

3.2 CUDA APIs 兼容版本

截至 PPU SDK V2.1 版本,我们对 CUDA APIs 正式版本支持到 13.0, 但是限于硬件架构的区别,我们对 CUDA APIs 兼容主要在 DeepLearning 范畴,下表是目前我们的支持版本列表。

CUDA Version

11.1

11.2

11.3

11.4

11.5

11.6

11.7

11.8

12.1

12.2

12.3

12.4

12.5

12.6

12.8

12.9

13.0

具体API支持状态参考文档目录 “CUDA软件生态兼容状态”。

3.3 支持的开源框架/库

4. 已知问题

4.1 驱动

4.2 编译器

  • 不支持大部分的Texture 和 Surface 相关的 Cuda C++ 扩展 API和相关inline PTX,编译会报错;

  • 不支持 Dynamic Parallelism 相关的 Cuda C++ 扩展 API、Inline PTX 指令的功能,编译会报错;

  • 针对ptx 8.7之上的新增指令,会存在因硬件架构不支持的部分功能,不影响编译,但运行时会报错;

  • 不支持 Inline PTX 中 128bit atomic 指令,不影响编译,但运行时会报错;

  • 不支持 Inline PTX 中 multimem 指令,不影响编译,但运行时会报错;

  • 不支持 Inline PTX 中 griddepcontrol 指令,忽略其定义但不影响编译和运行的过程;

  • 不支持 Inline PTX 中 ld/st 相关指令带有 {.level::eviction_priority} 和 {.level::prefetch_size} 的特性,忽略其定义但不影响编译和运行的过程;

  • 不支持 Inline PTX 中 cache eviction policy 相关的指令和操作数,忽略其定义但不影响编译和运行的过程;

  • Device 文件编译流程包括 Cuda Device C++ 代码 -> llvm(hgvm) IR -> Device Binary的过程, 但不包含输出 ptx 格式的文件过程; 针对其他平台的代码编译(或Codegen)环节,如带有 ptx 格式的编译环节,需要进行代码适配;

  • 兼容CUDA mma及相关数据搬运的ptx指令中大部分,范围包括特定数据类型(.u8/.s8/.tf32/.bf16/.f16)下的dense/sparse mma指令,相比于使用ppu specific tensor core ptx指令实现,性能会存在损失。如果此类kernel的性能在整个端到端中占比重要,则建议对该kernel代码实现进行算法的重构(参考ppu tensor core ptx用户编程手册及算法重构指南);对不支持的数据类型和功能(如u4/f64/sp::ordered_metadata),不影响编译,但运行时会报错;

4.3 加速库

  • 性能:性能泛化能力加强中。

  • acblas:

    • 仅支持列表API,详见CUBLAS APIs支持状态(v2.1)更多API持续按需支持中。

    • 不支持复数数据类型。

    • Gemm: 默认打开 FP32 Tensor Core,由于计算顺序等原因导致精度不能和 FP32 FMA 完全配置,matrixMulCUBLAS/simpleCUBLAS 示例会因此失败;

      • 可以通过 export PPU_FP32_TENSOR_OVERRIDE=0 关闭 FP32 Tensor Core 解决;

    • Gemv:仅支持 host 指针模式;

    • BlasLt:不支持 algo/perf 等指定属性;

  • acdnn:

    • 仅支持列表API,详见 CUDNN APIs支持状态(v2.1),更多API持续按需支持中。

    • Conv:不支持INT64/BOOLEAN数据类型,不支持输入FP16 + 输出FP32。

    • 3DConv:有限调优,性能待加强;

    • depthwise:某些dgrad用例性能待加强;

    • BN:1)仅支持alpha1beta0参数;2)不支持ACDNN_BATCHNORM_PER_ACTIVATION模式。

    • Pooling:不支持ACDNN_PROPAGATE_NAN。

    • RNN:仅支持acdnnRNNBiasMode_t DOUBLE;仅支持FP16/F32数据类型;仅支持ACDNN_RNN_ALGO_STANDARD;

    • Activation:不支持ACDNN_PROPAGATE_NAN;不支持SWISH Op。

    • Softmax:不支持SoftmaxAlgorithm_t FAST。

    • TensorOp:acdnnReduceTensor不支持MUL_NO_ZEROS。

    • MultiHeadAttn:仅支持前向op。

    • Backend:1)不支持前处理融合;2)仅支持最多4pointwise后处理融合;3)仅支持fp16/fp32/bf16数据类型;4)融合的pointwise操作仅支持alpha1 = 1alpha2 = 1。

  • acsolver:

  • acfft:

  • acrand:

    • 仅支持列表API,详见CURAND APIs支持状态(v2.1),更多API持续按需支持中。

    • 仅支持类型:XORWOW/MRG32K3A/PHILOX4_32_10;

    • 仅支持Legacy order;

    • 仅支持分布类型:default、uniform、uniform double、normal、normal double、lognormal、lognormal double;

  • acsparse:

  • acext:

    • 只支持TPNative EP运行模式,不支持DeepEP;

    • A8W4-Int8算子量化仅支持Channelwise,暂不支持Groupwise/Blockwise,部分模型可能存在精度损失;且目前只接入MoE接口,暂未支持DenseGemm接口;ACEXT算子性能泛化需配合AutoTune LUT;

  • cutlass:

    • 目前只能使用 cutlass/examples/CMakeLists.txt 上的示例。

  • cutlass3:

    • 目前只能使用 cutlass/examples/CMakeLists.txt 上的示例。

4.4 互联库

重要

由于对外云产品使用的网卡为EIC,对客界面可使用的加速库为ACCL-P(替代PCCL)和ACCLEP-P(替代DeepEP)。

  • 尚不支持的 nccl APIs 与环境变量列表可见:PCCL: NCCL APIs及环境变量支持情况(v2.1)

  • Collective操作性能:

    • 16 x 真武 810E 服务器

      • 采用 TP 2 / 4 / 8 mode 时建议 aware 自己所用机器的 icn topo 配置情况以来选择最佳的 devices placement。

  • 暂不支持 nccl netplugin v7 及以上的版本;

  • 执行收发 sizes 不均衡的 alltoallV 类型操作会有概率性死锁问题,dmesg 中能看到 "ERR_FAB_REQ_TO" 类型的 exceptions,此时需做 device reset 以恢复为正常状态;

  • 多个 pccl communicators 同时工作在相同 ppu 下面时,会有概率性死锁的问题, 需尽量规避多个 pccl kernels 同时 launch 在同一 ppu device 下的行为; 避免设置 NCCL_MIN_NCHANNELS 为大于 16 的值可帮助规避此问题 ;

4.5 Video Codec/Image 硬件加速

  • Video decode 不支持 MPEG1,MPEG2,MPEG4,VC1,VP8 等 legacy 格式;

  • JPEG 不支持 lossless,不支持 JPEG2000;

  • NPP 目前只支持 Image Process 接口,不支持 Signal Process 接口;

4.6 工具

4.7 框架与模型

开源框架已知问题说明

框架

版本

已知问题

PyTorch

2.9

UT FAIL的分类说明:

  1. CUDA API 暂不支持

  2. 预期结果:

    1. A100 FAIL

    2. 测试中Hardcode NV binary的名字

    3. 硬件SFU导致的差异 等

  3. 个别tensor元素误差略大于测试用例阈值

  4. 运行超时time out

SGlang

0.5.9

  • 社区的临时解决方案是改写 tilelang act_quant 成 triton kernel,从而在运行时不会 import tilelang,但因为我们复用了 tilelang mqa_logits 相关的 kernel,所以无法移除对 tilelang 的依赖,因此将 flashinfer 暂时回退到适配 tvm-ffi 前的 v0.4.0rc3 版本。

  • 待不同社区 flashinfer/tilelang/sglang 使用统一的 tvm-ffi 版本后,我们将会在后续版本回退这个改动。

  • AWQ 和 GPTQ 未深度优化存在性能问题,后续视业务需要进行针对性优化。建议优先使用PTG提供的int8量化方案。

  • 运行 a8w8(int8)量化模型需要加 --quantization w8a8_int8 选项。

VLLM

v0.16.0

  • 社区vLLM0.16.0上并未支持Qwen3.5系列模型,PPU进行了部分支持(TP),无法支持Qwen3.5系列模型的所有功能、所有case。如果需要使用Qwen3.5系列模型,请使用v0.17.1以上的版本。目前的已知问题:

    • 开启MTP功能后概率性fail

    • bfcl_v3 精度测试(enable tool call功能)fail

    • 不支持EP

  • GLM-5Qwen3.5系列模型运行需要升级 transformers ==5.2.0

  • DP+EP+DeepEP low latency问题:

    • 第一次启动server需要设置export VLLM_ENGINE_READY_TIMEOUT_S=6000,否则可能会因DeepGemm warmup编译超时导致server启动失败。社区在H20上存在相同问题。

    • 对于 Qwen3 MoE BF16 的case,存在精度问题。社区同样存在该精度问题。请不要使用DP+EP+DeepEP low latency这种组合启动Qwen3 MoE BF16 模型权重

  • MiniMax-M2.5模型 TP=16 server fail,请使用TP=8 运行int8量化权重。

  • --gpu_memory_utilization的值高于0.96时,会概率性遇到HGGC OOM问题,调低gpu_memory_utilization可避免该问题。

  • 开启Flashinfer Sampler可以获得性能提升,但可能会导致精度掉点(默认关闭该功能)。社区同样存在该问题。可通过 VLLM_USE_FLASHINFER_SAMPLER 环境变量控制该功能的开启或关闭。

  • 精度问题:

    • MiniMax-M2模型 ifeval精度测试得分较低,社区在H20上同样存在该问题

    • Kimi-linear模型存在精度问题,社区同样存在精度问题

  • Flashinfer Sampler可能会导致精度掉点(该功能默认开启)。如遇到模型精度掉点问题,可通过 export VLLM_USE_FLASHINFER_SAMPLER=0 关闭该功能

  • 如果使用 DeepGemm backend,server 启动阶段则会自动进行 warmup,编译出需要用到的 DeepGemm kernel,可能会带来较长的 warmup 时间:

    • warmup时间较长。推荐做法是通过环境变量指定deepgemm cache 路径:export DG_CACHE_DIR=<your path>,从而避免重复编译。

    • 可以通过export VLLM_DEEP_GEMM_WARMUP="skip"跳过 DeepGemm warmup,测试性能过程中请确保未出现 DeepGemm JiT 编译,否则会导致性能下降。

  • 量化方面,目前 SAIL vLLM 0.16.0 暂未对 Marlin kernel 进行适配和优化,AWQ(w4a16)、GPTQ (w4a16、w8a16)、mxfp4 性能较差,请使用 int8 w8a8 量化方案。

  • Kimi-linear | DeepSeek-OCR 系列 模型使用 DeepGemm 将会导致 server 启动失败。请使用 Acext MoE backend。

    # 设置环境变量,使用Acext MoE backend
    export VLLM_USE_DEEP_GEMM=0
    export VLLM_MOE_USE_ACEXT=1
    export VLLM_DENSE_USE_ACEXT=1
    export ACEXT_NUM_TOKENS_LIMIT=16385

v0.17.1

  • GLM-5模型运行需要升级 transformers ==5.2.0

  • DP+EP+DeepEP low latency问题:

    • 第一次启动server需要设置export VLLM_ENGINE_READY_TIMEOUT_S=6000,否则可能会因DeepGemm warmup编译超时导致server启动失败。社区在H20上存在相同问题。

    • 对于 Qwen3 MoE BF16 的case,存在精度问题。社区同样存在该精度问题。请不要使用DP+EP+DeepEP low latency这种组合启动Qwen3 MoE BF16 模型权重

  • MiniMax-M2.5模型 TP=16 server fail,请使用TP=8 运行int8量化权重。

  • --gpu_memory_utilization的值高于0.96时,会概率性遇到HGGC OOM问题,调低gpu_memory_utilization可避免该问题。

  • 开启Flashinfer Sampler可以获得性能提升,但可能会导致精度掉点(默认关闭该功能)。社区同样存在该问题。可通过 VLLM_USE_FLASHINFER_SAMPLER 环境变量控制该功能的开启或关闭。

  • 精度问题:

    • MiniMax-M2模型 ifeval精度测试得分较低,社区在H20上同样存在该问题

    • Kimi-linear模型存在精度问题,社区在H20上同样存在该问题

    • Qwen3-Next系列模型相比vLLM 0.16.0,存在精度掉点,社区在H20上同样存在该问题

    • Qwen2.5-7B模型存在精度问题,社区在H20上同样存在该问题

  • 如果使用 DeepGemm backend,server 启动阶段则会自动进行 warmup,编译出需要用到的 DeepGemm kernel,可能会带来较长的 warmup 时间:

    • warmup时间较长。推荐做法是通过环境变量指定deepgemm cache 路径:export DG_CACHE_DIR=<your path>,从而避免重复编译。

    • 可以通过export VLLM_DEEP_GEMM_WARMUP="skip"跳过 DeepGemm warmup,测试性能过程中请确保未出现 DeepGemm JiT 编译,否则会导致性能下降。

  • 量化方面,目前 SAIL vLLM 0.17.1 暂未对 Marlin kernel 进行适配和优化,AWQ(w4a16)、GPTQ (w4a16、w8a16)、mxfp4 性能较差,请使用 int8 w8a8 量化方案。

  • Kimi-linear | DeepSeek-OCR 系列 模型使用 DeepGemm 将会导致 server 启动失败。请使用 Acext MoE backend

    # 设置环境变量,使用Acext MoE backend
    export VLLM_USE_DEEP_GEMM=0
    export VLLM_MOE_USE_ACEXT=1
    export VLLM_DENSE_USE_ACEXT=1
    export ACEXT_NUM_TOKENS_LIMIT=16385

v0.18.0

  • GLM-5模型运行需要升级 transformers==5.2.0

  • DP+EP+DeepEP low latency问题:

    • 第一次启动server需要设置export VLLM_ENGINE_READY_TIMEOUT_S=6000,否则可能会因DeepGemm warmup编译超时导致server启动失败。社区在H20上存在相同问题。

    • 对于 Qwen3 MoE BF16 的case,存在精度问题。社区同样存在该精度问题。请不要使用DP+EP+DeepEP low latency这种组合启动Qwen3 MoE BF16 模型权重

  • MiniMax-M2.5模型 TP=16 server fail,请使用TP=8 运行int8量化权重。

  • --gpu_memory_utilization的值高于0.96时,会概率性遇到HGGC OOM问题,调低gpu_memory_utilization可避免该问题。

  • 开启Flashinfer Sampler可以获得性能提升,但可能会导致精度掉点(默认关闭该功能)。社区同样存在该问题。可通过 VLLM_USE_FLASHINFER_SAMPLER 环境变量控制该功能的开启或关闭。

  • 如果使用 DeepGemm backend,server 启动阶段则会自动进行 warmup,编译出需要用到的 DeepGemm kernel,可能会带来较长的 warmup 时间:

    • warmup时间较长。推荐做法是通过环境变量指定deepgemm cache 路径:export DG_CACHE_DIR=<your path>,从而避免重复编译。

    • 可以通过export VLLM_DEEP_GEMM_WARMUP="skip"跳过 DeepGemm warmup,测试性能过程中请确保未出现 DeepGemm JiT 编译,否则会导致性能下降。

  • 量化方面,目前 SAIL vLLM 0.18.0 暂未对 Marlin kernel 进行适配和优化,AWQ(w4a16)、GPTQ (w4a16、w8a16)、mxfp4 性能较差,请使用 int8 w8a8 量化方案。

TransformerEngine

v2.10

  • 为了更好的性能表现,我们推荐用户在TE里面使用flashattention的后端。因为暂时不支持cudnn中的fusedattention backend,在PPU版本的TE中做了保护性重定向,所有指定fusedattention backend的操作都会导向flashattention的后端。

  • 已知问题列表:

    • PPU1.0/PPU1.1上不支持fp8/fp4相关特性的计算与围绕着其构建的功能,包括FP8 current/delay scaling, MXFP8 and block scaling以及NVFP4的训练特性等。

    • 目前PPU Transformer Engine只支持torchbackend,暂时不支持使用jax

Flashinfer

v0.6.3

目前ppu版本flashinfer存在性能问题,在端到端性能中,TPOT指标与A100存在性能差距,需要继续优化:

  • test_batch_attention.py

    • Fail Instance:23039

    • Fail Reason:没有对mma适配,sglang暂时没有支持需求

  • test_attention_sink.py

    • Fail Instance:1336

    • Fail Reason:用了hopper的特性

  • test_cudnn_decode.py

    • Fail Instance:12

    • Fail Reason:cudnnFinalize failedptrDesc->finalize() cudnn_status: CUDNN_STATUS_BAD_PARAM,cudnn版本兼容问题

  • test_group_gemm.py

    • Fail Instance:26

    • Fail Reason:sglang暂时没有适配需求 AssertionError(AssertionError: Unknown cases: sf.dtype=torch.float32)

  • test_trtllm_alltoall.py

    • Fail Instance:6

    • Fail Reason:trtllm底层库不支持AssertionError: Process 0 failed with exit code 1test_trtllm_moe_allreduce_fusion.py:443: AssertionError

  • test_trtllm_gen_mla.py

    • Fail Instance:1232

    • Fail Reason:trtllm底层库不支持

Jax

0.8.0

  1. JAX 依赖 numpy >= 2.0.0, 如果需要配合 torch/tf 等框架,需要单独编译。

  2. 只支持在 ubuntu 上编译运行,不支持 alios。

  3. examples/advi.py 运行出现 nan。

  4. 暂不支持 xla_gpu_enable_llvm_module_compilation_parallelism = True。

  5. UT FAIL 的分类说明:

    1. cusolver/cusparse/cublas api 不支持导致的 fail。

    2. acdnn 内部错误导致 jaxlib._jax.XlaRuntimeError: UNKNOWN: Failed to determine best cudnn convolution algorithm。

Megatron

v0.15.0

  • UT 测试 pass-rate: 98.24% (810E), 98.51%(h20),部分 fail case 由于未安装必要依赖库导致;有部分case 要求 SM >= 90

v0.16.0

DeepSpeed

v0.18.7

  • UT 测试: 当前 UT 测试通过率为 95.74%

  • PPU DeepSpeedv0.18.7版本开始会支持每个小版本的whl包发布

ROLL

v0.1.3

部分UT存在Failed的情况:

  1. 一部份测试脚本需要多节点而导致Failed.

  2. 部分test case存在hang的问题,该情况与NVIDIA H20上的行为保持一致.

veRL

v0.7.0

  1. veRLFusedKernelsPPU上性能未达预期,请在训练脚本中设置:actor_rollout_ref.model.use_fused_kernels=False

  2. veRL0.7.0的全量 UT 测试:

    • 使用SGLang后端的镜像:PPU pass rate84.79%,GPU84.48%。未通过的case大部分与PPU无关。

    • 使用vLLM后端的镜像:PPU pass rate83.21%,GPU82.55%。未通过的case大部分与PPU无关。

    • 测试models/test_transformers_ulysses.py出现hang,该情况与GPU上的行为一致。

Slime

v0.2.1

  1. colocated模式下,即训练和推理在同一张卡上,在极少数复杂情况会出现Segfault 或 hang 的情况。如出现请切换到非colocated模式使用,取消启动脚本的--colocate ,并且合理分配训练和推理的卡数。

Warp

v1.10.0

  1. 目前 NVIDIA Warp 属于第一版支持,专注于客户场景如 Mujoco 和 Newton 的使能,目前 Warp 单软件的 UT 通过率是41.2% ,还有待提高。

  2. 某些硬件因为不支持conditional cuda graph,执行mujoco solver时会报错,改动mujoco-warp的一行代码即可解决。比如 /usr/local/lib/python3.12/site-packages/mujoco_warp/_src/solver.py,在文件的对应行加入and False关闭此功能即可。image.png

  3. 在运行Mujoco-Warpmake_constraint后,efc数据会在condim的维度反序,但是不影响每个contact的结果。