SDK Release Note

更新时间:
复制为 MD 格式

PPU SDK v1.7 Release Note

1. 版本概述

1.1 软件栈介绍

image

  • 为 CUDA 应用开发者使用PPU软硬件降本增效。

  • 三个层次交付物:兼容CUDA 生态的框架软件包与模型、CUDA SDK兼容软件包、T-Head SAIL SDK软件包。

1.2 核心组件

组件名称

概要说明

Firmware

PPU 固件

KMD

PPU 内核驱动

UMD / HGGC

PPU 用户态驱动和运行时

Compiler

PPU 编译器工具链

Acompute

PPU 计算加速库

Acext

PPU 量化加速库

PCCL

PPU 通信加速库

PPU SMI

PPU 设备管理工具

PPU DCGM

PPU 在线监控工具

Asight System

PPU 性能分析工具

Asight Compute

PPU 性能分析工具

PPU GDB

PPU 调试工具

PPU MemCheck

PPU Sanitizer工具

PPU hgobjdump

PPU Device Binary工具

CUDA SDK Wrapper

CUDA API 兼容库

1.3 主要功能

1.3.1 Firmware
  • 支持二进制包安装(rpmdeb);

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

  • 支持PPU锁频功能,详情请参见设备管理工具PPU-SMI

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

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

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

1.3.2 内核驱动
  • 支持二进制包安装(rpmdeb)和runfile包安装两种方式,用户可根据需要自由选择;

  • 支持内核驱动和PPU SDK解耦,v1.0版本之后的内核驱动和PPU SDK可任意组合使用;

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

  • 支持PPU故障上报和故障处理;

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

  • 支持MPS(Multi Pipe Service)功能,用户可通过ppudbg --config_submit_mode 1/0打开或关闭该功能;

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

  • 支持nvml GPM指标采集功能,包括:sm利用率、sm occupancy、tensor core利用率、显存带宽利用率、pcie读写速率、icn link读写速率;

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

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

  • 新增单机16卡真武810E的正式支持并优化默认的PPU设备顺序(按照机尾排序);

  • 新增L1/L2 cache命中率GPM指标采集,用户可通过DCGM工具查询获取;

  • 调整16卡真武810E设备的逻辑编号规则以获取更好的互联性能;

  • 新增MPS潮汐模式状态查询功能;

  • 新增Xid错误信息查询功能;

  • 新增GPM状态查询和GPM状态启停功能;

  • 取消HBM parity Xid错误上报;

1.3.3 用户态驱动和运行时
  • 兼容绝大多数cuda runtime api (cudaXXX) 和cuda driver api (cuXXX);

  • 新增发生xid 896错误时输出更多错误日志方便定位问题;

  • 优化2D/3D类型的cudaMemcpy/cudaMemset性能;

  • 支持stream memory operation v2版本API;

  • 支持graph managementedge data相关API;

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

  • 支持texture基本功能;

  • 优化CUDA_DEVICE_MAX_CONNECTIONS=1PPU性能;

  • 支持cuCtxCreate_v3 api;

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

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

  • 修复cuda graph不支持cooperative kernel node的问题;

  • 支持cudaLaunchKernelExC接口;

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

  • 支持AD扩展API接口;

  • 修复运行colmapcudaCreateTextureObject报错cudaErrorIllegalAddress的问题;

  • 修复cudaMallocAsync接口分配size0时报错的问题;

  • 支持cudaGetDriverEntryPointByVersion接口;

  • 支持cudaLaunchAttributeLaunchCompletionEvent特性;

  • 优化cudaCreateTextureObject接口性能;

  • 优化split section打开之后ppu-gdb工具性能问题;

  • 修复使用cudaLaunchHostFunc概率性挂死的问题;

  • 修复使用cuMemImportFromShareableHandle接口导入进程不可见设备显存报错的问题;

