| Files | Last commit | Last update |
|---|---|---|
feat(quant): add per_token_cast torch.compile support Co-authored-by: daisysj<songjing17@huawei.com> # message auto-generated for no-merge-commit merge: !53 merge graph_adapt into master feat(quant): add per_token_cast torch.compile support Created-by: daisysj Commit-by: daisysj Merged-by: cann-robot Description: <!-- 感谢您的合入申请! --> ### 当前PR是否有AI参与: [x] 否 [ ] 是 __1. AI Agent 平台: __2. AI 模型: __3. Prompt上下文 : ### PR功能描述 / 为什么需要这个合入**: <!-- 本 PR 做了什么,为什么需要 / What does this PR do and why --> 为 per_token_cast、per_token_cast_with_sf_only 和 per_token_cast_with_precomputed_sf 添加 torch.compile 入图支持,使它们能够参与 npugraph_ex 固定 shape 推理图的捕获与重放。 适配同时覆盖自动分配输出和调用方提供 out 的路径,保持原有 eager 调用方式,确保输入及只读 scale 不被修改、输出正确复用指定内存,并能感知 clamp 配置变化。 ### 该PR关联的issue *(格式为fixes #<issue号>, 或者resolves #<issue号>)*: fixes # ### 希望检视人员了解: ## 改动类型 / Change Type - [ ] Bug 修复 / Bug Fix - [x] 新功能 / New Feature - [ ] 性能优化 / Performance - [ ] 代码重构 / Refactoring - [ ] 文档更新 / Documentation - [ ] 测试相关 / Test - [ ] 其它 / Other ## 测试信息 / Testing <!-- 简要测试说明或关键结果 / Brief test description or key results --> - [ ] 单元测试通过 / UT passed - [ ] 集成测试通过 / ST passed - [ ] 人工验证通过 / Manual verified ## 检查清单 / Checklist - [ ] 代码符合规范 / Code follows style guide - [ ] 测试添加并通过 / Tests added and passed - [ ] 文档已更新 / Docs updated if needed - [ ] 无硬编码敏感信息 / No secrets hardcoded - [ ] 提交信息符合规范 / Commit message follows convention See merge request: cann/ops-tilelang!53 | 9 days ago | |
Add engram_gate_bwd/fwd torch.compile support and torch.compile skill. Co-authored-by: hanxu88<hanxu88@huawei.com> # message auto-generated for no-merge-commit merge: !49 merge graph into master Add engram_gate_bwd/fwd torch.compile support and torch.compile skill. Created-by: hanxu88 Commit-by: hanxu88 Merged-by: cann-robot Description: <!-- 感谢您的合入申请! --> ### 当前PR是否有AI参与: [x] 否 [ ] 是 __1. AI Agent 平台: __2. AI 模型: __3. Prompt上下文 : ### PR功能描述 / 为什么需要这个合入**: 以 engram_gate_fwd 和 engram_gate_bwd 为例,增加 torch.compile 支持,通过 npugraph_ex 实现固定形状的 ACL Graph 图接入; 另外,修改: 1. 拆分真实执行与元信息校验,注册 custom_op 和 fake 实现,避免追踪阶段执行环境检查或编译内核; 2. 适配前向可选统计量返回,以及反向依赖硬件核数的权重梯度分片形状推导; 3. 同时支持 eager 和 compile 调用与测试,覆盖统计量开关、空输入、原存储更新和新输入替换,并统一前后向图执行的 inference_mode 上下文。 此外,增加 ops-tilelang-graph-mode skill,指导 agent 为本仓库其它算子实现 torch.compile / ACL Graph 入图改造。 ### 该PR关联的issue 无。 ### 希望检视人员了解: ## 改动类型 / Change Type - [ ] Bug 修复 / Bug Fix - [x] 新功能 / New Feature - [ ] 性能优化 / Performance - [ ] 代码重构 / Refactoring - [x] 文档更新 / Documentation - [ ] 测试相关 / Test - [ ] 其它 / Other ## 测试信息 / Testing <!-- 简要测试说明或关键结果 / Brief test description or key results --> - [x] 单元测试通过 / UT passed - [x] 集成测试通过 / ST passed - [x] 人工验证通过 / Manual verified ## 检查清单 / Checklist - [x] 代码符合规范 / Code follows style guide - [x] 测试添加并通过 / Tests added and passed - [x] 文档已更新 / Docs updated if needed - [x] 无硬编码敏感信息 / No secrets hardcoded - [x] 提交信息符合规范 / Commit message follows convention See merge request: cann/ops-tilelang!49 | 10 days ago | |
feat(mhc): migrate operator suite Co-authored-by: wangliwei12<wangliwei12@huawei.com> # message auto-generated for no-merge-commit merge: !11 merge feat/mhc-operators into master feat(mhc): migrate operator suite Created-by: wangliwei12 Commit-by: wangliwei12 Merged-by: cann-robot Description: 当前PR是否有AI参与: [x] 否 [ ] 是 __1. AI Agent 平台: __2. AI 模型: __3. Prompt上下文 : PR功能描述 / 为什么需要这个合入**: 该PR关联的issue (格式为fixes #<issue号>, 或者resolves #<issue号>): fixes # 希望检视人员了解: 改动类型 / Change Type Bug 修复 / Bug Fix [ ✅] 新功能 / New Feature 性能优化 / Performance 代码重构 / Refactoring 文档更新 / Documentation 测试相关 / Test 其它 / Other 测试信息 / Testing [ ✅] 单元测试通过 / UT passed [ ✅] 集成测试通过 / ST passed [ ]✅ 人工验证通过 / Manual verified 检查清单 / Checklist [ ✅] 代码符合规范 / Code follows style guide [ ✅] 测试添加并通过 / Tests added and passed [ ✅] 文档已更新 / Docs updated if needed [ ✅] 无硬编码敏感信息 / No secrets hardcoded [ ✅] 提交信息符合规范 / Commit message follows convention See merge request: cann/ops-tilelang!11 | 1 month ago | |
feat(moe): migrate gate operators Co-authored-by: wangliwei12<wangliwei12@huawei.com> # message auto-generated for no-merge-commit merge: !9 merge feat/moe-gate-operators into master feat(moe): migrate gate operators Created-by: wangliwei12 Commit-by: wangliwei12 Merged-by: cann-robot Description: 当前PR是否有AI参与: [x] 否 [ ] 是 __1. AI Agent 平台: __2. AI 模型: __3. Prompt上下文 : PR功能描述 / 为什么需要这个合入**: 该PR关联的issue (格式为fixes #<issue号>, 或者resolves #<issue号>): fixes # 希望检视人员了解: 改动类型 / Change Type Bug 修复 / Bug Fix [ ✅] 新功能 / New Feature 性能优化 / Performance 代码重构 / Refactoring 文档更新 / Documentation 测试相关 / Test 其它 / Other 测试信息 / Testing [ ✅] 单元测试通过 / UT passed [ ✅] 集成测试通过 / ST passed [ ]✅ 人工验证通过 / Manual verified 检查清单 / Checklist [ ✅] 代码符合规范 / Code follows style guide [ ✅] 测试添加并通过 / Tests added and passed [ ✅] 文档已更新 / Docs updated if needed [ ✅] 无硬编码敏感信息 / No secrets hardcoded [ ✅] 提交信息符合规范 / Commit message follows convention See merge request: cann/ops-tilelang!9 | 1 month ago | |
fix(sampling): add ops Co-authored-by: wangliwei12<wangliwei12@huawei.com> # message auto-generated for no-merge-commit merge: !20 merge feat/vllm-operators into master fix(sampling): add ops Created-by: wangliwei12 Commit-by: wangliwei12 Merged-by: cann-robot Description: <!-- 感谢您的合入申请! --> ### 当前PR是否有AI参与: [x] 否 [ ] 是 __1. AI Agent 平台: __2. AI 模型: __3. Prompt上下文 : ### PR功能描述 / 为什么需要这个合入**: <!-- 本 PR 做了什么,为什么需要 / What does this PR do and why --> ### 该PR关联的issue *(格式为fixes #<issue号>, 或者resolves #<issue号>)*: fixes # ### 希望检视人员了解: ## 改动类型 / Change Type - [ ] Bug 修复 / Bug Fix - [ ] 新功能 / New Feature - [ ] 性能优化 / Performance - [ ] 代码重构 / Refactoring - [ ] 文档更新 / Documentation - [ ] 测试相关 / Test - [ ] 其它 / Other ## 测试信息 / Testing <!-- 简要测试说明或关键结果 / Brief test description or key results --> - [ ] 单元测试通过 / UT passed - [ ] 集成测试通过 / ST passed - [ ] 人工验证通过 / Manual verified ## 检查清单 / Checklist - [ ] 代码符合规范 / Code follows style guide - [ ] 测试添加并通过 / Tests added and passed - [ ] 文档已更新 / Docs updated if needed - [ ] 无硬编码敏感信息 / No secrets hardcoded - [ ] 提交信息符合规范 / Commit message follows convention See merge request: cann/ops-tilelang!20 | 1 month ago | |
feat(quant): add per_token_cast torch.compile support Co-authored-by: daisysj<songjing17@huawei.com> # message auto-generated for no-merge-commit merge: !53 merge graph_adapt into master feat(quant): add per_token_cast torch.compile support Created-by: daisysj Commit-by: daisysj Merged-by: cann-robot Description: <!-- 感谢您的合入申请! --> ### 当前PR是否有AI参与: [x] 否 [ ] 是 __1. AI Agent 平台: __2. AI 模型: __3. Prompt上下文 : ### PR功能描述 / 为什么需要这个合入**: <!-- 本 PR 做了什么,为什么需要 / What does this PR do and why --> 为 per_token_cast、per_token_cast_with_sf_only 和 per_token_cast_with_precomputed_sf 添加 torch.compile 入图支持,使它们能够参与 npugraph_ex 固定 shape 推理图的捕获与重放。 适配同时覆盖自动分配输出和调用方提供 out 的路径,保持原有 eager 调用方式,确保输入及只读 scale 不被修改、输出正确复用指定内存,并能感知 clamp 配置变化。 ### 该PR关联的issue *(格式为fixes #<issue号>, 或者resolves #<issue号>)*: fixes # ### 希望检视人员了解: ## 改动类型 / Change Type - [ ] Bug 修复 / Bug Fix - [x] 新功能 / New Feature - [ ] 性能优化 / Performance - [ ] 代码重构 / Refactoring - [ ] 文档更新 / Documentation - [ ] 测试相关 / Test - [ ] 其它 / Other ## 测试信息 / Testing <!-- 简要测试说明或关键结果 / Brief test description or key results --> - [ ] 单元测试通过 / UT passed - [ ] 集成测试通过 / ST passed - [ ] 人工验证通过 / Manual verified ## 检查清单 / Checklist - [ ] 代码符合规范 / Code follows style guide - [ ] 测试添加并通过 / Tests added and passed - [ ] 文档已更新 / Docs updated if needed - [ ] 无硬编码敏感信息 / No secrets hardcoded - [ ] 提交信息符合规范 / Commit message follows convention See merge request: cann/ops-tilelang!53 | 9 days ago | |
feat(quant): support torch.compile for per-block quantization ops | 2 days ago | |
chore: remove operators implementation Co-authored-by: wangliwei12<wangliwei12@huawei.com> # message auto-generated for no-merge-commit merge: !25 merge chore/remove-initial-vllm-operators into master chore: remove operators implementation Created-by: wangliwei12 Commit-by: wangliwei12 Merged-by: cann-robot Description: <!-- 感谢您的合入申请! --> ### 当前PR是否有AI参与: [x] 否 [ ] 是 __1. AI Agent 平台: __2. AI 模型: __3. Prompt上下文 : ### PR功能描述 / 为什么需要这个合入**: <!-- 本 PR 做了什么,为什么需要 / What does this PR do and why --> ### 该PR关联的issue *(格式为fixes #<issue号>, 或者resolves #<issue号>)*: fixes # ### 希望检视人员了解: ## 改动类型 / Change Type - [ ] Bug 修复 / Bug Fix - [ ] 新功能 / New Feature - [ ] 性能优化 / Performance - [ ] 代码重构 / Refactoring - [ ] 文档更新 / Documentation - [ ] 测试相关 / Test - [ ] 其它 / Other ## 测试信息 / Testing <!-- 简要测试说明或关键结果 / Brief test description or key results --> - [ ] 单元测试通过 / UT passed - [ ] 集成测试通过 / ST passed - [ ] 人工验证通过 / Manual verified ## 检查清单 / Checklist - [ ] 代码符合规范 / Code follows style guide - [ ] 测试添加并通过 / Tests added and passed - [ ] 文档已更新 / Docs updated if needed - [ ] 无硬编码敏感信息 / No secrets hardcoded - [ ] 提交信息符合规范 / Commit message follows convention See merge request: cann/ops-tilelang!25 | 1 month ago | |
fix(sampling): add ops Co-authored-by: wangliwei12<wangliwei12@huawei.com> # message auto-generated for no-merge-commit merge: !20 merge feat/vllm-operators into master fix(sampling): add ops Created-by: wangliwei12 Commit-by: wangliwei12 Merged-by: cann-robot Description: <!-- 感谢您的合入申请! --> ### 当前PR是否有AI参与: [x] 否 [ ] 是 __1. AI Agent 平台: __2. AI 模型: __3. Prompt上下文 : ### PR功能描述 / 为什么需要这个合入**: <!-- 本 PR 做了什么,为什么需要 / What does this PR do and why --> ### 该PR关联的issue *(格式为fixes #<issue号>, 或者resolves #<issue号>)*: fixes # ### 希望检视人员了解: ## 改动类型 / Change Type - [ ] Bug 修复 / Bug Fix - [ ] 新功能 / New Feature - [ ] 性能优化 / Performance - [ ] 代码重构 / Refactoring - [ ] 文档更新 / Documentation - [ ] 测试相关 / Test - [ ] 其它 / Other ## 测试信息 / Testing <!-- 简要测试说明或关键结果 / Brief test description or key results --> - [ ] 单元测试通过 / UT passed - [ ] 集成测试通过 / ST passed - [ ] 人工验证通过 / Manual verified ## 检查清单 / Checklist - [ ] 代码符合规范 / Code follows style guide - [ ] 测试添加并通过 / Tests added and passed - [ ] 文档已更新 / Docs updated if needed - [ ] 无硬编码敏感信息 / No secrets hardcoded - [ ] 提交信息符合规范 / Commit message follows convention See merge request: cann/ops-tilelang!20 | 1 month ago | |
feat(quant): add graph support for cast operators and fix FP4 unpack replay | 3 days ago | |
ci: 自动发现增量测试并移除算子映射表 Co-authored-by: lwwangcgz<wlw9393@gmail.com> # message auto-generated for no-merge-commit merge: !48 merge master into master ci: 自动发现增量测试并移除算子映射表 Created-by: wangliwei12 Commit-by: lwwangcgz Merged-by: cann-robot Description: <!-- 感谢您的合入申请! --> ### 当前PR是否有AI参与: [x] 否 [ ] 是 __1. AI Agent 平台: __2. AI 模型: __3. Prompt上下文 : ### PR功能描述 / 为什么需要这个合入**: 当前增量测试依赖手工维护算子与测试的对应关系,新增算子时还需要修改 .gitcode/ 配置,容易遗漏。 本 PR 新增基于 Python AST 的测试选择器,通过源码和测试之间的直接、间接导入关系自动发现受影响用例,并接入现有 CI 执行脚本。新增算子和测试无需登记映射。 同时,移除 .gitcode/* 的默认全量触发规则,并修正选择器兜底逻辑:仅修改 CI 文件时跳过算子用例;同时修改源码或测试时正常增量选测。保留管理员在机器故障等场景调整执行逻辑的通道。 ### 该PR关联的issue *(格式为fixes #<issue号>, 或者resolves #<issue号>)*: fixes # ### 希望检视人员了解: ## 改动类型 / Change Type - [❌ ] Bug 修复 / Bug Fix - [ ❌] 新功能 / New Feature - [❌ ] 性能优化 / Performance - [ ❌] 代码重构 / Refactoring - [❌ ] 文档更新 / Documentation - [ ✅] 测试相关 / Test - [ ❌] 其它 / Other ## 测试信息 / Testing <!-- 简要测试说明或关键结果 / Brief test description or key results --> - [✅ ] 单元测试通过 / UT passed - [ ✅] 集成测试通过 / ST passed - [ ✅] 人工验证通过 / Manual verified ## 检查清单 / Checklist - [✅ ] 代码符合规范 / Code follows style guide - [✅ ] 测试添加并通过 / Tests added and passed - [ ✅] 文档已更新 / Docs updated if needed - [ ✅] 无硬编码敏感信息 / No secrets hardcoded - [ ✅] 提交信息符合规范 / Commit message follows convention See merge request: cann/ops-tilelang!48 | 16 days ago | |
本 PR 初始化 ops-tilelang 高性能算子仓,为 CANN 社区提供基于 TileLang DSL 的 Ascend 算子开发、测试、性能回归和 wheel 发布基础框架。 项目采用运行时接口与硬件 kernel 分层设计,当前仅支持 Ascend A5 及以上架构。 Co-authored-by: ChaoyangJi<jichaoyang@huawei.com> # message auto-generated for no-merge-commit merge: !2 merge master into master add ops-tilelang init framework Created-by: ChaoyangJi Commit-by: ChaoyangJi Merged-by: cann-robot Description: <!-- 感谢您的合入申请! --> ### 当前PR是否有AI参与: 无。 ### PR功能描述 / 为什么需要这个合入**: <!-- 本 PR 做了什么,为什么需要 / What does this PR do and why --> 本 PR 初始化 ops-tilelang 高性能算子仓,为 CANN 社区提供基于 TileLang DSL 的 Ascend 算子开发、测试、性能回归和 wheel 发布基础框架。 项目采用运行时接口与硬件 kernel 分层设计,当前仅支持 Ascend A5 及以上架构。 主要改动 1. 基础工程框架 - 建立 src/ layout Python 工程。 - 发行包名为 ops-tilelang,Python 命名空间为 cann_ops_tilelang。 - 使用 setuptools、setuptools-scm 构建 wheel。 - 支持 Python 3.10、3.11 和 3.12。 - 运行时依赖 torch>=2.1 和 tilelang>=0.1.9。 - torch_npu、CANN Toolkit 和驱动由目标环境提供。 - License 使用 MIT。 - 包导入阶段不初始化 NPU、不编译 kernel、不申请设备内存。 2. Engram 算子 首批提供四个公开 API: - engram_fused_weight - BF16 权重逐元素融合,使用 FP32 输出。 - engram_gate_fwd - 支持动态 token 数。 - 输出 gate 结果及反向所需统计量。 - engram_gate_bwd - 融合计算输入梯度、KV 梯度和分片权重梯度。 - engram_grad_w_reduce - 归并分片权重梯度并原地累加两份 FP32 权重梯度。 算子采用: - PyTorch wrapper 与 *_kernel.py 分层。 - tilelang.jit kernel generator。 - GM/UB 分块、SimdVF、Persistent 调度及多级流水。 - 明确的 device、dtype、shape、stride、contiguous 和输出缓冲区校验。 - 无 CPU/CUDA 静默回退。 3. 测试与性能回归框架 参考 TileKernels-Nightly 迁移并适配完整 pytest 能力: - OPS_TILELANG_TEST_LEVEL=0|1|2 分级参数生成。 - 基于 --seed 和 node ID 的确定性随机数。 - PyTorch reference 数值验证。 - NPU 峰值显存画像和 large_npu_mem 自动分类。 - A5 benchmark JSONL 基线及回归检测。 - -K 参数表达式筛选。 - benchmark improvement 几何平均汇总。 - xdist 多卡绑定及全局 fail-fast。 - TileLang 编译告警汇总。 - 普通显存与大/未知显存测试分阶段执行。 4. 构建、CI 与脚本 新增统一执行入口: - scripts/check.sh - scripts/build_wheel.sh - scripts/run_test.sh - scripts/run_tests.sh - scripts/run_ci.sh scripts/run_ci.sh 支持: - lint - cpu - wheel - npu - benchmark - all 新增 GitHub workflows: - Ruff lint 检查。 - Wheel 构建、安装及元数据验证。 5. 文档 新增并完善: - README:安装、使用、测试、CI 和硬件支持范围。 - 架构说明。 - 完整的新增算子开发流程。 - pytest 插件、benchmark 和显存画像维护指南。 - CONTRIBUTING 贡献规范。 - MIT License。 ### 该PR关联的issue 无。 ### 希望检视人员了解: ops-tilelang的框架代码是否合理和完善。 ## 改动类型 / Change Type - [ ] Bug 修复 / Bug Fix - [ √] 新功能 / New Feature - [ ] 性能优化 / Performance - [ ] 代码重构 / Refactoring - [ ] 文档更新 / Documentation - [ ] 测试相关 / Test - [ ] 其它 / Other ## 测试信息 / Testing <!-- 简要测试说明或关键结果 / Brief test description or key results --> - [√] 单元测试通过 / UT passed - [ √] 集成测试通过 / ST passed - [ √] 人工验证通过 / Manual verified ## 检查清单 / Checklist - [√] 代码符合规范 / Code follows style guide - [ √] 测试添加并通过 / Tests added and passed - [ √] 文档已更新 / Docs updated if needed - [√ ] 无硬编码敏感信息 / No secrets hardcoded - [ √] 提交信息符合规范 / Commit message follows convention See merge request: cann/ops-tilelang!2 | 1 month ago | |
feat(quant): migrate quant operators Co-authored-by: wangliwei12<wangliwei12@huawei.com> # message auto-generated for no-merge-commit merge: !16 merge feat/quant-operators into master feat(quant): migrate quant operators Created-by: wangliwei12 Commit-by: wangliwei12 Merged-by: cann-robot Description: 当前PR是否有AI参与: [x] 否 [ ] 是 __1. AI Agent 平台: __2. AI 模型: __3. Prompt上下文 : PR功能描述 / 为什么需要这个合入**: 该PR关联的issue (格式为fixes #<issue号>, 或者resolves #<issue号>): fixes # 希望检视人员了解: 改动类型 / Change Type Bug 修复 / Bug Fix [ ✅] 新功能 / New Feature 性能优化 / Performance 代码重构 / Refactoring 文档更新 / Documentation 测试相关 / Test 其它 / Other 测试信息 / Testing [ ✅] 单元测试通过 / UT passed [ ✅] 集成测试通过 / ST passed [ ]✅ 人工验证通过 / Manual verified 检查清单 / Checklist [ ✅] 代码符合规范 / Code follows style guide [ ✅] 测试添加并通过 / Tests added and passed [ ✅] 文档已更新 / Docs updated if needed [ ✅] 无硬编码敏感信息 / No secrets hardcoded [ ✅] 提交信息符合规范 / Commit message follows convention See merge request: cann/ops-tilelang!16 | 1 month ago | |
fix: 省略可选bandwidth_gbs的合法benchmark在会话报告阶段崩溃 Co-authored-by: Ethan_Zou<tianjun.zou@foxmail.com> # message auto-generated for no-merge-commit merge: !37 merge fix/tilelang-optional-benchmark-bandwidth-20261014 into master fix: 省略可选bandwidth_gbs的合法benchmark在会话报告阶段崩溃 Created-by: Ethan_Zou Commit-by: Ethan_Zou Merged-by: cann-robot Description: ### 当前 PR 是否有 AI 参与 - [ ] 否 - [x] 是 1. AI Agent 平台:OpenAI Codex(桌面运行环境) 2. AI 模型:GPT-6 系列 3. Prompt 上下文:核查真实 pytest 插件和 JSONL 文件路径,按真人评审移除新增 Host 测试文件,并保留真实修复前后复现证据。 ### PR 功能描述 / 为什么需要这个合入 benchmark_record签名默认bandwidth_gbs=None,JSONL schema标注该字段可选,记录入口也只在传入时写入;但matched基线分支的终端报告使用rec[bandwidth_gbs],合法省略字段的成功测试会话最终KeyError退出。 终端报告使用.get读取可选带宽,并沿用已有_fmt_bw(None)占位;实际自测覆盖matched/missing基线与传入/省略带宽四种组合,不影响基线或记录字段。 ### 该 PR 关联的 Issue Fixes #9 https://gitcode.com/cann/ops-tilelang/issues/9 ### 希望检视人员了解 在精确基线 af459ebfa36cf7f15532cc63385e8d4289704adb 与同一最终回归文件上,真实 pytest 原版 1 failed、3 passed;最终 4 passed,0 skip。完整原始插件 record/terminal-summary 路径和实际嵌套pytest会话,matched/missing基线与省略/传入带宽四种组合;记录可选字段及基线字节核对。 公开 CPU 复现(基线或修复仓库根目录): bash python3 -m pip install pytest "torch==2.7.1" PYTEST_DISABLE_PLUGIN_AUTOLOAD=1 TORCH_DEVICE_BACKEND_AUTOLOAD=0 python3 - <<'PY' import ast import importlib.util from pathlib import Path import sys import types import pytest root = Path.cwd() import json import tempfile directory_owner = tempfile.TemporaryDirectory(prefix="tilelang-benchmark-host-") directory = Path(directory_owner.name) (directory / "pytest.ini").write_text("[pytest]\n") record = dict(kernel="host_fixture", operation="record", params={}, param_id=None, cluster="a5", time_us=1.0, bandwidth_gbs=1.0) baseline = directory / "benchmark/test_cpu_refresh.jsonl" baseline.parent.mkdir() original = json.dumps(record) + "\n" baseline.write_text(original) (directory / "test_cpu_refresh.py").write_text( 'def test_record(benchmark_record):\n' ' benchmark_record(kernel="host_fixture", operation="record", params={}, time_us=1.0)\n') sys.path[:0] = [str(root), str(root / "src")] # Isolate unrelated package kernel imports; every exercised plugin method remains real. package = types.ModuleType("cann_ops_tilelang") package.__path__ = [str(root / "src/cann_ops_tilelang")] package.__spec__ = importlib.util.spec_from_loader(package.__name__, loader=None, is_package=True) sys.modules[package.__name__] = package path = root / "tests/pytest_benchmark_plugin.py" tree = ast.parse(path.read_text()) excluded = [node for node in tree.body if isinstance(node, ast.ImportFrom) and node.module == "tilelang.profiler.bench"] assert len(excluded) == 1 # The timer backend is unavailable without TileLang/NPU and is never called here. tree.body = [node for node in tree.body if node not in excluded] module = types.ModuleType("host_benchmark_plugin") module.__file__ = str(path) sys.modules[module.__name__] = module exec(compile(tree, str(path), "exec"), vars(module)) setattr(module, "_TESTS_DIR", str(directory)) setattr(module, "_BASELINES_LAYOUT", str(directory / "benchmark")) args = [str(directory / "test_cpu_refresh.py"), "-q", "-c", str(directory / "pytest.ini"), "--confcutdir=" + str(directory), "--run-benchmark", "--benchmark-output=" + str(directory / "results.jsonl")] try: status = pytest.main(args, plugins=[module]) except KeyError as exc: print("KeyError:", str(exc)) status = 1 print("session_exit=", int(status)) print("baseline_time_us=", json.loads(baseline.read_text())["time_us"]) directory_owner.cleanup() PY 公开复现原版成功测试尾部 KeyError(bandwidth_gbs),session_exit=1;修复后 session_exit=0,基线字节不变。 已按 ChaoyangJi 的评审意见移除新增 Host 测试文件,最终源码只保留原生产修复。原真实回归的完整文件、命令和结果已归档到提交之外;上方内联 CPU 命令仍可在精确基线与当前修复源码直接复现,公开复现不依赖归档文件。 只排除一个未调用的 tilelang.profiler.bench 后端导入,并隔离无关 kernel 包初始化;执行的插件源码/pytest hooks/fixture/JSONL文件操作均为真实实现。受控时间字段用于Host文件/报告行为验证,不是实测NPU性能;没有模拟NPU计时,没有调用benchmark_timer。最终实际仓库 Ruff 检查/格式化、diff --check通过。 本地无 NPU;仅核查实际源码及 Host 行为,未运行 CANN/NPU 编译、设备精度或性能测试。 真实自测:aarch64 Host、Python 3.11.4、PyTorch 2.7.1+cpu、NumPy 1.26.4、pytest 8.3.2。执行原始真实回归时设置 PYTEST_DISABLE_PLUGIN_AUTOLOAD=1、TORCH_DEVICE_BACKEND_AUTOLOAD=0,并采用上方说明的 kernel 初始化/未调用 timer 导入隔离。当前保留的生产源码与通过回归的源码逐字相同。原项目完整 pre-commit 及 Ruff 0.14.14 检查/格式化、Git diff/object/LFS 检查通过。独立补充 OAT 实际扫描插件 1 文件,基线与最终均报告该既有文件缺少许可证头、版权头两项;未声称 OAT 通过,原许可、头部和检查规则保持。 ### 改动类型 / Change Type - [x] Bug 修复 - [ ] 测试补充(原回归已移出源码并保留为自测证据) See merge request: cann/ops-tilelang!37 | 28 days ago | |
本 PR 初始化 ops-tilelang 高性能算子仓,为 CANN 社区提供基于 TileLang DSL 的 Ascend 算子开发、测试、性能回归和 wheel 发布基础框架。 项目采用运行时接口与硬件 kernel 分层设计,当前仅支持 Ascend A5 及以上架构。 Co-authored-by: ChaoyangJi<jichaoyang@huawei.com> # message auto-generated for no-merge-commit merge: !2 merge master into master add ops-tilelang init framework Created-by: ChaoyangJi Commit-by: ChaoyangJi Merged-by: cann-robot Description: <!-- 感谢您的合入申请! --> ### 当前PR是否有AI参与: 无。 ### PR功能描述 / 为什么需要这个合入**: <!-- 本 PR 做了什么,为什么需要 / What does this PR do and why --> 本 PR 初始化 ops-tilelang 高性能算子仓,为 CANN 社区提供基于 TileLang DSL 的 Ascend 算子开发、测试、性能回归和 wheel 发布基础框架。 项目采用运行时接口与硬件 kernel 分层设计,当前仅支持 Ascend A5 及以上架构。 主要改动 1. 基础工程框架 - 建立 src/ layout Python 工程。 - 发行包名为 ops-tilelang,Python 命名空间为 cann_ops_tilelang。 - 使用 setuptools、setuptools-scm 构建 wheel。 - 支持 Python 3.10、3.11 和 3.12。 - 运行时依赖 torch>=2.1 和 tilelang>=0.1.9。 - torch_npu、CANN Toolkit 和驱动由目标环境提供。 - License 使用 MIT。 - 包导入阶段不初始化 NPU、不编译 kernel、不申请设备内存。 2. Engram 算子 首批提供四个公开 API: - engram_fused_weight - BF16 权重逐元素融合,使用 FP32 输出。 - engram_gate_fwd - 支持动态 token 数。 - 输出 gate 结果及反向所需统计量。 - engram_gate_bwd - 融合计算输入梯度、KV 梯度和分片权重梯度。 - engram_grad_w_reduce - 归并分片权重梯度并原地累加两份 FP32 权重梯度。 算子采用: - PyTorch wrapper 与 *_kernel.py 分层。 - tilelang.jit kernel generator。 - GM/UB 分块、SimdVF、Persistent 调度及多级流水。 - 明确的 device、dtype、shape、stride、contiguous 和输出缓冲区校验。 - 无 CPU/CUDA 静默回退。 3. 测试与性能回归框架 参考 TileKernels-Nightly 迁移并适配完整 pytest 能力: - OPS_TILELANG_TEST_LEVEL=0|1|2 分级参数生成。 - 基于 --seed 和 node ID 的确定性随机数。 - PyTorch reference 数值验证。 - NPU 峰值显存画像和 large_npu_mem 自动分类。 - A5 benchmark JSONL 基线及回归检测。 - -K 参数表达式筛选。 - benchmark improvement 几何平均汇总。 - xdist 多卡绑定及全局 fail-fast。 - TileLang 编译告警汇总。 - 普通显存与大/未知显存测试分阶段执行。 4. 构建、CI 与脚本 新增统一执行入口: - scripts/check.sh - scripts/build_wheel.sh - scripts/run_test.sh - scripts/run_tests.sh - scripts/run_ci.sh scripts/run_ci.sh 支持: - lint - cpu - wheel - npu - benchmark - all 新增 GitHub workflows: - Ruff lint 检查。 - Wheel 构建、安装及元数据验证。 5. 文档 新增并完善: - README:安装、使用、测试、CI 和硬件支持范围。 - 架构说明。 - 完整的新增算子开发流程。 - pytest 插件、benchmark 和显存画像维护指南。 - CONTRIBUTING 贡献规范。 - MIT License。 ### 该PR关联的issue 无。 ### 希望检视人员了解: ops-tilelang的框架代码是否合理和完善。 ## 改动类型 / Change Type - [ ] Bug 修复 / Bug Fix - [ √] 新功能 / New Feature - [ ] 性能优化 / Performance - [ ] 代码重构 / Refactoring - [ ] 文档更新 / Documentation - [ ] 测试相关 / Test - [ ] 其它 / Other ## 测试信息 / Testing <!-- 简要测试说明或关键结果 / Brief test description or key results --> - [√] 单元测试通过 / UT passed - [ √] 集成测试通过 / ST passed - [ √] 人工验证通过 / Manual verified ## 检查清单 / Checklist - [√] 代码符合规范 / Code follows style guide - [ √] 测试添加并通过 / Tests added and passed - [ √] 文档已更新 / Docs updated if needed - [√ ] 无硬编码敏感信息 / No secrets hardcoded - [ √] 提交信息符合规范 / Commit message follows convention See merge request: cann/ops-tilelang!2 | 1 month ago | |
本 PR 初始化 ops-tilelang 高性能算子仓,为 CANN 社区提供基于 TileLang DSL 的 Ascend 算子开发、测试、性能回归和 wheel 发布基础框架。 项目采用运行时接口与硬件 kernel 分层设计,当前仅支持 Ascend A5 及以上架构。 Co-authored-by: ChaoyangJi<jichaoyang@huawei.com> # message auto-generated for no-merge-commit merge: !2 merge master into master add ops-tilelang init framework Created-by: ChaoyangJi Commit-by: ChaoyangJi Merged-by: cann-robot Description: <!-- 感谢您的合入申请! --> ### 当前PR是否有AI参与: 无。 ### PR功能描述 / 为什么需要这个合入**: <!-- 本 PR 做了什么,为什么需要 / What does this PR do and why --> 本 PR 初始化 ops-tilelang 高性能算子仓,为 CANN 社区提供基于 TileLang DSL 的 Ascend 算子开发、测试、性能回归和 wheel 发布基础框架。 项目采用运行时接口与硬件 kernel 分层设计,当前仅支持 Ascend A5 及以上架构。 主要改动 1. 基础工程框架 - 建立 src/ layout Python 工程。 - 发行包名为 ops-tilelang,Python 命名空间为 cann_ops_tilelang。 - 使用 setuptools、setuptools-scm 构建 wheel。 - 支持 Python 3.10、3.11 和 3.12。 - 运行时依赖 torch>=2.1 和 tilelang>=0.1.9。 - torch_npu、CANN Toolkit 和驱动由目标环境提供。 - License 使用 MIT。 - 包导入阶段不初始化 NPU、不编译 kernel、不申请设备内存。 2. Engram 算子 首批提供四个公开 API: - engram_fused_weight - BF16 权重逐元素融合,使用 FP32 输出。 - engram_gate_fwd - 支持动态 token 数。 - 输出 gate 结果及反向所需统计量。 - engram_gate_bwd - 融合计算输入梯度、KV 梯度和分片权重梯度。 - engram_grad_w_reduce - 归并分片权重梯度并原地累加两份 FP32 权重梯度。 算子采用: - PyTorch wrapper 与 *_kernel.py 分层。 - tilelang.jit kernel generator。 - GM/UB 分块、SimdVF、Persistent 调度及多级流水。 - 明确的 device、dtype、shape、stride、contiguous 和输出缓冲区校验。 - 无 CPU/CUDA 静默回退。 3. 测试与性能回归框架 参考 TileKernels-Nightly 迁移并适配完整 pytest 能力: - OPS_TILELANG_TEST_LEVEL=0|1|2 分级参数生成。 - 基于 --seed 和 node ID 的确定性随机数。 - PyTorch reference 数值验证。 - NPU 峰值显存画像和 large_npu_mem 自动分类。 - A5 benchmark JSONL 基线及回归检测。 - -K 参数表达式筛选。 - benchmark improvement 几何平均汇总。 - xdist 多卡绑定及全局 fail-fast。 - TileLang 编译告警汇总。 - 普通显存与大/未知显存测试分阶段执行。 4. 构建、CI 与脚本 新增统一执行入口: - scripts/check.sh - scripts/build_wheel.sh - scripts/run_test.sh - scripts/run_tests.sh - scripts/run_ci.sh scripts/run_ci.sh 支持: - lint - cpu - wheel - npu - benchmark - all 新增 GitHub workflows: - Ruff lint 检查。 - Wheel 构建、安装及元数据验证。 5. 文档 新增并完善: - README:安装、使用、测试、CI 和硬件支持范围。 - 架构说明。 - 完整的新增算子开发流程。 - pytest 插件、benchmark 和显存画像维护指南。 - CONTRIBUTING 贡献规范。 - MIT License。 ### 该PR关联的issue 无。 ### 希望检视人员了解: ops-tilelang的框架代码是否合理和完善。 ## 改动类型 / Change Type - [ ] Bug 修复 / Bug Fix - [ √] 新功能 / New Feature - [ ] 性能优化 / Performance - [ ] 代码重构 / Refactoring - [ ] 文档更新 / Documentation - [ ] 测试相关 / Test - [ ] 其它 / Other ## 测试信息 / Testing <!-- 简要测试说明或关键结果 / Brief test description or key results --> - [√] 单元测试通过 / UT passed - [ √] 集成测试通过 / ST passed - [ √] 人工验证通过 / Manual verified ## 检查清单 / Checklist - [√] 代码符合规范 / Code follows style guide - [ √] 测试添加并通过 / Tests added and passed - [ √] 文档已更新 / Docs updated if needed - [√ ] 无硬编码敏感信息 / No secrets hardcoded - [ √] 提交信息符合规范 / Commit message follows convention See merge request: cann/ops-tilelang!2 | 1 month ago | |
fix: 合法上界种子与用例哈希组合后溢出导致 CPU 测试失败 Co-authored-by: Ethan_Zou<tianjun.zou@foxmail.com> # message auto-generated for no-merge-commit merge: !39 merge fix/tilelang-node-seed-overflow-20261014 into master fix: 合法上界种子与用例哈希组合后溢出导致 CPU 测试失败 Created-by: Ethan_Zou Commit-by: Ethan_Zou Merged-by: cann-robot Description: ## 描述 --seed 接受的合法 64 位基准 seed 与 pytest node hash 相加可能超过 torch.manual_seed 上界。仅对合法基准发生的上溢按2**64归一化;普通 seed 序列和无效基准的既有错误行为保持。 按原审核意见删除新增 tests/host/test_node_seed_overflow.py,当前 PR 只有 tests/pytest_random_plugin.py 两行既定修复;生产文件与此前公开源字节一致。13项验证保留在PR外,不改仓内功能/性能用例目录。 当前单提交 08be1f61cbed45198401b20b14b0e913adb7bbc7,目标 ce4c30075c5836ee71d18bb8eb7663987ce580da。版权、许可证、OAT配置和CI规则均未修改。 ## 关联的Issue Fixes #11 ## 实际测试 同一完整真实插件、pytest子进程和CPU Torch/Python RNG比较:先前干净基线3 failed/10 passed;删除测试后的最终源码13 passed,无skip。覆盖正常值、合法边界上溢、无效上/下界和随机序列保持。完整模块由实际pytest加载,不使用AST抽取或fake模块。 Python3.11.6 / PyTorch2.7.1+cpu / pytest9.1.1,仅Host;没有运行TileLang算子或NPU。外部验证脚本 evidence/expanded-repairs-20261019/ops-tilelang-39/test_node_seed_overflow_external.py,以CANN_TILELANG_SOURCE_ROOT指定源码。 正常提交及新HEAD原生pre-commit均通过(本仓配置只有Ruff check/format),全仓Ruff check和format --check通过;diff --check、LFS fetch/fsck、最新目标merge-tree通过。 此前额外完整OAT的plugin缺少许可头/版权头各1条,与未改目标一致。本次核对确认仓内pre-commit未配置OAT;保留此诊断,不称完整OAT通过,也未以改头、豁免或关闭门禁处理。新SHA官方CI仍待实际源码绑定及结束结果。 ## AI 参与 OpenAI Codex(GPT-6系列)辅助落实现有审核意见、真实Host验证和发布核对;由Ethan_Zou提交,保持一个提交。 ## 当前源码 CI 核验 当前源 08be1f61cbed45198401b20b14b0e913adb7bbc7 的官方业务运行 9b3a2d5ab97140b4bc1c9b62507c4cc5 已于 2026-10-19T11:25:38+00:00 实际结束;源码绑定、SCA 明确通过及 Test_finish 的 pipeline_failed=false 均已核验。此结论仅覆盖该实际运行;其他工作流、跳过或禁用项仍未证明,不能算完整 CI 全通过。 See merge request: cann/ops-tilelang!39 | 21 days ago | |
本 PR 初始化 ops-tilelang 高性能算子仓,为 CANN 社区提供基于 TileLang DSL 的 Ascend 算子开发、测试、性能回归和 wheel 发布基础框架。 项目采用运行时接口与硬件 kernel 分层设计,当前仅支持 Ascend A5 及以上架构。 Co-authored-by: ChaoyangJi<jichaoyang@huawei.com> # message auto-generated for no-merge-commit merge: !2 merge master into master add ops-tilelang init framework Created-by: ChaoyangJi Commit-by: ChaoyangJi Merged-by: cann-robot Description: <!-- 感谢您的合入申请! --> ### 当前PR是否有AI参与: 无。 ### PR功能描述 / 为什么需要这个合入**: <!-- 本 PR 做了什么,为什么需要 / What does this PR do and why --> 本 PR 初始化 ops-tilelang 高性能算子仓,为 CANN 社区提供基于 TileLang DSL 的 Ascend 算子开发、测试、性能回归和 wheel 发布基础框架。 项目采用运行时接口与硬件 kernel 分层设计,当前仅支持 Ascend A5 及以上架构。 主要改动 1. 基础工程框架 - 建立 src/ layout Python 工程。 - 发行包名为 ops-tilelang,Python 命名空间为 cann_ops_tilelang。 - 使用 setuptools、setuptools-scm 构建 wheel。 - 支持 Python 3.10、3.11 和 3.12。 - 运行时依赖 torch>=2.1 和 tilelang>=0.1.9。 - torch_npu、CANN Toolkit 和驱动由目标环境提供。 - License 使用 MIT。 - 包导入阶段不初始化 NPU、不编译 kernel、不申请设备内存。 2. Engram 算子 首批提供四个公开 API: - engram_fused_weight - BF16 权重逐元素融合,使用 FP32 输出。 - engram_gate_fwd - 支持动态 token 数。 - 输出 gate 结果及反向所需统计量。 - engram_gate_bwd - 融合计算输入梯度、KV 梯度和分片权重梯度。 - engram_grad_w_reduce - 归并分片权重梯度并原地累加两份 FP32 权重梯度。 算子采用: - PyTorch wrapper 与 *_kernel.py 分层。 - tilelang.jit kernel generator。 - GM/UB 分块、SimdVF、Persistent 调度及多级流水。 - 明确的 device、dtype、shape、stride、contiguous 和输出缓冲区校验。 - 无 CPU/CUDA 静默回退。 3. 测试与性能回归框架 参考 TileKernels-Nightly 迁移并适配完整 pytest 能力: - OPS_TILELANG_TEST_LEVEL=0|1|2 分级参数生成。 - 基于 --seed 和 node ID 的确定性随机数。 - PyTorch reference 数值验证。 - NPU 峰值显存画像和 large_npu_mem 自动分类。 - A5 benchmark JSONL 基线及回归检测。 - -K 参数表达式筛选。 - benchmark improvement 几何平均汇总。 - xdist 多卡绑定及全局 fail-fast。 - TileLang 编译告警汇总。 - 普通显存与大/未知显存测试分阶段执行。 4. 构建、CI 与脚本 新增统一执行入口: - scripts/check.sh - scripts/build_wheel.sh - scripts/run_test.sh - scripts/run_tests.sh - scripts/run_ci.sh scripts/run_ci.sh 支持: - lint - cpu - wheel - npu - benchmark - all 新增 GitHub workflows: - Ruff lint 检查。 - Wheel 构建、安装及元数据验证。 5. 文档 新增并完善: - README:安装、使用、测试、CI 和硬件支持范围。 - 架构说明。 - 完整的新增算子开发流程。 - pytest 插件、benchmark 和显存画像维护指南。 - CONTRIBUTING 贡献规范。 - MIT License。 ### 该PR关联的issue 无。 ### 希望检视人员了解: ops-tilelang的框架代码是否合理和完善。 ## 改动类型 / Change Type - [ ] Bug 修复 / Bug Fix - [ √] 新功能 / New Feature - [ ] 性能优化 / Performance - [ ] 代码重构 / Refactoring - [ ] 文档更新 / Documentation - [ ] 测试相关 / Test - [ ] 其它 / Other ## 测试信息 / Testing <!-- 简要测试说明或关键结果 / Brief test description or key results --> - [√] 单元测试通过 / UT passed - [ √] 集成测试通过 / ST passed - [ √] 人工验证通过 / Manual verified ## 检查清单 / Checklist - [√] 代码符合规范 / Code follows style guide - [ √] 测试添加并通过 / Tests added and passed - [ √] 文档已更新 / Docs updated if needed - [√ ] 无硬编码敏感信息 / No secrets hardcoded - [ √] 提交信息符合规范 / Commit message follows convention See merge request: cann/ops-tilelang!2 | 1 month ago | |
本 PR 初始化 ops-tilelang 高性能算子仓,为 CANN 社区提供基于 TileLang DSL 的 Ascend 算子开发、测试、性能回归和 wheel 发布基础框架。 项目采用运行时接口与硬件 kernel 分层设计,当前仅支持 Ascend A5 及以上架构。 Co-authored-by: ChaoyangJi<jichaoyang@huawei.com> # message auto-generated for no-merge-commit merge: !2 merge master into master add ops-tilelang init framework Created-by: ChaoyangJi Commit-by: ChaoyangJi Merged-by: cann-robot Description: <!-- 感谢您的合入申请! --> ### 当前PR是否有AI参与: 无。 ### PR功能描述 / 为什么需要这个合入**: <!-- 本 PR 做了什么,为什么需要 / What does this PR do and why --> 本 PR 初始化 ops-tilelang 高性能算子仓,为 CANN 社区提供基于 TileLang DSL 的 Ascend 算子开发、测试、性能回归和 wheel 发布基础框架。 项目采用运行时接口与硬件 kernel 分层设计,当前仅支持 Ascend A5 及以上架构。 主要改动 1. 基础工程框架 - 建立 src/ layout Python 工程。 - 发行包名为 ops-tilelang,Python 命名空间为 cann_ops_tilelang。 - 使用 setuptools、setuptools-scm 构建 wheel。 - 支持 Python 3.10、3.11 和 3.12。 - 运行时依赖 torch>=2.1 和 tilelang>=0.1.9。 - torch_npu、CANN Toolkit 和驱动由目标环境提供。 - License 使用 MIT。 - 包导入阶段不初始化 NPU、不编译 kernel、不申请设备内存。 2. Engram 算子 首批提供四个公开 API: - engram_fused_weight - BF16 权重逐元素融合,使用 FP32 输出。 - engram_gate_fwd - 支持动态 token 数。 - 输出 gate 结果及反向所需统计量。 - engram_gate_bwd - 融合计算输入梯度、KV 梯度和分片权重梯度。 - engram_grad_w_reduce - 归并分片权重梯度并原地累加两份 FP32 权重梯度。 算子采用: - PyTorch wrapper 与 *_kernel.py 分层。 - tilelang.jit kernel generator。 - GM/UB 分块、SimdVF、Persistent 调度及多级流水。 - 明确的 device、dtype、shape、stride、contiguous 和输出缓冲区校验。 - 无 CPU/CUDA 静默回退。 3. 测试与性能回归框架 参考 TileKernels-Nightly 迁移并适配完整 pytest 能力: - OPS_TILELANG_TEST_LEVEL=0|1|2 分级参数生成。 - 基于 --seed 和 node ID 的确定性随机数。 - PyTorch reference 数值验证。 - NPU 峰值显存画像和 large_npu_mem 自动分类。 - A5 benchmark JSONL 基线及回归检测。 - -K 参数表达式筛选。 - benchmark improvement 几何平均汇总。 - xdist 多卡绑定及全局 fail-fast。 - TileLang 编译告警汇总。 - 普通显存与大/未知显存测试分阶段执行。 4. 构建、CI 与脚本 新增统一执行入口: - scripts/check.sh - scripts/build_wheel.sh - scripts/run_test.sh - scripts/run_tests.sh - scripts/run_ci.sh scripts/run_ci.sh 支持: - lint - cpu - wheel - npu - benchmark - all 新增 GitHub workflows: - Ruff lint 检查。 - Wheel 构建、安装及元数据验证。 5. 文档 新增并完善: - README:安装、使用、测试、CI 和硬件支持范围。 - 架构说明。 - 完整的新增算子开发流程。 - pytest 插件、benchmark 和显存画像维护指南。 - CONTRIBUTING 贡献规范。 - MIT License。 ### 该PR关联的issue 无。 ### 希望检视人员了解: ops-tilelang的框架代码是否合理和完善。 ## 改动类型 / Change Type - [ ] Bug 修复 / Bug Fix - [ √] 新功能 / New Feature - [ ] 性能优化 / Performance - [ ] 代码重构 / Refactoring - [ ] 文档更新 / Documentation - [ ] 测试相关 / Test - [ ] 其它 / Other ## 测试信息 / Testing <!-- 简要测试说明或关键结果 / Brief test description or key results --> - [√] 单元测试通过 / UT passed - [ √] 集成测试通过 / ST passed - [ √] 人工验证通过 / Manual verified ## 检查清单 / Checklist - [√] 代码符合规范 / Code follows style guide - [ √] 测试添加并通过 / Tests added and passed - [ √] 文档已更新 / Docs updated if needed - [√ ] 无硬编码敏感信息 / No secrets hardcoded - [ √] 提交信息符合规范 / Commit message follows convention See merge request: cann/ops-tilelang!2 | 1 month ago | |
fix: reject non-finite quantization clamp values Co-authored-by: Ethan_Zou<tianjun.zou@foxmail.com> # message auto-generated for no-merge-commit merge: !28 merge docs/tilelang-finite-quant-clamp-20261002 into master fix: reject non-finite quantization clamp values Created-by: Ethan_Zou Commit-by: Ethan_Zou Merged-by: cann-robot Description: ## 版权维护与当前验证 个人版权已按仓库规定的组织署名修正,参考 LICENSE;保留原年份和许可证正文。代码AST与版权行以外的字节不变。当前源提交 c46874ad413484e3b03556ee13bee21a4c7e9070 保留单提交,本地原有 17 个 Host 回归用例通过;当前源 SHA c46874ad413484e3b03556ee13bee21a4c7e9070 的[流水线](https://gitcode.com/cann/ops-tilelang/actions/runs/93de23f225b14ca4b5be2b1111983956)已完成,15个任务 COMPLETED。未执行NPU测试。以下保留原功能问题及复现说明。 ### AI 参与 - [x] 是:OpenAI Codex(GPT-6 系列) Prompt上下文:在无NPU环境审查CANN仓库中的可复现问题,排除已有报告,提交独立修复。 ### 问题与修复 set_amax_clamp_for_quant只检查value <= 0。NaN与正无穷大均通过此判断,并被get_amax_clamp_for_quant原样返回。该值用于量化参考计算的torch.clamp下界以及quant/common.py的调度配置;非法值在Host配置入口处没有被拒绝,污染后续计算的全局最小幅值配置。 同时要求数值有限且为正;拒绝非法输入时保留已有配置。添加两个量化格式的非法值、合法正数和不支持格式的Host参数化回归。 ### 关联 Issue Fixes #3 ### 改动类型 - [x] Bug 修复 - [x] 测试相关 ### 验证 使用实际config.py和新增17项pytest,隔离包初始化,torch仅提供不调用设备的空导入stub;无NPU操作。原版的NaN与+Inf四项会因未抛异常失败,修复后17项通过;负无穷、零、负数仍拒绝,合法正数与格式校验正常。此验证不代表量化设备精度或完整算子套件通过。 仓库适用 pre-commit hooks 与 diff --check 通过。 仓库适用 pre-commit hooks 与 diff --check 通过。 没有NPU、msprof或TileLang运行环境,未声明完整算子测试或硬件精度验证通过。 基于 874d6410590472febdfe1b4ca262cb17a74974d5,PR仅一个提交;提交后评论 /compile。 See merge request: cann/ops-tilelang!28 | 1 month ago | |
本 PR 初始化 ops-tilelang 高性能算子仓,为 CANN 社区提供基于 TileLang DSL 的 Ascend 算子开发、测试、性能回归和 wheel 发布基础框架。 项目采用运行时接口与硬件 kernel 分层设计,当前仅支持 Ascend A5 及以上架构。 Co-authored-by: ChaoyangJi<jichaoyang@huawei.com> # message auto-generated for no-merge-commit merge: !2 merge master into master add ops-tilelang init framework Created-by: ChaoyangJi Commit-by: ChaoyangJi Merged-by: cann-robot Description: <!-- 感谢您的合入申请! --> ### 当前PR是否有AI参与: 无。 ### PR功能描述 / 为什么需要这个合入**: <!-- 本 PR 做了什么,为什么需要 / What does this PR do and why --> 本 PR 初始化 ops-tilelang 高性能算子仓,为 CANN 社区提供基于 TileLang DSL 的 Ascend 算子开发、测试、性能回归和 wheel 发布基础框架。 项目采用运行时接口与硬件 kernel 分层设计,当前仅支持 Ascend A5 及以上架构。 主要改动 1. 基础工程框架 - 建立 src/ layout Python 工程。 - 发行包名为 ops-tilelang,Python 命名空间为 cann_ops_tilelang。 - 使用 setuptools、setuptools-scm 构建 wheel。 - 支持 Python 3.10、3.11 和 3.12。 - 运行时依赖 torch>=2.1 和 tilelang>=0.1.9。 - torch_npu、CANN Toolkit 和驱动由目标环境提供。 - License 使用 MIT。 - 包导入阶段不初始化 NPU、不编译 kernel、不申请设备内存。 2. Engram 算子 首批提供四个公开 API: - engram_fused_weight - BF16 权重逐元素融合,使用 FP32 输出。 - engram_gate_fwd - 支持动态 token 数。 - 输出 gate 结果及反向所需统计量。 - engram_gate_bwd - 融合计算输入梯度、KV 梯度和分片权重梯度。 - engram_grad_w_reduce - 归并分片权重梯度并原地累加两份 FP32 权重梯度。 算子采用: - PyTorch wrapper 与 *_kernel.py 分层。 - tilelang.jit kernel generator。 - GM/UB 分块、SimdVF、Persistent 调度及多级流水。 - 明确的 device、dtype、shape、stride、contiguous 和输出缓冲区校验。 - 无 CPU/CUDA 静默回退。 3. 测试与性能回归框架 参考 TileKernels-Nightly 迁移并适配完整 pytest 能力: - OPS_TILELANG_TEST_LEVEL=0|1|2 分级参数生成。 - 基于 --seed 和 node ID 的确定性随机数。 - PyTorch reference 数值验证。 - NPU 峰值显存画像和 large_npu_mem 自动分类。 - A5 benchmark JSONL 基线及回归检测。 - -K 参数表达式筛选。 - benchmark improvement 几何平均汇总。 - xdist 多卡绑定及全局 fail-fast。 - TileLang 编译告警汇总。 - 普通显存与大/未知显存测试分阶段执行。 4. 构建、CI 与脚本 新增统一执行入口: - scripts/check.sh - scripts/build_wheel.sh - scripts/run_test.sh - scripts/run_tests.sh - scripts/run_ci.sh scripts/run_ci.sh 支持: - lint - cpu - wheel - npu - benchmark - all 新增 GitHub workflows: - Ruff lint 检查。 - Wheel 构建、安装及元数据验证。 5. 文档 新增并完善: - README:安装、使用、测试、CI 和硬件支持范围。 - 架构说明。 - 完整的新增算子开发流程。 - pytest 插件、benchmark 和显存画像维护指南。 - CONTRIBUTING 贡献规范。 - MIT License。 ### 该PR关联的issue 无。 ### 希望检视人员了解: ops-tilelang的框架代码是否合理和完善。 ## 改动类型 / Change Type - [ ] Bug 修复 / Bug Fix - [ √] 新功能 / New Feature - [ ] 性能优化 / Performance - [ ] 代码重构 / Refactoring - [ ] 文档更新 / Documentation - [ ] 测试相关 / Test - [ ] 其它 / Other ## 测试信息 / Testing <!-- 简要测试说明或关键结果 / Brief test description or key results --> - [√] 单元测试通过 / UT passed - [ √] 集成测试通过 / ST passed - [ √] 人工验证通过 / Manual verified ## 检查清单 / Checklist - [√] 代码符合规范 / Code follows style guide - [ √] 测试添加并通过 / Tests added and passed - [ √] 文档已更新 / Docs updated if needed - [√ ] 无硬编码敏感信息 / No secrets hardcoded - [ √] 提交信息符合规范 / Commit message follows convention See merge request: cann/ops-tilelang!2 | 1 month ago |
测试与性能回归框架
测试框架沿用 TileKernels-Nightly 的完整链路,并收敛为 Ascend/NPU 语义。测试目录根部的 tests/conftest.py 负责在 TileLang 导入前绑定 xdist worker,随后装载以下独立 pytest 插件:
pytest_random_plugin.py:--seed与 node ID 组合,得到逐用例稳定随机种子;pytest_npu_mem_plugin.py:记录峰值 NPU 显存,自动标记large_npu_mem;pytest_benchmark_plugin.py:计时、JSONL 输出、性能基线比较与刷新;pytest_improvement_plugin.py:通过--legacy-table汇总几何平均性能;pytest_xdist_failfast_plugin.py:并行-x时停止全部 worker;pytest_tilelang_warning_plugin.py:在终端汇总 TileLang 编译警告。
CI 增量测试
CI 根据 pr_filelist.txt 中的变更,通过 .gitcode/scripts/select_tests.py 静态分析源码和测试的导入关系,自动选择直接或间接受影响的测试。新增算子及测试无需登记映射;测试应通过常规 Python 导入引用被测实现和参考实现。
共享配置、包入口、删除文件或无法确定依赖关系的代码变更保守执行全仓 L0。全局触发策略保留在 .gitcode/config/full_test_triggers.conf,其中不维护算子与测试的对应关系。选择器的 CPU 回归测试可用 python -m unittest discover -s .gitcode/scripts/tests 运行。
.gitcode/ 下的 CI 文件默认不触发算子测试,便于管理员在机器故障等场景调整执行逻辑。仅修改 CI 文件或文档时跳过算子用例;同时修改源码或测试时,仍正常选择受影响的测试。显式配置的全量触发规则优先,管理员可按需添加 CI 路径规则;选择器的 CPU 回归测试仍会执行。
测试分级
OPS_TILELANG_TEST_LEVEL 取值为 0、1、2,分别表示核心、默认和完整参数集合。参数生成器位于 tests/testing/generator.py。
OPS_TILELANG_TEST_LEVEL=0 pytest tests/engram -m 'not benchmark'
OPS_TILELANG_TEST_LEVEL=2 pytest tests/engram -m 'not benchmark'
性能基线
benchmark 默认跳过,必须显式传入 --run-benchmark。Ascend 使用 msprof 后端,并将结果归入 a5 集群。基线存储在 tests/<operator>/benchmark/<测试文件>.jsonl。
# 全部 benchmark
pytest -m benchmark --run-benchmark
# 用参数表达式筛选;支持 and/or/not 和正则
pytest -m benchmark --run-benchmark -K 'hidden=4096 or hidden=7168'
# 指定输出、阈值和最小绝对变化
pytest -m benchmark --run-benchmark \
--benchmark-output=/tmp/ops-tilelang.jsonl \
--benchmark-regression-threshold=0.10 \
--benchmark-min-delta-us=1.0
# 确认性能变化是预期后刷新分片基线
pytest tests/engram/test_engram_fused_weight.py \
-m benchmark --run-benchmark --refresh-baseline
# Engram gate 前向/反向及梯度归并
pytest tests/engram -m benchmark --run-benchmark --refresh-baseline
--refresh-baseline 在 pytest_sessionfinish 收到 exitstatus=0 时更新基线;收到测试失败、fixture 错误或中断状态时不更新,原始结果输出仍可保留。后续终端报告或其他插件的错误不属于此守卫范围。
-K 的布尔运算优先级为 not、and、or(相邻正则也按 and 处理)。
解析器不支持用括号对布尔表达式分组;括号属于单个正则模式的语法。
例如,用 -K 'hidden=(4096|7168) and not round=True' 筛选两种 hidden 值并排除 round=True,
不要写成 -K '(hidden=4096 or hidden=7168) and not round=True'。
--benchmark-verbose 展示扩展列,--legacy-table 展示算子级几何平均汇总。
NPU 显存画像
显存画像存储在 tests/<operator>/npu_mem_profiles/<测试文件>.json。无画像的测试会先归入 large_npu_mem,避免以高并发运行未知用例。
# 运行并刷新选中用例的显存画像
pytest tests/engram/test_engram_fused_weight.py -m 'not benchmark' --refresh-mem
# 刷新整个 Engram 算子族的画像
pytest tests/engram -m 'not benchmark' --refresh-mem
# 只运行尚无画像的测试
pytest --refresh-mem --refresh-mem-missing-only
# 调整大显存阈值或保留原始 JSONL
pytest --refresh-mem --npu-mem-threshold=15000 \
--npu-mem-output=/tmp/ops-tilelang-mem.jsonl
--npu-mem-flush-interval 控制长任务周期写回间隔。
标准执行入口
scripts/run_test.sh 默认分两阶段运行:普通测试每张 NPU 两个 worker,大/未知显存测试每张 NPU 一个 worker。可通过 TILELANG_TEST_WORKERS_PER_NPU 和 TILELANG_TEST_WORKERS_PER_NPU_LARGE 调整。
./scripts/run_test.sh
OPS_TILELANG_RUN_BENCHMARK=1 ./scripts/run_test.sh tests/engram