Skip to content

Add GDN benchmarking - #501

Merged
Anerudhan merged 3 commits into
NVIDIA:developfrom
jhjpark:jopark/gdn_benchmark
Aug 6, 2026
Merged

Add GDN benchmarking#501
Anerudhan merged 3 commits into
NVIDIA:developfrom
jhjpark:jopark/gdn_benchmark

Conversation

@jhjpark

@jhjpark jhjpark commented Aug 6, 2026

Copy link
Copy Markdown
Collaborator

Before submitting

  • I agree to license this contribution under the terms of LICENSE.txt.
  • I ran pre-commit run and committed any formatting changes.
  • I added GitHub labels: one cat-*, one or more mod-*, and one orig-* (see label list).

Affected area

  • Benchmarks or performance

Summary

Add GDN benchmark results and FlashQLA as a backend for benchmarking.

Why

Related issues

API and compatibility impact

Testing

Summary by CodeRabbit

  • New Features
    • Added FlashQLA as a selectable backend for linear-attention benchmarks, supporting the GDN variant and grouped-value attention.
    • Added automated forward and backward TFLOPS comparison charts with per-batch PNG output and optional hardware labels.
  • Documentation
    • Documented FlashQLA support, supported configurations, benchmark usage, and measurement guidance.
  • Chores
    • Updated the benchmark environment for the latest runtime and FlashQLA installation.

@coderabbitai

coderabbitai Bot commented Aug 6, 2026

Copy link
Copy Markdown
Contributor

Review Change Stack

No actionable comments were generated in the recent review. 🎉

ℹ️ Recent review info
⚙️ Run configuration

Configuration used: Path: .coderabbit.yaml

Review profile: CHILL

Plan: Enterprise

Run ID: 975dbcae-3d5c-483b-8010-df2d1ae78ed6

📥 Commits

Reviewing files that changed from the base of the PR and between c34787b and b5d22e4.

📒 Files selected for processing (1)
  • benchmark/linear_attention/Dockerfile
🚧 Files skipped from review as they are similar to previous changes (1)
  • benchmark/linear_attention/Dockerfile

📝 Walkthrough

Walkthrough

The benchmark image now installs FlashQLA and its runtime dependencies. The linear-attention benchmark supports FlashQLA for GDN workloads. A new CLI utility generates per-batch performance charts from benchmark CSV files.

Changes

FlashQLA benchmark integration

Layer / File(s) Summary
FlashQLA runtime and benchmark usage
benchmark/linear_attention/Dockerfile, benchmark/linear_attention/README.md
The container installs Cutlass DSL 4.7.0, apache-tvm-ffi, FlashQLA, and plotting packages. The README documents FlashQLA commands, supported configurations, and FLA dispatch settings.
FlashQLA execution path
benchmark/linear_attention/benchmark_single_linear_attention.py
The CLI validates FlashQLA as a gdn backend without GQA. The benchmark invokes flash_qla.chunk_gated_delta_rule with dense tensors and preserves its output layout.
Benchmark chart generation
benchmark/linear_attention/plot_results.py
The new utility validates CSV data, filters benchmark results, and creates forward and backward TFLOPS charts with configurable labels and batch sizes.

Estimated code review effort: 3 (Moderate) | ~25 minutes

Sequence Diagram(s)

sequenceDiagram
  participant BenchmarkCLI
  participant run_benchmark
  participant flash_qla
  BenchmarkCLI->>run_benchmark: select and validate flash_qla
  run_benchmark->>flash_qla: pass dense GDN tensors
  flash_qla-->>run_benchmark: return attention output
  run_benchmark-->>BenchmarkCLI: preserve output for validation
Loading
🚥 Pre-merge checks | ✅ 3 | ❌ 2

❌ Failed checks (2 warnings)