1.3.4 编译器
  • 基于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的支持,本次扩展Compiler Optimization Hint Function : __builtin_assume_uniform()

  • 支持system level reserved shared memory特性;

  • gcc host compiler的版本支持范围在[5.5 - 14.2],clang host compiler的支持范围在[clang 9 - clang 18];

  • 支持triton 2.3.x、3.0.x、3.1.x、3.2.x、3.3.x、3.4x;扩展triton.lang的语法,引入对AIU的支持;

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

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

    • 运行时JIT编译库hgrtc, 详细介绍请参考。最新更新:

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

    • 发布了PPU 调试工具PPU-GDB,允许在同一个应用程序中同时调试GPUCPU代码。

    • 发布了PPU Memcheck,是一组用于功能性正确检查的工具套件。该套件中包含了一系列的检查工具,包括memcheck、initcheck、synccheck、racecheck。

    • 发布了PPU Binary 工具hgobjdump,用于提取binary中的device相关信息。

    • 发布了PPU Prune 工具hgprune,用于提取binary中的device相关信息。

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

    • 发布了PPU Link 工具hglink,用于对PPU device codelink操作。

    • 发布了PPU JitLink开发库hgjitlink,支持运行时对PPU device codelink操作。

1.3.5 加速库

计算加速库

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

    • acdnn支持算子

      Conv

      BatchNorm

      Pooling

      Softmax

      Activation

      CTCLoss

      Dropout

      LRN

      LSTM

      GRU

      MultiHeadAttn

      Tensor Ops

      SpartialTransform

      Backend fusion

    • acblas支持算子

      Level1 系列 Op

      Gemv

      Gemm

      Matmul + epilogue

      MatrixTransform

      trsm

      getrfBatched

      getrsBatched

      geqrfBatched

      gelsBatched

    • 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以及PerChanel/PerToken的各种Kernel变种;

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

    • 支持以下类型 MoE:FP16/BF16,a8w8 PerChanel/PerToken;

  • 开源项目支持:

    • PPU FlashMLA 1v5;

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

    • 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 量化、group-gemm、MoE、batch gemv 等实现;

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

    • PPU DeepGemm:

      • 支持 dense gemm、group gemm;

      • 支持 FP16、A8W8 format,不支持 FP8;

      • 支持 PerChanel/PerToken 量化,暂不支持 BlockWise 量化;

  • 关键更新:

    • 完成 DeepGemm 初步支持:EP 典型场景性能达竞对水平;

    • 优化 DPSK/Qwen3 等场景 MoE 性能,相对trition版本平均提升50%以上;

    • blaslt:增强 desc 跨版本支持;完善通信计算并行方案支持;

    • conv:完善不同format组合的支持;增加对 strided channel 的支持;

    • solver:新增 gesvda 支持;

    • 增强计算库对 Driver/Runtime 调用错误信息的处理;

    • 优化 deconv 性能:部分场景性能提升40%;

    • 优化 LSTM 性能,部分场景性能提升30%;

互联加速库

