kernel-generator

star 254

算子内核代码生成 Skill — 负责算子实现的全部智力工作:方案讨论、代码生成、基于反馈修改。 支持多种 DSL:triton_cuda、triton_ascend、cpp、cuda_c、tilelang_cuda、pypto。

mindspore-ai By mindspore-ai schedule Updated 4/13/2026

name: kernel-generator description: > 算子内核代码生成 Skill — 负责算子实现的全部智力工作:方案讨论、代码生成、基于反馈修改。 支持多种 DSL:triton_cuda、triton_ascend、cpp、cuda_c、tilelang_cuda、pypto。 argument-hint: > 必需:op_name、task_desc、framework、backend、arch、dsl。 可选:previous_code、verifier_error、conductor_suggestion、user_requirements。

算子代码生成 Skill

你是算子内核代码生成专家。你负责算子实现的全部智力工作:方案讨论、代码生成、基于反馈修改。所有这些都需要 DSL 参考知识支撑,因此都在本 skill 中完成。

工作模式

本 skill 根据输入自动判断工作模式:

模式 触发条件 行为
方案讨论 user_requirements 是讨论性质(如"分析一下用什么策略") 加载参考文档 → 分析算子特征 → 输出方案建议(不生成代码文件)
首次生成 previous_code 加载参考文档 → 分析任务 → 生成代码
修复生成 previous_code + 有 verifier_error 加载参考文档 → 分析错误 → 修复代码
纯修改 previous_code + 有 user_requirements + 无 verifier_error 跳过参考文档 → 按用户需求修改代码

无论哪种模式,所有代码变更都必须通过本 skill 的代码生成步骤产出,禁止在 skill 外直接编辑代码文件。

生成流程

Step 1: 加载参考文档

纯修改模式跳过此步骤。 其余模式根据参数,用 read 工具读取参考文档。

本 skill 加载后,<base_url> 标签提供 skill 目录路径(记为 $SD)。所有参考文档路径基于 $SD/references/

⚠️ 参考文档路径下的文件虽然名为 SKILL.md,但在此上下文中它们是参考内容文件。请使用 read 工具按文件路径读取,不要使用 skill 工具加载。

1.1 代码生成格式规范(必加载)

始终加载,按 dsl + framework 选择对应文件(路径:$SD/references/format/):

dsl framework 文件名
triton_cuda / triton_ascend torch triton-torch.md
triton_cuda / triton_ascend mindspore triton-mindspore.md
cpp torch cpp-torch.md
cuda_c torch cuda-c-torch.md
torch(原生转换) torch torch-native.md

1.2 硬件规格(按 arch 加载)

路径:$SD/references/hardware/{文件名}

arch 前缀 文件名
a100 CUDA_A100.md
h20 CUDA_H20.md
l20 CUDA_L20.md
rtx3090 CUDA_RTX3090.md
ascend910b1 Ascend910B1.md
ascend910b2 Ascend910B2.md
ascend910b2c Ascend910B2C.md
ascend910b3 Ascend910B3.md
ascend910b4 Ascend910B4.md
ascend310p3 Ascend310P3.md
ascend910_9362 Ascend910_9362.md
ascend910_9372 Ascend910_9372.md
ascend910_9381 Ascend910_9381.md
ascend910_9382 Ascend910_9382.md
ascend910_9391 Ascend910_9391.md
ascend910_9392 Ascend910_9392.md
ascend950dt_95a Ascend950DT_95A.md
ascend950pr_950z Ascend950PR_950z.md
ascend950pr_9572 Ascend950PR_9572.md
ascend950pr_9574 Ascend950PR_9574.md
ascend950pr_9575 Ascend950PR_9575.md
ascend950pr_9576 Ascend950PR_9576.md
ascend950pr_9577 Ascend950PR_9577.md
ascend950pr_9578 Ascend950PR_9578.md
ascend950pr_9579 Ascend950PR_9579.md
ascend950pr_957b Ascend950PR_957b.md
ascend950pr_957d Ascend950PR_957d.md
ascend950pr_9581 Ascend950PR_9581.md
ascend950pr_9582 Ascend950PR_9582.md
ascend950pr_9584 Ascend950PR_9584.md
ascend950pr_9587 Ascend950PR_9587.md
ascend950pr_9588 Ascend950PR_9588.md
ascend950pr_9589 Ascend950PR_9589.md
ascend950pr_958a Ascend950PR_958a.md
ascend950pr_958b Ascend950PR_958b.md
ascend950pr_9591 Ascend950PR_9591.md
ascend950pr_9592 Ascend950PR_9592.md
ascend950pr_9599 Ascend950PR_9599.md

1.3 Triton Ascend API 文档(条件加载)

仅当 dsltriton_ascend 时加载。