Check name Status Explanation Resolution
Description check ⚠️ Warning The description includes the affected area and summary, but the Why, Related issues, API impact, and Testing sections are incomplete. Complete the Why, Related issues, API and compatibility impact, and Testing sections with relevant details and exact test commands and results.
Docstring Coverage ⚠️ Warning Docstring coverage is 18.18% which is insufficient. The required threshold is 80.00%. Write docstrings for the functions missing them to satisfy the coverage threshold.
✅ Passed checks (3 passed)
Check name Status Explanation
Title check ✅ Passed The title clearly identifies the main change: adding GDN benchmarking.
Linked Issues check ✅ Passed Check skipped because no linked issues were found for this pull request.
Out of Scope Changes check ✅ Passed Check skipped because no linked issues were found for this pull request.
✨ Finishing Touches
🧪 Generate unit tests (beta)
  • Create PR with unit tests

Comment @coderabbitai help to get the list of available commands.

@coderabbitai coderabbitai Bot left a comment

Copy link
Copy Markdown
Contributor

Choose a reason for hiding this comment

The reason will be displayed to describe this comment to others. Learn more.

Actionable comments posted: 3

🧹 Nitpick comments (1)
benchmark/linear_attention/plot_results.py (1)

56-63: 📐 Maintainability & Code Quality | 🔵 Trivial | ⚡ Quick win

Use explicit optional annotations for cudnn_version and batch_sizes. Import Optional, then use Optional[str] for both cudnn_version parameters and Optional[list] for batch_sizes.

🤖 Prompt for AI Agents
Verify each finding against current code. Fix only still-valid issues, skip the
rest with a brief reason, keep changes minimal, and validate.

In `@benchmark/linear_attention/plot_results.py` around lines 56 - 63, Update the
type annotations in get_backend_display_name and generate_charts to use explicit
Optional types: import Optional, annotate both cudnn_version parameters as
Optional[str], and annotate generate_charts’ batch_sizes as Optional[list].

Source: Linters/SAST tools

🤖 Prompt for all review comments with AI agents
Verify each finding against current code. Fix only still-valid issues, skip the
rest with a brief reason, keep changes minimal, and validate.

Inline comments:
In `@benchmark/linear_attention/Dockerfile`:
- Around line 22-27: Pin the benchmark environment in the Dockerfile around the
FlashQLA installation: check out a specific FlashQLA tag or commit, require
flash-linear-attention==0.5.0, and pin cudnn-frontend to a specific source
revision. Add and apply a constraints or lock file covering these packages and
all transitive dependencies, including tilelang==0.1.9 and
apache-tvm-ffi==0.1.9, while recording the tested versions.

In `@benchmark/linear_attention/plot_results.py`:
- Around line 102-103: Update the chart title in the plotting flow around
fig.suptitle to use the selected variant value instead of hardcoding “GDN”,
while preserving the existing batch, heads, head dimension, and GPU information.
- Around line 19-24: Add pandas, matplotlib, and seaborn to the dependencies
installed by benchmark/linear_attention/Dockerfile so the imports in
plot_results.py resolve successfully when the linear-attention image runs.

---

Nitpick comments:
In `@benchmark/linear_attention/plot_results.py`:
- Around line 56-63: Update the type annotations in get_backend_display_name and
generate_charts to use explicit Optional types: import Optional, annotate both
cudnn_version parameters as Optional[str], and annotate generate_charts’
batch_sizes as Optional[list].
🪄 Autofix

Fix all unresolved CodeRabbit comments on this PR:

  • Push a commit to this branch (recommended)
  • Create a new PR with the fixes

ℹ️ Review info
⚙️ Run configuration

Configuration used: Path: .coderabbit.yaml

Review profile: CHILL

Plan: Enterprise

Run ID: 15da04bd-6e11-468b-9b7e-bb3d19890418

📥 Commits

Reviewing files that changed from the base of the PR and between 9f45b52 and a4640ea.