重要

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

  • PCCL:

    • 兼容支持绝大多数 nccl api (ncclXXX) 和环境变量;

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

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

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

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

    • 支持新的 comm config 选项 ‘CTA_POLICY ’, 当设置为 '1' 时可节省 真武810E 下 CUs 资源使用,对计算/通信并行部分场景有性能提升;

    • 支持新的 comm config 选项 'maxCTAs', 可以 per comm 的将最大可用 channels 数调整为 '64';

    • 完成了 ncclMemAlloc/ncclMemFree/ncclCommRegister/ncclCommDeregister 等内存管理功能的初步实现; 当前为实验性支持, by default 关闭,尚需设置 'NCCL_CUMEM_ENABLE' 为 '1' 来打开使用 'IPC Register' 等功能;

    • 完成初步的 'symmetric memory API' 功能支持, 支持了 'pcclCommWindowRegister' 与 'pcclCommWindowRegister' 等新接口及 allreduce、allgather 与 reducescatter 等 kernels 实现;

    • 完成初步的 'CE collective' 功能支持, 可通过 per comm 的设置 'CTA_POLICY' 为 'PCCL_CTA_POLICY_ZERO' 来使用;

    • 完成了 'ncclCommInitRankScalable' 功能支持以优化大规模通信组下的 init 时间;

    • 完成对第三方 net plugin 上 UCX 功能的初步支持;

    • 支持 cuda launch mode 为 'blocking' 时的正常运行;

    • Debug 功能增强:

      • 支持 NCCL RAS ,目前 by default 为关掉,需显式设置NCCL_RAS_ENABLE为 ‘1’ 来打开使用;

      • 部分 'hang monitor' 功能增强;

    • pccl tools:

      • pccl perf:

        • 支持 AllReduce、AllGather、ReduceScatter、AlltoAll、Broadcast 及 Reduce;

      • p2pBandwidthAndLatencyPerf 工具增强,具体可见:

        • 支持真武810E 等产品在各种多卡互联 topo server 下的 icn p2p 互联带宽与延迟性能评估;

      • DeviceOrderSearch tool:

        • 用于 Megatron 框架模型训练任务上基于并行训练配置得到真武810E 机器上的最佳 visible device order 设置,具体使用可见:

      • pccl check tools:

        • 支持 真武810E 等产品在多卡或多机多卡等场景下的功能 readiness 检查;

      • sailbandwidth:

        • 完成首个版本支持, 关键 P2P 性能可达到预期水平;

  • sailSHMEM:

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

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

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

    • 支持 bitcode 编译;

    • 提供了若干单机内做卡间 bulk copy 的新 APIs 支持,相较原有功能 bw 性能可提升约 30%;

  • DeepEP:

    • 兼容支持 DeepEP 绝大多数 python API 用法;

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

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

    • 支持 low latency kernel 传输的 int8 groupwise 与 channelwise 量化实现;

    • Rebase 到社区 v2.0 版本;

    • 优化并提升了多机 combine kernels 上的性能;

    • 增强 deepep kernel nvtx info;

    • 修复了 intranode kernel 执行 random 出错的问题;

1.3.6 Video/Image codec硬件加速

兼容Nvidia Video Codec SDK,包括cuvid decode、nvenc(up to v13.0)和nvjpeg(up to v12.4);兼容NPP(up to v12.9)接口。基于此,可以直接支持ffmpeg(up to v7.1.1)、OpenCV4、DALI(up to v1.44)、PyAV、TorchCodec 等上层框架的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

  • Up to FHD 192 streams

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

  • Up to FHD 32 streams

JPEG:

  • Support nvjpeg decoder & encoder

  • Resolution up to 32Kx32K

  • Support RGB format input and output with inlinePP

  • Support crop, scale, rotate with inlinePP

  • Up to UHD 960FPS

