awesome-agent-kernel-skills

FlashInfer — High-performance kernel library for LLM inference with optimized attention, paged KV-cache, FP8/FP4 quantization

Iteratively optimize cuTile kernel performance through systematic profiling, bottleneck analysis, IR comparison, and targeted tuning. Covers tile sizes, occupancy, autotune configs, TMA, latency hints, persistent scheduling, num_ctas, flush_to_zero, and IR-level debugging. Use when asked to "optimize cutile kernel", "improve kernel perf", "tune cutile performance", "make kernel faster", or iteratively benchmark and refine a cuTile GPU kernel in the TileGym project.

Integrate TileGym kernels into Hugging Face `transformers` models by replacing the library's submodule(s) and certain class(es)' implementations, and patching certain class(es)' init/forward/load weight methods prior to instantiating models. Used when the user requires integrating TileGym kernels into `transformers` models.

Use when the user wants to add, modify, debug, or review an xLLM TileLang Ascend kernel or specialization, including Python kernel definitions, generated Ascend-C source, runtime wrapper dispatch, TileLang CMake wiring, and NPU tests.

TileLang-Ascend 新增 Ascend 专属 T.tile.xxx 小 API 的端到端开发流程。用户要求新增、封装、暴露、实现或测试 ascend_tile.py 中的 T.tile API / Ascend tile primitive 时必须使用本 skill，尤其适用于需要同时打通 Python 前端、C++ lowering/codegen、Ascend C helper、文档和 CI 测试的任务。