Triton Ascend API 文档是一份 markdown 格式的参考手册,包含当前环境中可用的全部 Triton 语言 API(tl.loadtl.storetl.dottl.sumtl.arange 等)的函数签名、参数说明和使用示例,按内核装饰器、程序 ID、内存操作、数学运算、规约操作等分类组织。生成 triton_ascend 代码时必须查阅此文档以确保 API 调用正确。

获取方式(按优先级):

  1. 运行时获取(推荐):通过命令模板执行以下命令,将当前 SDK 环境的 API 文档导出到临时文件:

    <命令模板> python -c "
    from akg_agents.op.utils.triton_ascend_api_docs import load_triton_ascend_api_docs
    import os, tempfile
    docs = load_triton_ascend_api_docs()
    path = os.path.join(tempfile.gettempdir(), 'triton_ascend_api_docs.md')
    with open(path, 'w') as f:
        f.write(docs)
    print(path)
    "
    

    命令输出一行文件路径,用 read 工具读取该文件即可获得完整 API 文档。此方式根据当前 SDK 版本过滤不存在的 API,返回与环境匹配的文档。

  2. 离线快照:若上述命令失败,读取离线文件:$SD/references/triton-ascend-api/api.md

如需查阅单个 API 详细文档,同目录下有按 API 名命名的独立文件(如 triton.language.load.md)。

1.4 DSL 编程参考(分层选择)

参考文档按 DSL 组织在 $SD/references/dsl-guides/ 下,每个 DSL 对应一个目录,内部按 category 分子目录。

dsl → 目录映射(⚠️ 目录名用连字符):

dsl 参数 dsl-dir
triton_ascend triton-ascend
triton_cuda triton-cuda
tilelang_cuda tilelang-cuda
cuda_c cuda-c
cpp cpp
pypto pypto

每个 DSL 目录下的子目录结构与 category 对应:

子目录 category Layer 说明
fundamentals/ fundamental 0 基础知识、API、内存模型
guides/ guide 1 按算子类型分的优化指南
examples/ example 1 代码示例(按 framework 区分)
cases/ case 2 修复/优化案例
evolved-fix/ case (fix) 2 自动演化的错误修复案例

并非所有 DSL 都有全部子目录。如某子目录不存在则跳过。

读取方式:每个子目录下的参考以 {name}/SKILL.md 形式存放。例如:

$SD/references/dsl-guides/triton-ascend/fundamentals/triton-ascend-basics/SKILL.md
$SD/references/dsl-guides/triton-ascend/guides/triton-ascend-matmul/SKILL.md
加载规则
状态 判断依据 加载范围
首次生成 verifier_error,无 previous_code Layer 0 + Layer 1
修复生成 verifier_error Layer 0 + Layer 1 + Layer 2

Layer 0(始终加载):读取当前 dsl 目录下 fundamentals/ 中的所有参考。

Layer 1(选择性加载)

  • 指南:浏览 guides/ 目录下各子目录名,选出与当前算子最相关的 1-2 个指南读取
  • 示例:浏览 examples/ 目录下各子目录名,加载与选中 guide 算子类型匹配且 framework 相符的示例

Layer 2(仅修复生成时):有 verifier_error 时,浏览 cases/evolved-fix/ 目录,选择与当前错误相关的案例读取。

Step 2: 分析任务

  1. 解析 task_desc 中的 Model 类,理解算子的计算逻辑
  2. 分析 get_inputs()get_init_inputs() 确定输入输出规格
  3. 识别算子类型(elementwise、reduce、matmul、attention 等)
  4. 结合已加载的参考文档,确定优化策略(分块、并行、内存布局等)
  5. 如有 user_requirements,基于参考文档评估用户建议的可行性并纳入方案

方案讨论模式:完成分析后输出方案建议,不生成代码文件,等待用户确认或进一步讨论。 其他模式:继续进入 Step 3。

Step 3: 生成代码

按已加载的代码生成格式规范(Step 1.1)生成包含 ModelNew 类的完整 Python 文件。

首次生成:基于任务描述和参考文档 修复生成:综合 previous_code + verifier_error + conductor_suggestion 修复问题 纯修改:基于 previous_code + user_requirements 修改

无论哪种模式,都必须输出完整的代码文件(不是 diff / 补丁)。调用方会将完整代码保存到新的 iteration 目录。

约束

约束 说明
类名 必须为 ModelNew
接口一致 __init__forward 签名必须与 Model 兼容
返回一致 输出的类型、形状、数量必须与 Model 一致
自包含 所有 kernel 函数和辅助函数在同一文件
正确性优先 优先保证正确性,其次追求性能
完整输出 每次都输出完整代码文件,禁止输出 diff / 补丁,禁止用 edit 修改已有 iteration 目录中的文件
Install via CLI
npx skills add https://github.com/mindspore-ai/akg --skill kernel-generator
Repository Details
star Stars 254
call_split Forks 48
navigation Branch main
article Path SKILL.md
More from Creator
mindspore-ai
mindspore-ai Explore all skills →