Image Process:

  • Support Nvidia 2D Image NPP

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

    • PPU设备管理工具ppu-smi类似Nvidia-smi。详细介绍请参考设备管理工具PPU-SMI

    • ppu-smi v1.3 新功能主要包含:

      • 增加查询HBM厂家的描述

      • 增加设置和查询drain状态的描述

      • 增加查询机尾ID的描述

      • 增加topo子命令显示拓扑图支持按照机尾ID过滤和分组的描述

      • 增加topo子命令拓扑图显示RDMA网卡信息的描述

    • ppu-smi v1.4 新功能主要包含:

      • 增加查询XID错误码的描述

      • 增加查询performance counter状态的描述

      • 增加gpm子命令查询和设置GPM的描述

      • 增加查询潮汐模式的描述

      • 增加查询产品架构和Minor number的描述

      • 增加设置和查询MPS模式的描述

    • ppu-smi v1.5新功能主要包含:

      • 更新已知问题

      • 增加查询SDK各组件版本信息的描述

    • 数据中心管理和监控工具PPU DCG类似Nvidia DGCM,详情请参考管理监控工具DCGM

    • DCGM v1.3新功能主要包含:

      • 支持cache命中率相关field ID

      • 支持tensor core利用率相关field ID

    • DCGM v1.4新功能主要包含:

      • 支持ICN每链路收发速率相关field ID

      • 更新field ID支持情况列表

    • DCGM v1.5新功能主要包含:

      • 增加Docker环境使用DCGM的说明

    • DCGM v1.6 增强包括:

      • 支持整数计算pipe利用率相关field ID

    • DCGM v1.7新功能主要包括:

      • 增加nvbandwidth诊断测试程序的说明

      • 标明各子命令未支持的功能

      • 更新DCGM 4.2.3版本Field Id支持状态

    • hgml类似Nvidia nvml;

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

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

    • Asight Systems v1.3 新功能主要包含:

      • 支持 Python Functions Trace,在不修改用户 Python 代码的情况下对指定的 Python 函数进行 profile

      • 支持采集应用程序日志输出;支持采集设备内存申请失败跟踪

      • 新增专家系统和统计系统,支持15条分析规则

        • asys stats 支持 HGGC API / OSRT API / HGTX range / HGGC kernel 耗时统计和跟踪导出

        • asys stats 支持关联 HGTX 和 PPU 侧跟踪,输出统计报告和导出关联数据

      • Timeline View增强:

        • 支持HGGC Launch APITimeline View上按其启动的Kernel名字显示

        • User created streamTimeline View上支持显示blocking属性

        • Timeline View新增WASD快捷键缩放平移; 在kernel 类型的Timeline item上右键生成的acu命令不准确时提示用户

        • Top-Down Viewfunction table支持搜索module name

      • 高级打开功能支持指定更多事件类型(PCCL, Log, Backtrace, NIC metrics)

    • Asight Systems v1.4 新功能主要包含:

      • UI样式美化与布局调整,更多tab 样式用于切换;支持在报告打开过程中关闭报告标签页。

      • asys增强: asys支持指定CPUPython Profiling调用栈深度;asys支持指定采集应用程序部分进程的跟踪数据;asys支持采集PPU频率和温度等基本信息;asys支持采集CPU调用栈信息更短的采样周期;asys stats支持HGGC kernel grid block统计和跟踪导出;asys 支持HGGC Python backtrace调用栈采集;asys 支持 memory Python backtrace调用栈采集;asys python functions trace支持采集python进程下的所有线程。

      • Timeline View 增强: Timeline View支持显示HGTX自定义颜色;Timeline View支持filterPPU节点时间占比更新;Timeline View支持显示RDMA网卡metrics指标信息;Timeline View支持显示PPU Activity依赖关系;Timeline View支持独立显示HGGC Graph信息;Timeline View支持标记时间线;Timeline View 支持以不同的颜色显示不同类型的Video时间线;Timeline ViewHGGC Launch API支持以kernel名显示;Events View增加了平铺模式,增强搜索记录的自动填充;Function View增加了火焰图和冰川图;HGTX range汇总支持指定进程和线程列表。

    • Asight Systems v1.5新功能主要包含:

      • asys stats支持HGTX PPU projection summary统计功能

      • asys stats支持HGGC PPU memory跟踪数据量 / 耗时统计功能

      • asys stats子命令PPU kernel相关统计功能支持添加HGTX range前缀

      • asys stats子命令HGTX Range Summary支持按照进程过滤

      • asys stats子命令PPU跟踪相关统计功能支持指定统计的PPU device列表

      • asys stats子命令Device Memory Uasge Details / Summary支持2 / 4种查看模式

      • asys analyze子命令PPU时间利用率分析支持指定时间范围的选择模式

      • asys专家系统和统计系统description框说明文本简化,增加说明文档跳转链接

      • asys支持限制生成报告文件尺寸

      • asys支持通过--hggc-memory-usage选项使能采集devicepinned内存使用情况

      • 支持采集CPU侧动态分配内存的使用情况,支持内存泄露分析 / 内存使用量统计 / 申请次数统计

      • 支持采集MPS模式使能时HGGC context关联的计算资源信息

      • Memory View支持4种查看模式以及火焰图

      • Timeline View 优化PPU activity依赖关系的显示方式

      • Timeline View支持分组显示功能,允许将感兴趣的时间线行添加到Group View中并单独显示

      • Timeline View支持Device Memory Timeline,帮助用户分析各模块的内存的使用情况,支持三种模式:普通模式;分组模式;分组着色模式

      • Timeline View支持通信算子分层显示

      • Timeline View支持不同kernel以不同颜色显示

      • Timeline View支持CPUheap内存时间线显示

      • Time range tooltip中支持显示PPU active/idle时间占比

      • Timeline View 选择time range时按住Shift键支持吸附至最近的时间线起点/终点

      • Timeline View支持按shift/ctrl多选时间线行

      • Timeline View性能优化,解决了打开超大报告GUI卡顿的问题

      • HGGC API支持C调用栈和Python调用栈的混合显示

    • Asight Systems v1.6新功能主要包含:

      • asys支持CPU metrics跟踪采集

      • asys analyze子命令PPU time utilization分析功能支持指定统计的MPS模式Compute Pipe列表

      • asys支持采集Ray框架启动的多节点应用跟踪数据,每台服务器生成一份报告

      • asys支持采集mpirun启动的多节点应用跟踪数据,每台服务器生成一份报告

      • asys支持采集HGGC stream memory operation相关跟踪数据

      • asys stats子命令HGTX PPU Projection Trace / Summary支持按照HGTX range过滤

      • asys stats子命令HGTX PPU Projection Summary支持输出HGTX投影的时间占比和总的投影时间

      • asys stats子命令HGTX Range Kernel Trace / Summary支持按照HGTX range过滤

      • asys stats子命令HGTX Range Kernel Trace支持导出HGGC Graph相关信息

      • asys stats / analyze子命令支持指定ratio类数据输出精度

      • asys stats子命令支持分析PCCL通信kernel之间的不同步情况,支持导出跟踪和汇总数据

      • asys支持投影HGGC graph capture阶段HGTX rangePPUHGGC graph node对应的活动

      • asys支持通过事件控制capture range重复启停,多次capture range采集的跟踪写入同一份报告

      • asys支持通过指定PPU设备ID限制仅采集使用这些PPU设备的进程跟踪数据

      • Timeline View支持局部options,支持隐藏用户不感兴趣的行和定制报告的显示行为

      • Timeline View中支持kernel节点按kernel名字组织并展示

      • Timeline View添加全局tips,支持提示快捷操作信息

      • Timeline Viewtime range tooltip中单独计算PCCL的时间占比

      • CPU时间线支持按线程名称显示为不同的颜色

      • HGGC PPU Kernel Summary规则支持对TimeTotal Time两列数据占比进行自动求和

      • 统计系统新增PCCL Desynchronization Summary / Trace规则,分析PCCL kernel间的不同步情况

      • Bookmarks支持导入/导出和加载/保存

      • 统计系统/专家系统支持对表格进行搜索、过滤

    • Asight Systems v1.7新功能主要包含:

      • asys stats子命令支持导出PPU设备属性信息

      • asys stats子命令支持PPU operator summary统计功能,支持PPU算子分类统计,支持硬件利用率估算

      • asys stats子命令支持MoE operator summary统计功能,支持MoE算子中GEMM运算硬件利用率估算

      • asys stats子命令支持PCCL operator summary统计功能,支持PCCL算子的传输时间统计和带宽估算

      • asys statsasys analyze子命令支持指定输出的列过滤器,仅输出指定的列

      • asys stats子命令支持PPU metric trace跟踪导出功能,支持导出PPU metrics采样跟踪数据

      • asys stats子命令支持HGTX push pop range trace跟踪导出功能,支持导出HGTX range父子关系信息

      • asys支持采集PPU执行kernel / memcpy等使用的硬件队列索引

      • PPU Metrics跟踪功能PCIe总线收发数据量指标已包含PPU和其他外设(如RDMA网卡)的通信数据量

      • 查询PPU活动依赖关系支持查询HGGC graphPPU活动的依赖关系

      • 多报告打开时,允许按报告的起始时间自动对齐

      • Events View中支持跳转到目标行

      • 统一memory类型时间线的高度比例尺,方便定位OOM发生在哪个device/进程

      • Events view和专家/统计系统跳转行为优化,双击后可直接跳转至时间线

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

  • Asight Compute v1.3新功能主要包含:

    • Metric Details Widget支持显示metricinstance列表

    • Raw Page/Summary Page中支持选择多实例metric的显示格式,选择模式由非按行选中改为按单元格选中

    • 新增ICN link吞吐表格

    • acu 支持 --clock-control 选项

  • Asight Compute v1.4新功能主要包含:

    • 新增真武810E 16卡拓扑图显示

    • Source Page支持显示Source Marker,支持采集和显示Kernel的汇编指令的执行信息,包括Instructions ExecutedThread Instructions Executed

    • metrics修正和稳定性增强

  • Asight Compute v1.5新功能主要包含:

    • Details PageMemory Chart支持按transfer size/throughput显示

    • Details PageMemory Chart连接链路tooltip中显示链路利用率

    • Details Page中的tooltip支持pin住,pin住后支持选择文本拷贝

    • Details PageLaunch Statistics中增加 Stack Size Per Thread

    • Details PageLaunch Statistics中增加 Driver SharedMemory Per Block

    • Baseline按钮布局更新,明确各按钮功能

    • Source Page支持点击分支指令跳转,支持高亮所有搜索结果

    • Source Page增强resolve源代码,支持自动resolve同目录结构其它文件

  • Asight Compute v1.6新功能主要包含:

    • Source Page提示源码是否导入以及导入来源、提示如何采集lineinfo信息

    • Source Page、Details Page支持查看not issued stall reasons

    • Source Page提供Source Lookup入口按钮,提供“记住导入目录”选项

    • Source Page支持点击file not found提示打开源码导入窗口

    • Source Page支持Python、Pasm语法高亮

    • Source Page支持表头tooltip

    • Source Pagefunction列表按地址升序排序

    • Source Page支持显示报告附带的源码

    • Metric Details窗口对搜索到的metric进行排序

  • Asight Compute v1.7新功能主要包含:

    • 加强rule系统,支持用户自定义rules

    • 加强section tool窗口,新增rule文件显示

    • 打开报告时从报告中读取section文件,并按其定义展示metrics

    • 适配890P/890L ICN link拓扑图

    • 表格中多pass metrics超过界限时,显示警告标志

    • Source Page中支持列选择器,允许定制感兴趣列的显示

    • Source Pagetooltip显示样式增强,以图形化的方式显示stall reason占比

    • 添加了一些CU级的metrics,涵盖了CU active cycle,tensor指令,L1缓存的数据传输等信息,可以在Metrics Details中查看具体信息

    • 提升了低负载下利用率相关的metrics的准确性

    • 支持对更多类型的launch API启动的kernel进行跟踪

    • 表格中单元格间距调整,使其更便于查看

    • 其它GUI细节美化