TileLang-Ascend 算子测试设计技能。支持多种场景：(1) 从 design.md 设计测试配置 (2) 从 examples/{op}/*.py 补充测试 (3) 手动提供算子信息生成测试 (4) 测试覆盖率分析。理解算子实现逻辑后智能判断测试策略。触发：设计算子测试、生成测试用例、补充测试、测试覆盖率不足。

Generates optimized CUDA kernel code based on performance analysis reports or algorithm type. Reads NCU analysis reports (e.g. *_analysis.md) and optionally existing kernel code, then produces high-quality compilable .cu files with applied optimizations. Use when the user provides an NCU analysis report or requests CUDA kernel generation, optimization, or implementation of techniques like Shared Memory Tiling, vectorized loads, bank conflict elimination, or double buffering. Does not handle compilation, execution, or profiling.

CUDA kernel development, debugging, performance optimization, linear algebra, and multi-GPU communication for Claude Code. Use when writing, debugging, or optimizing CUDA code, GPU kernels, parallel algorithms, or CUDA library calls. Covers cuBLAS/cuBLASLt GEMM operations, CUDA Math API (half, bfloat16, FP8, FP6, FP4), NCCL multi-GPU collectives, non-interactive profiling with nsys/ncu, debugging with cuda-gdb/compute-sanitizer, binary inspection with cuobjdump, and performance analysis workflows. Triggers on CUDA, GPU programming, kernel optimization, nsys, ncu, cuda-gdb, compute-sanitizer, PTX, GPU profiling, parallel performance, cuBLAS, cublasLtMatmul, GEMM, GemmEx, FP8, bfloat16, half precision, __half, __nv_bfloat16, cublasGemmEx, cublasGemmStridedBatchedEx, NCCL, ncclAllReduce, ncclReduceScatter, ncclAllGather, ncclCommInitRank, tensor parallel, pipeline parallel, all-reduce, vLLM CUDA kernels.

Orchestrates a full profiling-driven CUDA kernel optimization loop (write → validate → profile → analyze → optimize) until performance converges or no further gains are possible. Capabilities include generating reference implementations, writing initial kernels via cuda-code-generator, running correctness validation and benchmarks via kernel-benchmarker, profiling and analyzing NCU reports via ncu-rep-analyzer, and applying targeted optimizations via cuda-code-generator. Use when the user wants to optimize a .cu file or CUDA kernel, improve GPU kernel performance, run a CUDA optimization workflow.

Curated index of NVIDIA official CUDA Samples with code patterns, GitHub permalinks, and key snippets. Use when looking for working CUDA code examples, GPU kernel patterns (reduction, scan, GEMM, transpose, histogram), Tensor Core WMMA usage, CUDA Graphs API, cuBLAS/cuFFT/cuSPARSE library integration, stream/event async patterns, multi-GPU P2P/IPC, or framework interop (PyTorch/TensorFlow custom kernels). Covers 50+ curated samples organized by topic. Triggers on CUDA sample, code example, kernel pattern, GEMM example, reduction example, CUDA graph example, cuBLAS example, Tensor Core WMMA, how to write CUDA kernel, CUDA code reference.

This skill should be used when reviewing FlagGems operator PRs, performing code review, self-reviewing before submission, or when the user mentions "review PR", "审PR", "代码review", "code review", "review #123", "self-review", "自审", "检查PR", "审查算子", "review operator". It fetches PR diffs, applies FlagGems domain-specific review rules (structural checks, naming, registration, tests, benchmarks), and posts inline review comments directly on GitHub.

This skill should be used when submitting FlagGems operator PRs, reviewing operator code before submission, preparing operator code for PR, or when the user mentions "提PR", "提交算子", "submit operator", "PR提交", "代码审核", "pre-commit". It automates code review, validates completeness and compliance, runs pre-commit and worktree tests, and directly submits PR to upstream with full description including speedup data.

Compiles, validates, and benchmarks a CUDA kernel (.cu file) against a Python reference (*_ref.py). Auto-detects GPU arch and infers dimension args from the `extern "C" void solve(...)` signature. Runs benchmark.py to: compile with nvcc, optionally validate outputs (exits on failure), benchmark both kernel and reference with CUDA-event timing, and print a latency/bandwidth/speedup summary. Use when the user wants to validate or benchmark a .cu file.

Use when the user asks about optimizing NVIDIA Blackwell (SM100, B200) or Hopper (SM90, H100) GPU kernels — tcgen05/TMEM/CLC/NVFP4/2-SM cooperative, warp specialization, FlashAttention-4, DeepGEMM, FlashMLA, MoE, grouped GEMM, CuTe-DSL/PTX/Triton on Blackwell, or wants concrete PR references from CUTLASS/SGLang/vLLM/FlashInfer/PyTorch. Do NOT use for generic CUDA Q&A that is not Blackwell/Hopper-specific, host-side framework integration, or distributed systems (DeepEP/EPLB/DualPipe).

Profiles a CUDA kernel with NCU and analyzes the resulting .ncu-rep report to diagnose performance bottlenecks and generate optimization recommendations. Use when the user provides a .cu file or a .ncu-rep file and asks for performance analysis, NCU profiling, or bottleneck diagnosis. If given a .cu file, runs NCU via benchmark.py to produce a .ncu-rep, then imports it with `ncu --import` to extract metrics (SM throughput, DRAM/L1 bandwidth, occupancy), classifies the bottleneck (DRAM-bound, compute-bound, latency-bound, etc.), and saves a structured *_analysis.md report alongside the .ncu-rep file.

Query NVIDIA PTX ISA 9.1, CUDA Runtime API 13.1, Driver API 13.1, Programming Guide v13.1, Best Practices Guide, Nsight Compute, Nsight Systems local documentation. Debug and optimize GPU kernels with nsys/ncu/compute-sanitizer workflows. Use when writing, debugging, or optimizing CUDA code, GPU kernels, PTX instructions, inline PTX, TensorCore operations (WMMA, WGMMA, TMA, tcgen05), or when the user mentions CUDA API functions, error codes, device properties, memory management, profiling, GPU performance, compute capabilities, CUDA Graphs, Cooperative Groups, Unified Memory, dynamic parallelism, CUDA programming model concepts, bank conflicts, shared memory optimization, warp divergence, memory coalescing, occupancy tuning, register pressure, L2 cache control, async copy, mbarrier, thread block clusters, or CUDA architecture questions (Ampere sm_80, Hopper sm_90, Blackwell sm_100).