⛔ Files ignored due to path filters (2)
  • benchmark/linear_attention/results/gdn/b300/gdn.png is excluded by !**/*.png
  • benchmark/linear_attention/results/gdn/b300/gdn_20260806.csv is excluded by !**/*.csv
📒 Files selected for processing (4)
  • benchmark/linear_attention/Dockerfile
  • benchmark/linear_attention/README.md
  • benchmark/linear_attention/benchmark_single_linear_attention.py
  • benchmark/linear_attention/plot_results.py

Comment on lines +22 to +27
# Install the Cutlass DSL runtime (cuDNN FROST engines) and FLA.
RUN pip install nvidia-cutlass-dsl[cu13]==4.7.0 apache-tvm-ffi flash-linear-attention

# Install FlashQLA from source.
RUN git clone https://github.com/QwenLM/FlashQLA.git
RUN pip install -v /workspace/FlashQLA

Copy link
Copy Markdown
Contributor

Choose a reason for hiding this comment

The reason will be displayed to describe this comment to others. Learn more.

🗄️ Data Integrity & Integration | 🟠 Major | ⚡ Quick win

🧩 Analysis chain

🏁 Script executed:

#!/bin/bash
set -eu

printf '%s\n' '--- Dockerfile ---'
cat -n benchmark/linear_attention/Dockerfile

printf '%s\n' '--- dependency and benchmark references ---'
rg -n -i --glob '!node_modules' --glob '!dist' \
  'FlashQLA|flash-linear-attention|apache-tvm-ffi|nvidia-cutlass-dsl|constraints|requirements|lock' \
  benchmark README* .github 2>/dev/null || true

printf '%s\n' '--- tracked files near the benchmark ---'
git ls-files benchmark/linear_attention

Repository: NVIDIA/cudnn-frontend

Length of output: 36203


🌐 Web query:

QwenLM FlashQLA GitHub README benchmark instructions pinned dependencies release notes API alignment

💡 Result:

FlashQLA is a high-performance linear attention library by QwenLM designed to accelerate Gated Delta Networks (GDN) [1]. Benchmark Instructions To benchmark FlashQLA, you must have the required dependencies installed, including flash-linear-attention (FLA) and flashinfer [2][3]. The benchmark scripts are located in the benchmark directory [2][4]. For example, to run the gated delta rule benchmark, ensure the correct versions are installed and run the script [2][4]: pip install flash_linear_attention==0.5.0 flashinfer-python==0.6.13 python benchmark/bench_gated_delta_rule.py Detailed benchmark results for H200 (SM90) and GB200 (SM100) are available in the repository's benchmark directory [2][5]. Pinned Dependencies The project enforces specific dependency versions for compatibility [1]. As of version 0.1.2, these include: - torch: >=2.8 [1] - tilelang: ==0.1.9 [1] - apache-tvm-ffi: ==0.1.9 [1] - flash_linear_attention: ==0.5.0 (for comparison/API alignment) [2][3] Release Notes - v0.1.2 (July 2026): Added forward pass support for SM120 (Blackwell) architectures and established the library as a backend for the flash-linear-attention GDN, enabling plug-and-play acceleration [1]. - v0.1.1 (June 2026): Introduced intra-card sequence parallelism for the backward pass and added SM100 architecture support [5][1]. This release also upgraded the tilelang dependency to v0.1.9 and aligned entry function signatures with the latest flash-linear-attention interface [5][6][1]. API Alignment FlashQLA is aligned with the flash-linear-attention (FLA) API, allowing it to serve as a high-performance backend [1]. The high-level API is accessed via chunk_gated_delta_rule, and the low-level API provides separate forward (chunk_gated_delta_rule_fwd) and backward (chunk_gated_delta_rule_bwd) functions, mirroring standard interfaces used in linear attention research [1].

Citations:


🏁 Script executed:

#!/bin/bash
set -eu