1.4 支持的操作系统

类别

操作系统

架构

内核版本

GCC

Ubuntu

Ubuntu 24.04 LTS

x86_64

6.8.0-53-generic

13.3.0

Ubuntu 22.04 LTS

6.2.0-39-generic

11.4.0

Ubuntu 20.04 LTS

5.4.0-131-generic (GA)

9.5.0

5.4.0-92-generic

9.5.0

Ubuntu 18.04 LTS

4.15.0-112-generic (GA)

7.5.0

4.18.0-15-generic

7.5.0

1.5 SDK版本兼容性说明

1.5.1 KMD兼容性

V1.7 版本向前兼容V1.6.x、V1.5.x 与 V1.4.x 的 kmd。

1.5.2 Firmware兼容性

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

1.5.3 SDK兼容性

SDK V1.7 同 SDK V1.6、V1.5之间,保持在编程API接口、Device二进制格式、库文件的兼容性;同V1.5一样,对SDK V1.4在这些方面存在不兼容,具体不兼容的情况描述如下。

SDK v1.4 在文件、编程接口和编译环境上存在四类调整,会导致同之前SDK版本的组件或编译产生的组件共存情况下,存在不兼容的情形。相关的调整内容如下:

  • 将先前SDK中包含的libgomp.so、libomp.so文件移除,改为直接依赖相应OS环境以及host compiler(gcc/clang)所提供的libgomp.so、libomp.so库文件;

    此改动可能导致旧版程序执行在v1.5/v1.6/v1.7 SDK环境中存在找不到相应so文件的报错;

  • 在编译命令option层面,将“-alippu-xxxoption 选项值 修改为 “-ppu-xxx”的格式;

    这些options主要包含在一些定制化编译的场景中,此改动可能导致在混合版本的源代码编译或JIT运行过程中出现编译报错;

  • ptx扩展指令层面,有几类扩展ptx指令的调整:

    1. 将“alippu.mma.xxx”的ptx指令修改为“ppu.mma.xxx";

    2. 将“alippu.ldmatrix.xxx”的ptx指令修改为“ppu.ldmatrix.xxx";

    3. 将“cp.async.aiu.xxx”的ptx指令修改为“ppu.cp.async.aiu.xxx”;

  • SDK 头文件中自定义类型“__ali_bfloat16”修改为“__ppu_bfloat16”;

