メニュー
NVIDIA-specific caps and tuning: Hopper SM90 (WGMMA, TMA, cluster SMEM), Ampere SM80 (cp.async, TF32, 2:4 sparsity), Ada SM89 consumer (4090-class). Use when targeting a known compute capability or interpreting NCU metrics for a specific chip.
Codex または Claude でインストール この Prompt をコピーして Codex、Claude、または他のアシスタントに貼り付けると、Skill ページを確認してインストールできます。
SOC 職業分類に基づく
Optimize fused cross-entropy loss kernels in Triton for NVIDIA and AMD GPUs. Covers fused log-softmax + NLL, online log-sum-exp, and large vocabulary handling. Use when writing or optimizing cross-entropy, label-smoothed CE, or similar loss kernels.
Optimize FlashAttention-style fused attention kernels in Triton for NVIDIA and AMD GPUs. Covers online softmax, tiled QK/AV GEMM, causal masking, and memory-efficient attention. Use when writing or optimizing self-attention, cross-attention, or any QKV attention kernel.
Optimize dense matrix multiplication (GEMM) kernels in Triton for NVIDIA and AMD GPUs. Covers tiled blocking, L2 cache grouping, tensor core utilization, and dual-platform autotune. Use when writing or optimizing matmul, linear layers, or any GEMM-based kernel.
Optimize RMS Normalization kernels in Triton for NVIDIA and AMD GPUs. Covers row-parallel reduction, vectorized loads, warp shuffle patterns, and rsqrt usage. Use when writing or optimizing RMSNorm, LayerNorm, or similar normalization kernels.
Optimize Rotary Position Embedding (RoPE) kernels in Triton for NVIDIA and AMD GPUs. Covers interleaved/sequential layouts, sincos precomputation, and multi-row processing. Use when writing or optimizing RoPE, position embeddings, or rotary transformations.
Optimize fused softmax kernels in Triton for NVIDIA and AMD GPUs. Covers online max/sum reduction, multi-row processing, numerical stability, and warp-level patterns. Use when writing or optimizing softmax, log-softmax, or similar row-wise normalization kernels.
| name | gpu-opt-tier5-nvidia-arch |
| description | NVIDIA-specific caps and tuning: Hopper SM90 (WGMMA, TMA, cluster SMEM), Ampere SM80 (cp.async, TF32, 2:4 sparsity), Ada SM89 consumer (4090-class). Use when targeting a known compute capability or interpreting NCU metrics for a specific chip. |
AutoKernel Tier 5 maps generic techniques to SKU capabilities. This file is the NVIDIA branch.
| Arch | CC | Examples | Notes |
|---|---|---|---|
| Ampere | 8.0 | A100 | TF32 TC, cp.async, 2:4 structured sparsity (sparse TC) |
| Ada Lovelace | 8.9 | RTX 4090 | 16384 "CUDA cores" (marketing aggregate), ~1008 GB/s ref bandwidth class |
| Hopper | 9.0 | H100 | WGMMA, TMA, cluster SMEM, up to ~228 KB SMEM/SM (config-dependent) |
Exact SMEM/L1 splits and clocks vary by SKU; always verify in NCU "Launch Statistics" and occupancy calculator.
num_stages with occupancy.| Symptom in NCU | Likely lever |
|---|---|
Low sm__throughput + low dram__throughput | Occupancy / launch / tail (Tier 1/4) |
| High DRAM, low TC active | Memory path / vectorization (Tier 2) |
| High TC, reg spills | Tile size / WGMMA shape / fewer live tensors |
nvidia-smi, NCU header) and load SKU-specific peak FLOPS/BW for roofline.