不用写Tiling也能快速上线NPU算子ops-transformer简易算子调用实战指南【免费下载链接】ops-transformer本项目是CANN提供的transformer类大模型算子库实现网络在NPU上加速计算。项目地址: https://gitcode.com/cann/ops-transformerops-transformer 是 CANN 提供的 transformer 类大模型 NPU 算子库而它内置的 examples/fast_kernel_launch_example 示例展示了如何用语法一行代码启动核函数、无需单独编写 Tiling 工程就能把自定义 NPU 算子打包成 wheel 并在 PyTorch 中直接调用。本文以 add 简易算子为例带你走完构建 → 调用 → 新增算子 → 测试验证的完整实战流程。为什么调用适合算子快速上线 传统 CANN 算子开发通常要维护 op_hostTiling 生成与检查、op_kernel、schema 注册等一整套多文件工程。而 ops-transformer 提供的Fast Kernel Launch快速核函数启动方式把这一切压缩到了极致单交付件一个.cpp文件同时完成算子开发 PyTorch 框架适配高效调用使用类 CUDA 的语法直接启动 Ascend C 核函数流程简单高效无缝集成 PyTorch算子通过torch.ops分发机制注册用起来和原生 PyTorch 算子一模一样后续可支持图加速等能力。官方示例文档见examples/fast_kernel_launch_example/README.md下面开始实战。先看目录结构心里有数 整个示例工程组织得非常清晰核心文件如下文件路径作用csrc/add/ascend910b/add.cppadd 算子的单文件完整实现Schema 注册、Meta 函数、Kernel、调用接口全在一个文件里csrc/grouped_matmul/ascend910b/grouped_matmul.cpp分组矩阵乘算子示例方便用户自定义使用csrc/incre_flash_attention/、csrc/moe_distribute_dispatch_v2/等更复杂的真实算子接入示例tests/add/test_add.py基于 pytest 的算子功能测试setup.py、requirements.txt、build_and_test.shwheel 构建与一键脚本一键构建从源码到 wheel 安装包 前提条件环境部署可参考docs/QUICKSTART.md已安装 NPU 驱动与 CANN 包gcc 9.4.0、Python 3.8、torch ≥ 2.6.0、对应版本的 TorchNPU。进入示例目录后三步完成构建安装cd examples/fast_kernel_launch_example python3 -m pip install -r requirements.txt # 设置编译款型A2 系列默认 ascend910bA3 系列用 ascend910_93950 系列用 ascend950 export NPU_SOC_VERSIONascend910b python3 -m build --wheel -n # -n 表示复用当前环境构建 python3 -m pip install dist/*.whl --force-reinstall --no-deps 再次构建前建议先执行python setup.py clean清理编译缓存避免脏产物干扰。Python 调用像用 PyTorch 算子一样用 NPU 算子 安装完成后算子就注册进了 PyTorch 的torch.ops命名空间调用格式为torch.ops.库名.算子名。以 add 算子为例import torch import torch_npu import ascend_ops # 上面构建出的 python 包 x torch.randn(10, 32, dtypetorch.float32).npu() y torch.randn(10, 32, dtypetorch.float32).npu() # 一行调用自定义 NPU 算子 npu_result torch.ops.ascend_ops.add(x, y) # 与 CPU 结果对比验证 assert torch.allclose(x.cpu() y.cpu(), npu_result.cpu(), rtol1e-6) print(Verification successful!)一个文件里到底发生了什么打开csrc/add/ascend910b/add.cpp会发现开发一个 AI Core 算子所需的全部模块都在这一张文件里Schema 注册TORCH_LIBRARY_FRAGMENT告诉框架有这么一个算子及其签名Meta 函数实现 InferShape InferDtype先算出输出的形状和类型不做实际计算——这也是图模式torch.compile/AclGraph 等可用的关键Ascend C 核函数__global__ __aicore__标记的add_kernel内部用 TPipe/TQue 完成拷贝进 UB → 向量计算 → 拷贝回 GM的完整流水NPU 调用接口与注册通过TORCH_LIBRARY_IMPL(..., PrivateUse1, ...)注册到 NPU 设备。其中最关键的启动代码就是形如这样的一行add_kernelscalar_tnumBlocks, nullptr, stream(x_ptr, y_ptr, z_ptr, totalLength, blockLength, tileSize);语法与 CUDA kernel launch 类似第一个参数是启动的核数Block 数最后一个参数是 NPU 流。而大家最头疼的Tiling 参数numBlocks / blockLength / tileSize不用单独建 Tiling 工程直接在 Host 侧用一个calc_tiling_params()函数根据张量长度和芯片核数、UB 大小内联算出来这就是不用写 Tiling 也能快速上线的核心所在。调用还通过OpCommand::RunOpApi(Add, acl_call)包装保证时序与 TorchNPU 调用 aclnn 接口一致。动手实战新增一个属于你自己的算子 ✍️想把自己的算子接进来流程只有五步建目录在csrc下用算子名建文件夹并在其中建当前 SoC 名子目录例如csrc/my_op/ascend910b/建 CMakeLists.txt内容只需一行add_sources(--npu-archdav-2201)dav-2201是 ascend910b 对应的 NpuArch 编译参数写算子实现新建my_op.cpp建议与算子同名参考add.cpp补齐四件套——Schema 注册、Meta 函数、__aicore__Kernel、调用接口与 PrivateUse1 注册重建安装回到示例目录重复python3 -m build --wheel -n与 pip 安装测试仿照tests/add/test_add.py编写 pytest跑通即上线。测试验证一次覆盖 54 组用例 ✅官方测试文件tests/add/test_add.py的做法很值得借鉴接口存在性测试断言torch.ops.ascend_ops.add可被发现——这能拦截schema 签名与 C 注册不一致导致算子被框架隐藏这类隐蔽问题功能参数化测试18 种 shape × 3 种数据类型float32 / float16 / int32的笛卡尔组合NPU 结果与 CPU 基准逐组比对浮点用torch.allclose校验。直接执行pytest tests/add/test_add.py即可回归你的算子。常见问题速查 ️为什么torch.ops里找不到我的算子九成是 schema 字符串参数名/类型/重载与 C 注册签名对不上对照add.cpp的m.def(...)与m.impl(...)逐项核对。编译报错找不到 CANN/Torch 路径确认已sourceCANN 环境变量set_env.sh并检查 torch、torch_npu 版本是否满足要求。改了代码没生效先python setup.py clean再重建 wheel最后加--force-reinstall安装。延伸学习 示例完整文档examples/fast_kernel_launch_example/README.md复杂算子接入参考csrc/grouped_matmul/分组矩阵乘、csrc/moe_distribute_dispatch_v2/MoE 分发算子零基础算子开发入门标准 op_host/op_kernel 流程docs/QUICKSTART.md算子贡献流程项目根目录CONTRIBUTING.md掌握快速核函数启动模式后一个简易 NPU 算子从想法到上线只需一个文件、两条命令——这正是 ops-transformer 为算子快速迭代提供的最高效路径。【免费下载链接】ops-transformer本项目是CANN提供的transformer类大模型算子库实现网络在NPU上加速计算。项目地址: https://gitcode.com/cann/ops-transformer创作声明：本文部分内容由AI辅助生成（AIGC），仅供参考
不用写Tiling也能快速上线NPU算子：ops-transformer简易算子＜＜＜＞＞＞调用实战指南
不用写Tiling也能快速上线NPU算子ops-transformer简易算子调用实战指南【免费下载链接】ops-transformer本项目是CANN提供的transformer类大模型算子库实现网络在NPU上加速计算。项目地址: https://gitcode.com/cann/ops-transformerops-transformer 是 CANN 提供的 transformer 类大模型 NPU 算子库而它内置的 examples/fast_kernel_launch_example 示例展示了如何用语法一行代码启动核函数、无需单独编写 Tiling 工程就能把自定义 NPU 算子打包成 wheel 并在 PyTorch 中直接调用。本文以 add 简易算子为例带你走完构建 → 调用 → 新增算子 → 测试验证的完整实战流程。为什么调用适合算子快速上线 传统 CANN 算子开发通常要维护 op_hostTiling 生成与检查、op_kernel、schema 注册等一整套多文件工程。而 ops-transformer 提供的Fast Kernel Launch快速核函数启动方式把这一切压缩到了极致单交付件一个.cpp文件同时完成算子开发 PyTorch 框架适配高效调用使用类 CUDA 的语法直接启动 Ascend C 核函数流程简单高效无缝集成 PyTorch算子通过torch.ops分发机制注册用起来和原生 PyTorch 算子一模一样后续可支持图加速等能力。官方示例文档见examples/fast_kernel_launch_example/README.md下面开始实战。先看目录结构心里有数 整个示例工程组织得非常清晰核心文件如下文件路径作用csrc/add/ascend910b/add.cppadd 算子的单文件完整实现Schema 注册、Meta 函数、Kernel、调用接口全在一个文件里csrc/grouped_matmul/ascend910b/grouped_matmul.cpp分组矩阵乘算子示例方便用户自定义使用csrc/incre_flash_attention/、csrc/moe_distribute_dispatch_v2/等更复杂的真实算子接入示例tests/add/test_add.py基于 pytest 的算子功能测试setup.py、requirements.txt、build_and_test.shwheel 构建与一键脚本一键构建从源码到 wheel 安装包 前提条件环境部署可参考docs/QUICKSTART.md已安装 NPU 驱动与 CANN 包gcc 9.4.0、Python 3.8、torch ≥ 2.6.0、对应版本的 TorchNPU。进入示例目录后三步完成构建安装cd examples/fast_kernel_launch_example python3 -m pip install -r requirements.txt # 设置编译款型A2 系列默认 ascend910bA3 系列用 ascend910_93950 系列用 ascend950 export NPU_SOC_VERSIONascend910b python3 -m build --wheel -n # -n 表示复用当前环境构建 python3 -m pip install dist/*.whl --force-reinstall --no-deps 再次构建前建议先执行python setup.py clean清理编译缓存避免脏产物干扰。Python 调用像用 PyTorch 算子一样用 NPU 算子 安装完成后算子就注册进了 PyTorch 的torch.ops命名空间调用格式为torch.ops.库名.算子名。以 add 算子为例import torch import torch_npu import ascend_ops # 上面构建出的 python 包 x torch.randn(10, 32, dtypetorch.float32).npu() y torch.randn(10, 32, dtypetorch.float32).npu() # 一行调用自定义 NPU 算子 npu_result torch.ops.ascend_ops.add(x, y) # 与 CPU 结果对比验证 assert torch.allclose(x.cpu() y.cpu(), npu_result.cpu(), rtol1e-6) print(Verification successful!)一个文件里到底发生了什么打开csrc/add/ascend910b/add.cpp会发现开发一个 AI Core 算子所需的全部模块都在这一张文件里Schema 注册TORCH_LIBRARY_FRAGMENT告诉框架有这么一个算子及其签名Meta 函数实现 InferShape InferDtype先算出输出的形状和类型不做实际计算——这也是图模式torch.compile/AclGraph 等可用的关键Ascend C 核函数__global__ __aicore__标记的add_kernel内部用 TPipe/TQue 完成拷贝进 UB → 向量计算 → 拷贝回 GM的完整流水NPU 调用接口与注册通过TORCH_LIBRARY_IMPL(..., PrivateUse1, ...)注册到 NPU 设备。其中最关键的启动代码就是形如这样的一行add_kernelscalar_tnumBlocks, nullptr, stream(x_ptr, y_ptr, z_ptr, totalLength, blockLength, tileSize);语法与 CUDA kernel launch 类似第一个参数是启动的核数Block 数最后一个参数是 NPU 流。而大家最头疼的Tiling 参数numBlocks / blockLength / tileSize不用单独建 Tiling 工程直接在 Host 侧用一个calc_tiling_params()函数根据张量长度和芯片核数、UB 大小内联算出来这就是不用写 Tiling 也能快速上线的核心所在。调用还通过OpCommand::RunOpApi(Add, acl_call)包装保证时序与 TorchNPU 调用 aclnn 接口一致。动手实战新增一个属于你自己的算子 ✍️想把自己的算子接进来流程只有五步建目录在csrc下用算子名建文件夹并在其中建当前 SoC 名子目录例如csrc/my_op/ascend910b/建 CMakeLists.txt内容只需一行add_sources(--npu-archdav-2201)dav-2201是 ascend910b 对应的 NpuArch 编译参数写算子实现新建my_op.cpp建议与算子同名参考add.cpp补齐四件套——Schema 注册、Meta 函数、__aicore__Kernel、调用接口与 PrivateUse1 注册重建安装回到示例目录重复python3 -m build --wheel -n与 pip 安装测试仿照tests/add/test_add.py编写 pytest跑通即上线。测试验证一次覆盖 54 组用例 ✅官方测试文件tests/add/test_add.py的做法很值得借鉴接口存在性测试断言torch.ops.ascend_ops.add可被发现——这能拦截schema 签名与 C 注册不一致导致算子被框架隐藏这类隐蔽问题功能参数化测试18 种 shape × 3 种数据类型float32 / float16 / int32的笛卡尔组合NPU 结果与 CPU 基准逐组比对浮点用torch.allclose校验。直接执行pytest tests/add/test_add.py即可回归你的算子。常见问题速查 ️为什么torch.ops里找不到我的算子九成是 schema 字符串参数名/类型/重载与 C 注册签名对不上对照add.cpp的m.def(...)与m.impl(...)逐项核对。编译报错找不到 CANN/Torch 路径确认已sourceCANN 环境变量set_env.sh并检查 torch、torch_npu 版本是否满足要求。改了代码没生效先python setup.py clean再重建 wheel最后加--force-reinstall安装。延伸学习 示例完整文档examples/fast_kernel_launch_example/README.md复杂算子接入参考csrc/grouped_matmul/分组矩阵乘、csrc/moe_distribute_dispatch_v2/MoE 分发算子零基础算子开发入门标准 op_host/op_kernel 流程docs/QUICKSTART.md算子贡献流程项目根目录CONTRIBUTING.md掌握快速核函数启动模式后一个简易 NPU 算子从想法到上线只需一个文件、两条命令——这正是 ops-transformer 为算子快速迭代提供的最高效路径。【免费下载链接】ops-transformer本项目是CANN提供的transformer类大模型算子库实现网络在NPU上加速计算。项目地址: https://gitcode.com/cann/ops-transformer创作声明：本文部分内容由AI辅助生成（AIGC），仅供参考
图像增强原理与实战)
![[光学原理与应用-521]：对光的错误理解与纠偏](http://pic.xiahunao.cn/yaotu/[光学原理与应用-521]：对光的错误理解与纠偏)