cat -n benchmark/linear_attention/Dockerfile
rg -n -i 'FlashQLA|flash-linear-attention|apache-tvm-ffi|nvidia-cutlass-dsl|constraints|requirements|lock' \
  benchmark README* .github 2>/dev/null || true
git ls-files benchmark/linear_attention

Repository: NVIDIA/cudnn-frontend

Length of output: 36099


🌐 Web query:

QwenLM FlashQLA GitHub README benchmark pinned dependencies release notes API alignment

💡 Result:

FlashQLA is a high-performance linear attention kernel library developed by QwenLM that leverages the TileLang compiler to optimize GDN (Gated Delta Network) Chunked Prefill operations [1][2]. Key technical aspects include: Benchmarks: The library demonstrates significant performance improvements over the FLA (Flash Linear Attention) Triton kernel, achieving 2-3x speedup in forward passes and 2x speedup in backward passes on NVIDIA Hopper (SM90) and Blackwell (SM100) architectures [2][3]. Detailed benchmark results for H200 and GB200 systems are provided in the repository's benchmark directory [3]. Pinned Dependencies: The project maintains strict dependency requirements to ensure compatibility with its TileLang-based kernels. According to setup.py and README documentation, the library requires: - Python >= 3.10 [4] - PyTorch >= 2.8 [2][4] - TileLang == 0.1.9 [4] - Apache-tvm-ffi == 0.1.9 [4] - CUDA 12.8 or above [2] Release Notes & Alignment: Recent updates, such as v0.1.1, introduced support for the SM100 architecture and the state_v_first parameter [5][3]. A key focus in recent releases has been aligning the entry function signatures with the latest interfaces from the flash-linear-attention library, ensuring better interoperability [5][3]. API: FlashQLA provides both high-level and low-level APIs [1][6]: - High-level: Access via chunk_gated_delta_rule, which handles the core forward/backward computations and state management [1]. - Low-level: Access via separate chunk_gated_delta_rule_fwd and chunk_gated_delta_rule_bwd functions for fine-grained control over forward and backward passes [1]. The project is released under the MIT License [1][4].

Citations:


🏁 Script executed:

#!/bin/bash
set -eu

python3 - <<'PY'
import json
import urllib.request

urls = {
    "tags": "https://api.github.com/repos/QwenLM/FlashQLA/tags?per_page=20",
    "setup": "https://raw.githubusercontent.com/QwenLM/FlashQLA/main/setup.py",
    "readme": "https://raw.githubusercontent.com/QwenLM/FlashQLA/main/README.md",
    "releases": "https://api.github.com/repos/QwenLM/FlashQLA/releases?per_page=10",
}
for name, url in urls.items():
    print(f"--- {name} ---")
    request = urllib.request.Request(url, headers={"Accept": "application/vnd.github+json"})
    with urllib.request.urlopen(request) as response:
        data = response.read().decode()

    if name in {"tags", "releases"}:
        for item in json.loads(data):
            if name == "tags":
                print(item["name"], item["commit"]["sha"])
            else:
                print(item["tag_name"], item["target_commitish"])
    else:
        for number, line in enumerate(data.splitlines(), 1):
            if any(term in line.lower() for term in (
                "install_requires", "apache-tvm-ffi",
                "flash_linear_attention", "tilelang",
                "pip install", "benchmark", "0.1.",
            )):
                print(f"{number}: {line}")
PY

cat -n benchmark/linear_attention/README.md | sed -n '45,110p'

Repository: NVIDIA/cudnn-frontend

Length of output: 2459


🏁 Script executed:

#!/bin/bash
set -eu

fetch() {
  curl -k -fsSL --retry 2 "$1"
}

printf '%s\n' '--- tags ---'
fetch 'https://api.github.com/repos/QwenLM/FlashQLA/tags?per_page=20' |
  jq -r '.[] | [.name, .commit.sha] | `@tsv`'

printf '%s\n' '--- releases ---'
fetch 'https://api.github.com/repos/QwenLM/FlashQLA/releases?per_page=10' |
  jq -r '.[] | [.tag_name, .target_commitish] | `@tsv`'