后面的三类改动会导致之前开源的部分PPU代码库存在编译问题(vllm/cutlass/flash-attn/flashinfer等),需要拉取适配v1.5及之后版本改动的代码库。

总的来说,以上四类调整会导致在源代码不更新、运行组件局部升级情况下存在编译/运行的兼容性问题,这在内部SDK兼容测试中也得到了体现,建议大家基于完整的v1.5/v1.6/v1.7 SDK环境来做编译和运行环境的部署。

1.5.4 asight文件兼容性

v1.7 版本兼容v1.5/v1.6的报告文件。

2. CUDA生态兼容

2.1 介绍

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

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

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

image

2.2 CUDA APIs兼容版本

截至PPU SDK V1.7版本,对CUDA APIs版本支持到12.9.0,但是限于硬件架构的区别,对CUDA APIs兼容主要在DeepLearning范畴,下表是目前支持的版本列表。更多CUDA兼容性内容请参见兼容性

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

在本次发布周期内CUDA最新版本已经升级到13.0,但我们考虑到CUDA 13.0相对于以前的版本来说有了很大的变化,甚至影响到NV原生场景下可执行程序的Legacy版本兼容性,特别体现在CUDA Runtime API定义中。对于PPU SDK来说存在更多额外的开发代价,我们计划在未来v2.1的版本来支持。

