FastVideo Kernel
CUDA kernels for FastVideo video generation.
Installation
Standard Installation (Local Development)
This will automatically detect your GPU architecture. If an NVIDIA Hopper (H100/sm_90a) GPU is detected, ThunderKittens kernels will be enabled. Otherwise, they will be skipped, and the package will use Triton fallbacks at runtime.
Before installation, set CUDA toolchain paths:
export CUDA_HOME=/usr/local/cuda
export CUDACXX=$CUDA_HOME/bin/nvcc
git submodule update --init --recursive
cd fastvideo-kernel
./build.sh
Rocm Build
If you are in a rocm environment without the compilation toolchaine of CUDA.
cd fastvideo-kernel
./build.sh --rocm
Optional: FA4 CuTe block-sparse backend (VSA-256 fastpath)
The VSA-256 fastpath (tile volume 256, on NVIDIA Blackwell / sm_100) routes to the
FlashAttention-4 CuTe-DSL block-sparse kernel exposed as flash_attn.cute. This is
an optional dependency: it is imported lazily, and video_sparse_attn
transparently falls back to the Triton backend when it is absent (so the package is
fully usable without it).
The symbols the fastpath needs (flash_attn.cute.block_sparsity.BlockSparseTensorsTorch,
flash_attn.cute.interface._flash_attn_fwd) are provided upstream by
Dao-AILab/flash-attention. Pin to
commit c19cd20e: the wrapper targets that revision's _flash_attn_fwd signature
(m_block_size / n_block_size); later upstream revisions reshaped it into a
tile_mn tuple and are not drop-in compatible.
pip install "nvidia-cutlass-dsl>=4.5.0" torchvision
pip install "git+https://github.com/Dao-AILab/flash-attention.git@c19cd20e#subdirectory=flash_attn/cute"
The CuTe kernel JIT-compiles on first use. Verified on Blackwell (sm_100) against
tests/test_vsa256_forward*.py.
Usage
Sliding Tile Attention (STA) & Video Sparse Attention (VSA)
For detailed usage, please check the Attention Documentation.
from fastvideo_kernel import sliding_tile_attention, video_sparse_attn, moba_attn_varlen
# Example: Sliding Tile Attention
out = sliding_tile_attention(q, k, v, window_sizes, text_len)
# Example: Video Sparse Attention (with Triton fallback)
out = video_sparse_attn(q, k, v, block_sizes, block_sizes, topk=5)
# Example: VMoBA
out = moba_attn_varlen(q, k, v, cu_seqlens_q, cu_seqlens_k, ...)
Benchmark
VSA (block-sparse) TFLOPs
After building/installing fastvideo-kernel, run:
cd fastvideo-kernel
python benchmarks/bench_vsa.py --batch_size 1 --num_heads 16 --head_dim 128 --q_seq_lens 49152 --topk 64
TurboDiffusion Kernels
This package also includes kernels from TurboDiffusion, including INT8 GEMM, Quantization, RMSNorm and LayerNorm.
Requirements
- Runtime:
- NVIDIA H100 (sm_90a) for C++ optimized kernels.
- Any CUDA GPU for Triton-based fallbacks.
- Build:
- CUDA Toolkit 12.3+
CUDA_HOMEmust be set (for example,/usr/local/cuda)CUDACXXmust be set (for example,$CUDA_HOME/bin/nvcc)- C++20 compatible compiler (GCC 10+, Clang 11+)
Acknowledgement
This package structure and build system are based on sgl-kernel from the SGLang project.
The implementation of turbodiffusion kernels is adapted from TurboDiffusion. If you use these kernels, please cite:
@article{zhang2025turbodiffusion,
title={TurboDiffusion: Accelerating Video Diffusion Models by 100-200 Times},
author={Zhang, Jintao and Zheng, Kaiwen and Jiang, Kai and Wang, Haoxu and Stoica, Ion and Gonzalez, Joseph E and Chen, Jianfei and Zhu, Jun},
journal={arXiv preprint arXiv:2512.16093},
year={2025}
}
Download files
Download the file for your platform. If you're not sure which to choose, learn more about installing packages.
Source Distribution
Built Distributions
Filter files by name, interpreter, ABI, and platform.
If you're not sure about the file name format, learn more about wheel file names.
Copy a direct link to the current filters
File details
Details for the file fastvideo_kernel-0.3.1.tar.gz.
File metadata
- Download URL: fastvideo_kernel-0.3.1.tar.gz
- Upload date:
- Size: 25.6 MB
- Tags: Source
- Uploaded using Trusted Publishing? Yes
- Uploaded via: twine/6.1.0 CPython/3.13.12
File hashes
| Algorithm | Hash digest | |
|---|---|---|
| SHA256 |
0cf25c6c7975c724e97ca5a43054758b41d1a64be4076f82ed11efca3dfaaf21
|
|
| MD5 |
1936962815e667a9a9396ba43adab3cf
|
|
| BLAKE2b-256 |
fec5c59e43c1d5d97c179cb168405b7a19e39dcfa8a876590fbf19b3fd2f413a
|
Provenance
The following attestation bundles were made for fastvideo_kernel-0.3.1.tar.gz:
Publisher:
publish-kernel.yml on hao-ai-lab/FastVideo
-
Statement:
-
Statement type:
https://in-toto.io/Statement/v1 -
Predicate type:
https://docs.pypi.org/attestations/publish/v1 -
Subject name:
fastvideo_kernel-0.3.1.tar.gz -
Subject digest:
0cf25c6c7975c724e97ca5a43054758b41d1a64be4076f82ed11efca3dfaaf21 - Sigstore transparency entry: 2025926883
- Sigstore integration time:
-
Permalink:
hao-ai-lab/FastVideo@31e26abec43bec9b25a68dd1f2618db4972fe293 -
Branch / Tag:
refs/heads/main - Owner: https://github.com/hao-ai-lab
-
Access:
public
-
Token Issuer:
https://token.actions.githubusercontent.com -
Runner Environment:
github-hosted -
Publication workflow:
publish-kernel.yml@31e26abec43bec9b25a68dd1f2618db4972fe293 -
Trigger Event:
push
-
Statement type:
File details
Details for the file fastvideo_kernel-0.3.1-cp312-cp312-manylinux_2_34_x86_64.manylinux_2_35_x86_64.whl.
File metadata
- Download URL: fastvideo_kernel-0.3.1-cp312-cp312-manylinux_2_34_x86_64.manylinux_2_35_x86_64.whl
- Upload date:
- Size: 21.3 MB
- Tags: CPython 3.12, manylinux: glibc 2.34+ x86-64, manylinux: glibc 2.35+ x86-64
- Uploaded using Trusted Publishing? Yes
- Uploaded via: twine/6.1.0 CPython/3.13.12
File hashes
| Algorithm | Hash digest | |
|---|---|---|
| SHA256 |
351d8c419a398bc63c7a9d3777b57ce806dca8a356292d563ce4f9d304060f83
|
|
| MD5 |
0530871b532ffe2e9435360eac8ffff5
|
|
| BLAKE2b-256 |
6dfbaee3afd273aa8b5e712451b495d7e64712ea8084fa9fbf3372ea96f20a25
|
Provenance
The following attestation bundles were made for fastvideo_kernel-0.3.1-cp312-cp312-manylinux_2_34_x86_64.manylinux_2_35_x86_64.whl:
Publisher:
publish-kernel.yml on hao-ai-lab/FastVideo
-
Statement:
-
Statement type:
https://in-toto.io/Statement/v1 -
Predicate type:
https://docs.pypi.org/attestations/publish/v1 -
Subject name:
fastvideo_kernel-0.3.1-cp312-cp312-manylinux_2_34_x86_64.manylinux_2_35_x86_64.whl -
Subject digest:
351d8c419a398bc63c7a9d3777b57ce806dca8a356292d563ce4f9d304060f83 - Sigstore transparency entry: 2025926976
- Sigstore integration time:
-
Permalink:
hao-ai-lab/FastVideo@31e26abec43bec9b25a68dd1f2618db4972fe293 -
Branch / Tag:
refs/heads/main - Owner: https://github.com/hao-ai-lab
-
Access:
public
-
Token Issuer:
https://token.actions.githubusercontent.com -
Runner Environment:
github-hosted -
Publication workflow:
publish-kernel.yml@31e26abec43bec9b25a68dd1f2618db4972fe293 -
Trigger Event:
push
-
Statement type:
File details
Details for the file fastvideo_kernel-0.3.1-cp312-cp312-manylinux_2_34_aarch64.manylinux_2_35_aarch64.whl.
File metadata
- Download URL: fastvideo_kernel-0.3.1-cp312-cp312-manylinux_2_34_aarch64.manylinux_2_35_aarch64.whl
- Upload date:
- Size: 5.0 MB
- Tags: CPython 3.12, manylinux: glibc 2.34+ ARM64, manylinux: glibc 2.35+ ARM64
- Uploaded using Trusted Publishing? Yes
- Uploaded via: twine/6.1.0 CPython/3.13.12
File hashes
| Algorithm | Hash digest | |
|---|---|---|
| SHA256 |
e6890ccb986d4a3480426768ef1c2d16e6312fce15ad9c4c32a3659eff7f0ba8
|
|
| MD5 |
1931855c0c6b6b536ce9fb62bd410a25
|
|
| BLAKE2b-256 |
83c7f8710eba11d46758c026f8e64a2339a85e209638b571940bdee374872a6c
|
Provenance
The following attestation bundles were made for fastvideo_kernel-0.3.1-cp312-cp312-manylinux_2_34_aarch64.manylinux_2_35_aarch64.whl:
Publisher:
publish-kernel.yml on hao-ai-lab/FastVideo
-
Statement:
-
Statement type:
https://in-toto.io/Statement/v1 -
Predicate type:
https://docs.pypi.org/attestations/publish/v1 -
Subject name:
fastvideo_kernel-0.3.1-cp312-cp312-manylinux_2_34_aarch64.manylinux_2_35_aarch64.whl -
Subject digest:
e6890ccb986d4a3480426768ef1c2d16e6312fce15ad9c4c32a3659eff7f0ba8 - Sigstore transparency entry: 2025927092
- Sigstore integration time:
-
Permalink:
hao-ai-lab/FastVideo@31e26abec43bec9b25a68dd1f2618db4972fe293 -
Branch / Tag:
refs/heads/main - Owner: https://github.com/hao-ai-lab
-
Access:
public
-
Token Issuer:
https://token.actions.githubusercontent.com -
Runner Environment:
github-hosted -
Publication workflow:
publish-kernel.yml@31e26abec43bec9b25a68dd1f2618db4972fe293 -
Trigger Event:
push
-
Statement type: