메뉴
Optimizes global, shared/LDS, and cache behavior for GPU kernels: coalescing, pipelining, bank conflicts, vector loads, and register vs shared tradeoffs. Use when NCU/rocprof shows low memory throughput, poor coalescing, or L2/LDS pressure.
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-tier2-memory-hierarchy |
| description | Optimizes global, shared/LDS, and cache behavior for GPU kernels: coalescing, pipelining, bank conflicts, vector loads, and register vs shared tradeoffs. Use when NCU/rocprof shows low memory throughput, poor coalescing, or L2/LDS pressure. |
AutoKernel Tier 2 focus: after launch geometry (Tier 1), maximize bytes moved usefully per cycle across HBM → L2 → L1/LDS → registers.
float4, uint4, Triton block_ptr with wide block_shape): fewer instructions, better line utilization when aligned.cp.async from global to shared; Hopper: TMA bulk tensor loads if using supported APIs/kernels.num_stages in Triton) to overlap load with math.+1 column) to skew addresses across banks.GROUP_SIZE_M) and problem splitting (Split-K, stream-K in Tier 4) change L2 reuse; pair Tier 2 vectorization with Tier 1 grouping.bank = (offset_in_bytes / 4) % 32 for classic layouts; 32 lanes hitting the same bank on the same cycle → serial replays.N_stride + 1 (FP32) is the minimal skew; larger pads sometimes help double-buffer pointer math stay conflict-free across stages.(row,col) to physical addresses so MMA loads spread across banks (see CUTLASS/Triton shared layouts).num_stages with buffer counts and verify bank conflict-free SMEM/LDS layouts (pad or swizzle).