2.3 支持的开源框架/库

请参见:

3. 已知问题

3.1 驱动

不支持的CUDA API列表可参考CUDA APIs不支持列表

3.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),不影响编译,但运行时会报错;

3.3 加速库

  • 性能:性能泛化能力有待加强。

  • aublas:

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

    • 不支持复数数据类型。

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

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

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

  • audnn:

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

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

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

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

    • BN:1)仅支持alpha==1beta==0参数;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。

  • ausolver:

  • aufft:

  • aurand:

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

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

    • 仅支持Legacy order;

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

  • acsparse:

3.4 互联库

  • 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的值可帮助规避此问题;

  • NCCL comm register 及 window register 等功能需要显式设置 'NCCL_CUMEM_ENABLE' 来使用;目前打开 'CUMEM' 支持后会有分配 device mem 膨胀的问题,会在下个版本修复;

3.5 Video Codec/Image硬件加速

  • Video decode不支持MPEG1,MPEG2,MPEG4,VC1,VP8legacy格式;

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

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

3.6 工具

3.7. 框架与模型

开源框架已知问题说明:

框架

版本

已知问题

Pytorch

2.6

  • 兼容性Unit Test测试Fail Case分类:

    • CUDA API 暂不支持

    • "预期结果,包括:

      • A100 FAIL

      • 测试中Hardcode NV binary的名字

      • 硬件SFU导致的差异等

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

    • 运行超时time out

    • 已知issue

2.7

  • SDPA mem_efficient后端计算结果non-deterministic(A100也存在该问题)

  • SDPA mem_efficient后端failed 1 UT (A100也存在该问题)

  • SDPA 新增后端backend cudnn_attention, 调用了cudnn里的FA实现 目前PPU还未支持,用export TORCH_CUDNN_SDPA_ENABLED=1控制是否打开这个backend ,默认关闭

  • 新增一些HGGC api不支持接口,如cuptiActivityEnableRuntimeApi

  • 加速库优化LSTM性能,gemm使用TF32精度导致部分UT case(test_data_parallel.py ) regression fail,A100H20同样fail。

2.8

  • 兼容性Unit Test测试Fail Case分类:

    1. CUDA API 暂不支持

    2. 预期结果:

      1. A100 FAIL

      2. 测试中Hardcode NV binary的名字

      3. 硬件SFU导致的差异 等

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

    4. 运行超时time out

    5. 已知issue,包含带dropout mem_efficient attention结果差异

