# LightOp KernelPilot
**基于 Humanize 的 LightOp/DCU 算子开发与优化 skill 包。**
LightOp KernelPilot 是面向 LightOp DCU 算子库的 KernelPilot 工作流改造版。
它保留了自主算子开发循环里真正有用的部分：明确算子语义、正确性参考、
workload 分布、benchmark 证据、profile digest、尝试记录、优化记录，以及
带 review gate 的迭代。
它移除了 NVIDIA 优先的假设，例如 Nsight Compute、CUTLASS/CuTe、PTX/SASS、
Blackwell/Hopper、TMA、WGMMA、tcgen05 等。默认面向 DCU/ROCm/HIP/DTK 环境。
## Skills
| Skill | 作用 |
| --- | --- |
| [`lightop-kernel-agent-loop`](humanize/skills/humanize-kernel-agent-loop/) | 主循环。用于新增或优化 LightOp 算子：恢复 `K/R/W/E`，检查 wrapper/binding/kernel/test/config/benchmark，实现 HIP/ROCm 改动，build、test、benchmark、profile、tune，并启动 Humanize RLCR。 |
| [`lightop-kernel-knowledge`](knowledge/SKILL.md) | 证据检索。优先查本地 LightOp 源码，其次查 ROCm/DCU 上游和官方文档，最后才把 bundled CUDA PR corpus 当作跨平台灵感。 |
| [`dcu-profiler-report`](humanize/skills/ncu-report/) | 性能剖析。把 `hipprof`、ROCm/DTK profiler、benchmark log 和可选 AMDGPU ISA 证据整理成可复现 digest，并给出一个明确的下一步 LightOp 修改。 |
磁盘上的目录名 `humanize-kernel-agent-loop` 和 `ncu-report` 是为了兼容上游
Humanize installer；真正暴露给 agent 的 skill 名称以上表 frontmatter 为准。
## 手动安装
如果没有 Claude 或 Codex CLI installer，可以直接安装这三个 LightOp/DCU
skills：
```bash
./install-lightop-skills-manual.sh --target both
```
只安装 Claude：
```bash
./install-lightop-skills-manual.sh --target claude
```
只安装 Codex：
```bash
./install-lightop-skills-manual.sh --target codex
```
默认路径：
```text
Claude: ~/.claude/skills
Codex: ${CODEX_HOME:-~/.codex}/skills
```
脚本会：
- symlink `lightop-kernel-knowledge`
- symlink `dcu-profiler-report`
- hydrate `lightop-kernel-agent-loop` 中的 `{{HUMANIZE_RUNTIME_ROOT}}` 和
`{{KERNELPILOT_ROOT}}`
- 可选安装 `knowledge/requirements.txt`
## 请求格式
一个清晰的请求最好包含：算子名、正确性参考、workload、执行环境、目标
DCU/gfx arch、baseline、benchmark 方法、成功阈值。
示例：
```text
[$lightop-kernel-agent-loop] 给 LightOp 添加 fused rmsnorm + rope + fp8
kv-cache store 算子，目标 gfx936。正确性参考使用 PyTorch/native LightOp
组合路径，覆盖 Qwen decode 的 batch/token/head_dim shape。验证环境使用
Docker 容器 lightop-dtk，容器内 repo 路径是 /workspace/lightop。
性能要求：p50 latency 比现有 unfused 路径快 15%。
```
优化已有算子的示例：
```text
[$lightop-kernel-agent-loop] 优化 gfx938 上的 lightop.moe_gemm_w8a8，
目标 workload 是 DeepSeek EP8 decode。保持现有 Python API 不变，
和当前 LightOp baseline 对比；benchmark plateau 时使用 hipprof 证据继续分析。
```
如果宿主机路径和容器路径不同，建议直接写清楚：
```text
宿主机 LightOp 路径：/public/wanghl6/lightop
验证容器：wanghl_lightop209
容器内 LightOp 路径：/home/lightop
所有 build、correctness test、benchmark、profiling 都必须在容器内执行：
docker exec wanghl_lightop209 bash -lc 'cd /home/lightop && '
```
## LightOp 接入位置
主循环会在一个 LightOp checkout 中工作。该目录通常包含：
```text
setup.py
setup_torch29.py
lightop/__init__.py
lightop/csrc/export.cpp
lightop/csrc//
lightop/config*.py
test/
```
新增算子时，agent 通常会检查或修改：
- `lightop/csrc//`：HIP/C++ kernel 和 launcher
- `lightop/csrc/export.cpp`：`m.def(...)` binding
- `lightop/.py`：Python wrapper
- `lightop/__init__.py`：公开 API export
- `setup.py`：只有新增 `csrc` family 需要 glob 覆盖时才改
- `test/test_.py`：正确性测试
- benchmark 脚本：性能测试
- `lightop/config*.py`：需要 shape/gfx-aware dispatch 时才改
LightOp KernelPilot 的 build 规则固定为：
```bash
python setup.py install
```
无论 PyTorch 版本是什么，都不切到 `setup_torch29.py`。正常调优循环中也不删除
`build/`，以便复用增量编译结果；只有用户明确要求 clean build，或证明 cache
损坏时才清理。
## DCU Profiling
`dcu-profiler-report` 使用 SourceFind DCU 性能分析工具指南作为一阶段官方参考：
```text
https://developer.sourcefind.cn/gitbook//dcu_developer/DeveloperGuide/dcu_programming/DCU_programming_chapter3_7.html
```
默认一阶段 profiling：
```bash
cd /path/to/lightop
mkdir -p .humanize/lightop-agent/profile-artifacts/v000_baseline
python test/test_.py 2>&1 \
| tee .humanize/lightop-agent/profile-artifacts/v000_baseline/benchmark.log
hipprof python test/test_.py 2>&1 \
| tee .humanize/lightop-agent/profile-artifacts/v000_baseline/hipprof.txt
```
当 `hipprof` 和 benchmark log 不够解释问题时，可以进一步使用：
```text
rocprof
rocprofv3
rocprof-compute
AMDGPU ISA / code-object inspection
```
profile digest 不是只写一句 “memory-bound”。它必须给出：
```text
测到什么信号 -> 可能机制是什么 -> 为什么其他假设较弱 -> 下一步具体改哪里
```
## Build And Test 默认规则
build、install、test、benchmark、profiling 必须在同一个环境里完成。如果目标环境是
Docker，优先使用可重复的非交互命令：
```bash
docker exec bash -lc 'cd /workspace/lightop && '
```
在该环境的 LightOp root 下先探测版本和设备：
```bash
python - <<'PY'
import torch
print("torch:", torch.__version__)
print("hip:", torch.version.hip)
print("device:", torch.cuda.get_device_name(0))
print("gcn:", torch.cuda.get_device_properties(0).gcnArchName)
PY
hipcc --version
```
然后 build：
```bash
PYTORCH_ROCM_ARCH='gfx928;gfx936;gfx938' python setup.py install
```
再做 import smoke test：
```bash
python - <<'PY'
import torch, lightop
print("lightop:", getattr(lightop, "__file__", "unknown"))
print("gcn:", torch.cuda.get_device_properties(0).gcnArchName)
PY
```
优先跑最窄的目标测试：
```bash
cd test
python test_.py
```
benchmark 必须包含 warmup，并且在计时区间前后显式使用
`torch.cuda.synchronize()` 或 HIP 同步。不能只凭 Python wall-clock 异步计时宣布
加速。
一个 LightOp 算子改动不能只靠 build 通过就算完成。完成条件至少包括：
- install 成功
- import smoke test 成功
- targeted correctness test 成功
- benchmark 和 baseline 对比完成
- 当结果接近阈值或异常时，有 profiler 证据
## 常规安装
从当前 checkout 安装 Humanize skill pack：
```bash
cd /path/to/kernel-pilot
./humanize/scripts/install-skill.sh --target codex --kernelpilot-root "$PWD"
```
Claude Code 用户可以使用：
```bash
./humanize/scripts/install-skills-claude.sh --kernelpilot-root "$PWD"
```
安装后相关 skill 名称是：
```text
lightop-kernel-agent-loop
lightop-kernel-knowledge
dcu-profiler-report
```
## 证据规则
- 本地 LightOp 源码、测试、配置和 benchmark 是第一优先级证据。
- ROCm/DCU 官方文档和上游源码是第二优先级证据。
- bundled CUDA PR corpus 只能作为跨平台灵感，除非已经明确翻译并在 DCU 上验证。
- 复制或改写外部源码时，必须记录来源路径/URL、commit 或 version、
license/notice，以及优化收益。
- profile digest 最后必须给出一个具体的下一步修改；如果 profiling 不可执行，
要说明原因。