triton-skill

Triton と Gluon の開発

API と実装パターンの主な参照先として、ローカルの Triton チェックアウトを使用してください。Triton と Gluon は急速に進化しているため、記憶しているシグネチャよりも、最新のチュートリアルとソースを優先してください。

チェックアウトの場所

この SKILL.md を含むディレクトリを特定し、その中の repos/triton/ を使用してください。インストーラーは、このパスを agent-gpu-skills/third_party/triton/ または TRITON_REPO を通じて指定されたチェックアウトにリンクします。

以下のコマンドでは、TRITON_REPO を特定した絶対パスに置き換えてください。

TRITON_REPO=/absolute/path/to/triton-skill/repos/triton

チェックアウトが見つからない場合は、agent-gpu-skills リポジトリから以下を実行し、Skill を再インストールしてください。

bash update-repos.sh triton
bash install.sh --skill triton-skill

参照するソースの選択

タスク	まず確認する場所
Triton 言語の構文と入門パターン	python/tutorials/
Gluon のレイアウトとアーキテクチャレベルのパターン	python/tutorials/gluon/
カーネルの完全な例	python/examples/
本番環境向けの matmul、reduction、top-k、SwiGLU	python/triton_kernels/triton_kernels/
tl.* の定義とセマンティクス	python/triton/language/
JIT、オートチューニング、ランタイムの動作	python/triton/runtime/
Python コンパイラーのエントリーポイント	python/triton/compiler/
Triton と GPU の方言定義	include/triton/Dialect/
コンパイラーの解析、変換、低レベル化	lib/

チュートリアル、完全な例、本番用カーネルの実装を選ぶ際は、quick-reference.md を読んでください。

確認の進め方

1. タスクが Triton、Gluon、コンパイラー内部のいずれに関するものかを特定します。
2. 対象の操作とアーキテクチャに最も近い、最新のチュートリアルまたは実装を探します。
3. 使用するすべての API を、その定義または別の最新の呼び出し箇所で確認します。
4. 対象ワークロードの形状、dtype、ストライド、レイアウト、マスキング、数値上の契約を維持します。
5. 起動パラメーターや最適化戦略を変更する前に、正しさを確認します。

記憶しているファイル名を当てにする前に、最新の例を探してください。

find "$TRITON_REPO/python/tutorials" -maxdepth 2 -type f | sort
find "$TRITON_REPO/python/examples" -type f | sort

Triton 言語の使用方法と定義を確認します。

rg -n 'tl\.dot|tl\.dot_scaled' "$TRITON_REPO/python/tutorials"
rg -n '@triton\.autotune' "$TRITON_REPO/python/tutorials"
rg -n '^def (load|store|dot|dot_scaled)' \
  "$TRITON_REPO/python/triton/language"

Gluon のアーキテクチャパターンを確認します。

rg -n '@gluon\.jit' "$TRITON_REPO/python/tutorials/gluon"
rg -n 'wgmma|tcgen05|mbarrier|tma' \
  "$TRITON_REPO/python/tutorials/gluon" \
  "$TRITON_REPO/python/examples"

本番用カーネルを追跡します。

rg -n 'persistent|TensorDescriptor' \
  "$TRITON_REPO/python/triton_kernels/triton_kernels/matmul_details"

rg -n 'mxfp|flexpoint' \
  "$TRITON_REPO/python/triton_kernels/triton_kernels/numerics_details"

コンパイラーの定義と低レベル化を追跡します。

rg -n 'def.*Op' "$TRITON_REPO/include/triton/Dialect/Triton/IR"
rg -n 'Encoding' "$TRITON_REPO/include/triton/Dialect/TritonGPU/IR"
rg -n 'wgmma|tma|tcgen05' \
  "$TRITON_REPO/include/triton/Dialect/TritonNvidiaGPU"
rg -n 'Pattern|Rewrite' "$TRITON_REPO/lib/Conversion/TritonGPUToLLVM"

実装時の原則

診断中は、次の各レイヤーを分けて考えてください。

Python kernel and launch metadata
  → Triton/Gluon IR and compiler transforms
  → generated GPU code on the selected target

ソースレベルのパターンだけでは、コンパイル後のカーネルが意図した命令やメモリーパスを使用しているとは限りません。この違いが重要な場合は、コンパイラーの出力またはプロファイルデータを調べてください。PTX のセマンティクス、Compute Capability、Nsight、Compute Sanitizer の詳細が必要な場合は、cuda-skill を追加してください。

正しさを検証する場合：

• 代表的な形状と境界形状について、独立した参照実装と比較します。
• マスクされた末尾、2 のべき乗でない次元、ストライド、dtype 変換をテストします。
• コンパイルエラー、実行時の正しさ、数値許容誤差を切り分けます。
• ケースを縮小する前に、元のディスパッチと起動メタデータを再現します。

パフォーマンスを検証する場合：

• ベンチマークの形状セットと測定方法を固定します。
• 最適な設定が変わるすべての次元を、オートチューニングのキーが網羅していることを確認します。
• タイル、ワープ、ステージ、永続化、特殊化の選択肢を一度に一つずつ変更します。
• 結果を TMA、WGMMA、tcgen05、ワープ特殊化に帰属させる前に、生成された実行パスを確認します。

ソースの更新

agent-gpu-skills リポジトリから実行します。

bash update-repos.sh triton
python3 scripts/validate_repo.py --require-sources

チェックアウトは Triton の main を追跡し、third_party/UPSTREAMS.toml には、この Skill が最後に採用したコミットが記録されています。その記録を更新する前に、ソースマップの差異を確認してください。