SGlang

0.4.3

  • 兼容性Unit Test测试Fail Case分类:

    • 个别模型perf UT fail,NV GPU也同样fail

    • 社区已知问题:不支持运行GPTQ量化的模型

0.4.7

  • SGLang-V0.4.7在运行DeepSeek-R1 \ DeepSeek-V3模型时,默认会使用shared expert,SGLang-V0.4.6.post1则不会,这会导致SGLang-V0.4.7运行该模型时性能不如SGLang-V0.4.6.post1。SGLang-V0.4.7可通过--disable-shared-experts-fusion禁用shared expert,此时SGLang-V0.4.7运行该模型时性能与SGLang-V0.4.6.post1相当;

  • SGLang-V0.4.7--mem-fraction-static设置的较大导致单机显存容量不足时,会出现decode out of memory的问题,例如真武810E单机运行DeepSeek-R1。这是社区已知问题;

0.5.2

  • DeepGemm为运行时编译,测试性能时请确保warmup充分

  • DeepSeek v3.2目前必须打开dp-attention,尚不支持tp并行

  • Qwen-235B-A22B短输出(output_len=400)有性能回退。

  • DeepSeek-R1 in900, out400 SLA测试场景最大并发未达到最大吞吐。

VLLM

0.7.3

  • 量化问题

    • VLLM默认使用Marlin KernelGPTQ/AWQ/W8A8/WOQ进行量化推理加速,可以通过显式指定quantization的方法来使用vllm原本的Marlin/W8A8/awq/gptq等量化推理kernel,例如可以指定使用gptq_acext/awq_acext来做量化推理,但是acext目前不支持act_order=True的量化,以及GPTQ-Int8量化权重,下个版本优化

    • VLLM FP8量化依赖SM89/90架构,没有进行PPU适配与优化,支持计划待定

    • VLLM量化Dense A8W8对接acext库实现,性能需要持续优化,MOE W8A8量化暂时不支持,预计下一版支持

    • 支持DPSK FP8权重load并按照W8A16 量化推理,性能与BF16相当

    • VLLM 默认使用 vllm-flash-attn 作为attention的后端,该性能在PPU上尚未超过GPU版本,下个版本优化

  • 稳定性问题

0.9.1

  • 量化问题

    • VLLM 默认使用 Marlin Kernel 对 GPTQ/AWQ 进行量化推理加速,可以通过显式指定 quantization gptq_acext/awq_acext 使用 acext 来做量化推理,但是 acext 目前不支持 act_order=True 的量化,以及 GPTQ-Int8 量化权重

    • VLLM FP8 量化依赖 SM89/90 架构,PPU 暂无原生支持;VLLM 0.9.1 提供了FP8 marlin 的实现方式(实际是 W8A16 量化推理),可在 SM < 89 架构上运行 FP8 模型,PPU 可通过这种方式运行 FP8,但暂未进行适配与优化

0.10.2

  • 性能问题

    • Qwen 模型由于 fused moe 中 fused_topk 没有使用 torch.compile 包裹,导致模型性能相比前序版本有回退,开源社区上已经有 patch 修复,预计在 vllm 0.11.1 上合入 https://github.com/vllm-project/vllm/pull/26717

0.11.0

  • 本版本属于定向测试模型预览版本,有潜在的稳定性风险

TransformerEngine

2.0

  • 兼容性Unit Test测试Fail Case分类:

    • TE Context Parallelism UT test_fused_attn_with_cp.py::test_cp_with_flash_attention 在torch 2.1.0版本环境下运行会有FAIL,建议使用torch 2.1.2以上的版本

    • TE numeric UT存在FAIL CASE

    • 不支持cublasLt fused_attention

    • A100同样failcase

  • Pai-Megatron-Patch暂时不支持TE v2.0,直接使用可能出现潜在bug

flashinfer

0.2.6.post1

  • 以下模块存在已知的精度问题,导致UT fail

    1. PoD Attention

    2. Segment gemm