printf '%s\n' '--- setup.py dependency lines ---'
fetch 'https://raw.githubusercontent.com/QwenLM/FlashQLA/main/setup.py' |
  nl -ba |
  grep -Ei 'apache-tvm-ffi|flash[_-]linear[_-]attention|tilelang|install_requires|version'

printf '%s\n' '--- README install and benchmark lines ---'
fetch 'https://raw.githubusercontent.com/QwenLM/FlashQLA/main/README.md' |
  nl -ba |
  grep -Ei -C 3 'apache-tvm-ffi|flash[_-]linear[_-]attention|tilelang|benchmark|pip install'

Repository: NVIDIA/cudnn-frontend

Length of output: 406


🏁 Script executed:

#!/bin/bash
set -eu

fetch() {
  curl -k -fsSL --retry 2 "$1"
}

for ref in v0.1.2 main; do
  printf '%s\n' "--- setup.py at $ref ---"
  fetch "https://raw.githubusercontent.com/QwenLM/FlashQLA/$ref/setup.py" |
    grep -Ei -C 3 'apache-tvm-ffi|flash[_-]linear[_-]attention|tilelang|install_requires|version' || true

  printf '%s\n' "--- README at $ref ---"
  fetch "https://raw.githubusercontent.com/QwenLM/FlashQLA/$ref/README.md" |
    grep -Ei -C 3 'apache-tvm-ffi|flash[_-]linear[_-]attention|tilelang|benchmark|pip install' || true
done

Repository: NVIDIA/cudnn-frontend

Length of output: 8408


Pin the complete benchmark dependency set.

Pin the FlashQLA tag or commit, flash-linear-attention==0.5.0, and all transitive dependencies. FlashQLA v0.1.2 requires tilelang==0.1.9 and apache-tvm-ffi==0.1.9. Also pin the cudnn-frontend source clone. Record the tested versions in a constraints or lock file to keep benchmark results comparable across rebuilds.

🤖 Prompt for AI Agents
Verify each finding against current code. Fix only still-valid issues, skip the
rest with a brief reason, keep changes minimal, and validate.

In `@benchmark/linear_attention/Dockerfile` around lines 22 - 27, Pin the
benchmark environment in the Dockerfile around the FlashQLA installation: check
out a specific FlashQLA tag or commit, require flash-linear-attention==0.5.0,
and pin cudnn-frontend to a specific source revision. Add and apply a
constraints or lock file covering these packages and all transitive
dependencies, including tilelang==0.1.9 and apache-tvm-ffi==0.1.9, while
recording the tested versions.

Source: MCP tools

Comment thread benchmark/linear_attention/plot_results.py
Comment thread benchmark/linear_attention/plot_results.py Outdated
@Anerudhan Anerudhan added cat-feature Requests for new functionality, APIs, examples, or behavior improvements. mod-frontend cuDNN frontend APIs, operation graph construction, plans, and user-facing wrappers. orig-nv-eng Reported or requested by NVIDIA engineering. labels Aug 6, 2026
@Anerudhan Anerudhan added this to the Frontend 1.27.0 milestone Aug 6, 2026
@Anerudhan
Anerudhan merged commit 67c7634 into NVIDIA:develop Aug 6, 2026
1 check passed
Anerudhan pushed a commit that referenced this pull request Aug 6, 2026
@coderabbitai coderabbitai Bot mentioned this pull request Aug 6, 2026
Sign up for free to join this conversation on GitHub. Already have an account? Sign in to comment

Labels

cat-feature Requests for new functionality, APIs, examples, or behavior improvements. mod-frontend cuDNN frontend APIs, operation graph construction, plans, and user-facing wrappers. orig-nv-eng Reported or requested by NVIDIA engineering.

Projects

None yet

Development

Successfully merging this pull request may close these issues.

3 participants