Skip to main content

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_HOME must be set (for example, /usr/local/cuda)
    • CUDACXX must 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

fastvideo_kernel-0.3.1.tar.gz (25.6 MB view details)

Uploaded Source

Built Distributions

If you're not sure about the file name format, learn more about wheel file names.

fastvideo_kernel-0.3.1-cp312-cp312-manylinux_2_34_x86_64.manylinux_2_35_x86_64.whl (21.3 MB view details)

Uploaded CPython 3.12manylinux: glibc 2.34+ x86-64manylinux: glibc 2.35+ x86-64

fastvideo_kernel-0.3.1-cp312-cp312-manylinux_2_34_aarch64.manylinux_2_35_aarch64.whl (5.0 MB view details)

Uploaded CPython 3.12manylinux: glibc 2.34+ ARM64manylinux: glibc 2.35+ ARM64

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

Hashes for fastvideo_kernel-0.3.1.tar.gz
Algorithm Hash digest
SHA256 0cf25c6c7975c724e97ca5a43054758b41d1a64be4076f82ed11efca3dfaaf21
MD5 1936962815e667a9a9396ba43adab3cf
BLAKE2b-256 fec5c59e43c1d5d97c179cb168405b7a19e39dcfa8a876590fbf19b3fd2f413a

See more details on using hashes here.

Provenance

The following attestation bundles were made for fastvideo_kernel-0.3.1.tar.gz:

Publisher: publish-kernel.yml on hao-ai-lab/FastVideo

Attestations: Values shown here reflect the state when the release was signed and may no longer be current.

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

File hashes

Hashes for fastvideo_kernel-0.3.1-cp312-cp312-manylinux_2_34_x86_64.manylinux_2_35_x86_64.whl
Algorithm Hash digest
SHA256 351d8c419a398bc63c7a9d3777b57ce806dca8a356292d563ce4f9d304060f83
MD5 0530871b532ffe2e9435360eac8ffff5
BLAKE2b-256 6dfbaee3afd273aa8b5e712451b495d7e64712ea8084fa9fbf3372ea96f20a25

See more details on using hashes here.

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

Attestations: Values shown here reflect the state when the release was signed and may no longer be current.

File details

Details for the file fastvideo_kernel-0.3.1-cp312-cp312-manylinux_2_34_aarch64.manylinux_2_35_aarch64.whl.

File metadata

File hashes

Hashes for fastvideo_kernel-0.3.1-cp312-cp312-manylinux_2_34_aarch64.manylinux_2_35_aarch64.whl
Algorithm Hash digest
SHA256 e6890ccb986d4a3480426768ef1c2d16e6312fce15ad9c4c32a3659eff7f0ba8
MD5 1931855c0c6b6b536ce9fb62bd410a25
BLAKE2b-256 83c7f8710eba11d46758c026f8e64a2339a85e209638b571940bdee374872a6c

See more details on using hashes here.

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

Attestations: Values shown here reflect the state when the release was signed and may no longer be current.

Supported by

AWS Cloud computing and Security Sponsor Datadog Monitoring Depot Continuous Integration Fastly CDN Google Download Analytics Pingdom Monitoring Sentry Error logging StatusPage Status page