Buckets:

hf-doc-build
/

doc-dev

Files

xet

hf-doc-build/doc-dev / kernels /pr_607 /en /llms-full.txt

HuggingFaceDocBuilder

about 1 month ago

download

raw

160 kB

	# Kernels

	## Docs

	- [Environment variables](https://huggingface.co/docs/kernels/pr_607/env.md)
	- [Quickstart](https://huggingface.co/docs/kernels/pr_607/basic-usage.md)
	- [Installation](https://huggingface.co/docs/kernels/pr_607/installation.md)
	- [Why kernels?](https://huggingface.co/docs/kernels/pr_607/why_kernels.md)
	- [Kernels CLI Reference](https://huggingface.co/docs/kernels/pr_607/cli.md)
	- [kernels benchmark](https://huggingface.co/docs/kernels/pr_607/cli-benchmark.md)
	- [Kernels](https://huggingface.co/docs/kernels/pr_607/index.md)
	- [FAQ](https://huggingface.co/docs/kernels/pr_607/faq.md)
	- [kernels lock](https://huggingface.co/docs/kernels/pr_607/cli-lock.md)
	- [CLI reference for kernel-builder](https://huggingface.co/docs/kernels/pr_607/builder-cli.md)
	- [Layers](https://huggingface.co/docs/kernels/pr_607/layers.md)
	- [Kernel requirements](https://huggingface.co/docs/kernels/pr_607/kernel-requirements.md)
	- [Locking kernel/layer versions](https://huggingface.co/docs/kernels/pr_607/locking.md)
	- [Integrating kernels](https://huggingface.co/docs/kernels/pr_607/integrating-kernels.md)
	- [kernels versions](https://huggingface.co/docs/kernels/pr_607/cli-versions.md)
	- [kernel-builder skills add](https://huggingface.co/docs/kernels/pr_607/cli-skills.md)
	- [Talks](https://huggingface.co/docs/kernels/pr_607/talks.md)
	- [Migrating from older versions](https://huggingface.co/docs/kernels/pr_607/migration.md)
	- [kernels download](https://huggingface.co/docs/kernels/pr_607/cli-download.md)
	- [Kernels API Reference](https://huggingface.co/docs/kernels/pr_607/api/kernels.md)
	- [Layers API Reference](https://huggingface.co/docs/kernels/pr_607/api/layers.md)
	- [IDE setup with direnv and the kernel devshell](https://huggingface.co/docs/kernels/pr_607/builder/ide-setup.md)
	- [Writing custom kernels with code agents](https://huggingface.co/docs/kernels/pr_607/builder/agents-guide.md)
	- [Writing Hub kernels with kernel-builder](https://huggingface.co/docs/kernels/pr_607/builder/writing-kernels.md)
	- [Using the kernel builder with Nix](https://huggingface.co/docs/kernels/pr_607/builder/build.md)
	- [Design overview](https://huggingface.co/docs/kernels/pr_607/builder/design-overview.md)
	- [Build variants](https://huggingface.co/docs/kernels/pr_607/builder/build-variants.md)
	- [Why Nix?](https://huggingface.co/docs/kernels/pr_607/builder/why-nix.md)
	- [Nix Builder design](https://huggingface.co/docs/kernels/pr_607/builder/design-nix-builder.md)
	- [Metal kernels 🤘](https://huggingface.co/docs/kernels/pr_607/builder/metal.md)
	- [Security](https://huggingface.co/docs/kernels/pr_607/builder/security.md)
	- [Local development of kernels](https://huggingface.co/docs/kernels/pr_607/builder/local-dev.md)

	### Environment variables
	https://huggingface.co/docs/kernels/pr_607/env.md

	# Environment variables

	## `KERNELS_CACHE`

	The directory to use as the local kernel cache. If not set, the cache
	of the `huggingface_hub` package is used.

	## `DISABLE_KERNEL_MAPPING`

	Disables kernel mappings for [`layers`](layers).

	### Quickstart
	https://huggingface.co/docs/kernels/pr_607/basic-usage.md

	# Quickstart

	## Loading Kernels

	Here is how you would use the [activation](https://huggingface.co/kernels-community/activation) kernels from the Hugging Face Hub:

	```python
	import torch
	from kernels import get_kernel

	# Download optimized kernels from the Hugging Face hub
	activation = get_kernel("kernels-community/activation", version=1)

	# Create a random tensor
	x = torch.randn((10, 10), dtype=torch.float16, device="cuda")

	# Run the kernel
	y = torch.empty_like(x)
	activation.gelu_fast(y, x)

	print(y)
	```

	This fetches version `1` of the kernel `kernels-community/activation`.
	Kernels are versioned using a major version number. Using `version=1` will
	get the latest kernel build from the `v1` branch.

	Kernels within a version branch must never break the API or remove builds
	for older PyTorch versions. This ensures that your code will continue to work.

	Hub kernels must be loaded with either a `version` or an explicit `revision`.

	## Checking Kernel Availability

	You can check if a particular version of a kernel supports the environment
	that the program is running on:

	```python
	from kernels import has_kernel

	# Check if kernel is available for current environment
	is_available = has_kernel("kernels-community/activation", version=1)
	print(f"Kernel available: {is_available}")
	```

	When no compatible kernel is found, [has_kernel()](/docs/kernels/pr_607/en/api/kernels#kernels.has_kernel) does not say why.
	[get_kernel_variants()](/docs/kernels/pr_607/en/api/kernels#kernels.get_kernel_variants) returns the full resolution trace instead: one
	decision per build variant in the repository, with compatible variants listed
	first. Each decision is a `VariantAccepted` or a `VariantRejected`, and rejected
	variants carry a human-readable `reason`:

	```python
	from kernels import get_kernel_variants, VariantAccepted

	for decision in get_kernel_variants("kernels-community/activation", version=1):
	name = decision.variant.variant_str
	if isinstance(decision, VariantAccepted):
	print(f"{name}: compatible")
	else:
	print(f"{name}: rejected ({decision.reason})")
	```

	## Inspecting Loaded Kernels

	[get_loaded_kernels()](/docs/kernels/pr_607/en/api/kernels#kernels.get_loaded_kernels) returns a snapshot of every kernel that has been loaded
	into the current process. Each entry is a [LoadedKernel](/docs/kernels/pr_607/en/api/kernels#kernels.LoadedKernel) namedtuple with the
	imported `module`, the `package_name`, and `repo_infos` (repo id, resolved
	revision, and the backend argument that was passed).

	```python
	from kernels import get_kernel, get_loaded_kernels

	get_kernel("kernels-community/activation", version=1)

	for loaded in get_loaded_kernels():
	print(loaded.package_name, loaded.repo_infos)
	```

	`repo_infos` is populated only for kernels loaded with [get_kernel()](/docs/kernels/pr_607/en/api/kernels#kernels.get_kernel). Kernels
	loaded from a local path ([get_local_kernel()](/docs/kernels/pr_607/en/api/kernels#kernels.get_local_kernel)) or via a lockfile
	([get_locked_kernel()](/docs/kernels/pr_607/en/api/kernels#kernels.get_locked_kernel), [load_kernel()](/docs/kernels/pr_607/en/api/kernels#kernels.load_kernel)) have `repo_infos=None`.

	Browse through different kernels compatible with `kernels` from [here](https://huggingface.co/kernels).

	A kernel can provide layers in addition to kernel functions. Refer to [Layers](./layers) to know more.

	### Installation
	https://huggingface.co/docs/kernels/pr_607/installation.md

	# Installation

	> [!WARNING]
	> `kernels` has not reached `1.0` yet. Until then, minor releases may contain
	> breaking changes. If you depend on `kernels` in a library or application, we
	> strongly recommend pinning a version range rather than an unbounded
	> dependency. For example, in `pyproject.toml`:
	>
	> ```toml
	> dependencies = [
	> "kernels>=0.15,<0.16",
	> ]
	> ```
	>
	> or equivalently `kernels~=0.15` (compatible release). This protects your
	> project from unexpected breakage when a new `kernels` version is released.

	Install the `kernels` package with `pip` (requires `torch>=2.5` and CUDA):

	```bash
	pip install kernels
	```

	or with `uv`

	```bash
	uv pip install kernels
	```

	or if you want the latest version from the `main` branch:

	```bash
	pip install "kernels[benchmark] @ git+https://github.com/huggingface/kernels#subdirectory=kernels"
	```

	> [!IMPORTANT]
	> On Windows, we recommend using the Linux version of Torch through
	> [WSL 2](https://learn.microsoft.com/en-us/windows/wsl/install), since
	> many more kernels support Linux. If you want to use GPU acceleration,
	> check out the [CUDA on WSL](https://docs.nvidia.com/cuda/wsl-user-guide/index.html#getting-started-with-cuda-on-wsl-2)
	> and [PyTorch with DirectML on WSL 2](https://learn.microsoft.com/en-us/windows/ai/directml/pytorch-wsl)
	> guides.

	> [!IMPORTANT]
	> We strongly recommend not using a free-threaded Python build yet.
	> These builds are not only experimental, but do not support the stable ABI
	> on Python versions before 3.15. Kernels are compiled with the stable ABI
	> to support a wide range of Python versions.

	### Why kernels?
	https://huggingface.co/docs/kernels/pr_607/why_kernels.md

	# Why kernels?

	Our goal with the `kernels` package is manifold:

	* Establish a standard way of structuring and building kernels. The
	`builder` component takes kernel source in a pre-defined layout, with
	a declarative build configuration, and produces compiled kernels for
	a wide matrix of compute backends (e.g., CUDA, ROCM, and XPU), operating
	systems, and architectures. The builder also enforces the reproducibility of
	the build environment and build steps.
	* Provide a standard way of distributing and loading kernels. Kernels
	are distributed through the Hugging Face Hub and can be loaded through
	the kernels Python package. `kernels` fetches the right kernel build for
	the system that it runs, avoiding long local, sometimes hours-long,
	builds. It also supports loading of multiple versions of the same
	kernel, effectively eliminating "dependency hell".
	* Provide kernel builds across all supported PyTorch versions,
	accelerators, and capabilities. This is particularly important
	because local kernel builds can become unusable once the base
	machine learning framework (e.g., PyTorch) is updated to a recent version.

	Additionally, there are several advantages to hosting the pre-built
	kernels on the Hugging Face Hub platform:

	* Trends -- kinds of models using a kernel more than others,
	[applications](https://huggingface.co/spaces) using a particular kernel, etc.
	* Users immediately know if their hardware supports a kernel
	without having to run any installation locally.
	* General features of the Hub platform:
	* Seamless versioning
	* Fast download and upload powered by [XET](https://huggingface.co/docs/hub/xet/index)
	* Visibility into the download stats

	### Kernels CLI Reference
	https://huggingface.co/docs/kernels/pr_607/cli.md

	# Kernels CLI Reference

	The `kernels` CLI provides commands for managing compute kernels.

	## Commands

	\| Command \| Description \|
	\| ------------------------------------------------------- \| -------------------------------------------------------- \|
	\| [benchmark](cli-benchmark) \| Run benchmark results for a kernel \|
	\| [check](cli-check) \| Check a kernel for compliance \|
	\| [versions](cli-versions) \| Show kernel versions \|
	\| [lock](cli-lock) \| Lock kernel revisions \|

	## Quick Start

	For building and writing kernels, please refer [building kernels](./builder/build) and
	[writing kernels](./builder/writing-kernels).

	### Use kernels in your project

	#### Directly from the Hub

	```python
	import torch

	from kernels import get_kernel

	# Download optimized kernels from the Hugging Face hub
	my_kernel = get_kernel("my-username/my-kernel", version=1)

	# Random tensor
	x = torch.randn((10, 10), dtype=torch.float16, device="cuda")

	# Run the kernel
	y = torch.empty_like(x)
	my_kernel.my_kernel_function(y, x)

	print(y)
	```

	or

	#### Locked and downloaded

	Add to `pyproject.toml`:

	```toml
	[tool.kernels.dependencies]
	"my-username/my-kernel" = "1"
	```

	Then lock and download:

	```bash
	kernels lock .
	kernels download .
	```

	### See help

	```bash
	kernels --help
	```

	### kernels benchmark
	https://huggingface.co/docs/kernels/pr_607/cli-benchmark.md

	# kernels benchmark

	Use `kernels benchmark` to run benchmark scripts shipped with a kernel repository.

	The command:

	- Downloads the kernel repo at a specific branch or version
	- Runs all `benchmarks/benchmark*.py` scripts
	- Times each `benchmark_*` workload and prints a results table
	- Optionally saves results as JSON

	## Installation

	`kernels benchmark` requires extra dependencies:

	```bash
	uv pip install 'kernels[benchmark]' # or pip install 'kernels[benchmark]'
	```

	## Example

	```bash
	kernels benchmark kernels-community/activation --version 1
	```

	Example output:

	```text
	Downloading kernels-community/activation@v1...
	Running benchmark.py...

	GPU Apple M3 Max (30 cores)
	CPU Apple M3 Max
	OS Darwin 25.2.0
	PyTorch 2.10.0

	Running SiluWorkloads on mps

	┌───────────────┬────────────┬─────┬───────────┬────────────┬───────────┬───────────┬───────────┬───────────┬────────────┬───────────┬─────────┐
	│ Benchmark │ Workload │ N │ Speedup │ Mean(ms) │ Std(ms) │ Min(ms) │ Max(ms) │ IQR(ms) │ Outliers │ Ref(ms) │ Match │
	├───────────────┼────────────┼─────┼───────────┼────────────┼───────────┼───────────┼───────────┼───────────┼────────────┼───────────┼─────────┤
	│ SiluWorkloads │ large │ 100 │ 1.72x │ 6.5153 │ 0.4343 │ 6.2883 │ 8.4699 │ 0.1701 │ 8 │ 11.2048 │ ✓ │
	│ SiluWorkloads │ medium │ 100 │ 2.48x │ 1.1813 │ 0.3976 │ 1.04 │ 4.2146 │ 0.0698 │ 5 │ 2.9332 │ ✓ │
	│ SiluWorkloads │ small │ 100 │ 1.96x │ 0.4909 │ 0.2175 │ 0.4407 │ 2.6438 │ 0.0085 │ 16 │ 0.9622 │ ✓ │
	└───────────────┴────────────┴─────┴───────────┴────────────┴───────────┴───────────┴───────────┴───────────┴────────────┴───────────┴─────────┘

	large: 1.72x faster (95% CI: 6.4302-6.6004ms vs ref 11.2048ms) ✓ significant
	medium: 2.48x faster (95% CI: 1.1034-1.2592ms vs ref 2.9332ms) ✓ significant
	small: 1.96x faster (95% CI: 0.4483-0.5335ms vs ref 0.9622ms) ✓ significant

	Kernel: 2385e44 Benchmark: 5b53516
	```

	## Usage

	You must specify which revision to benchmark, either via flags or with `@...` in the repo id:

	```bash
	kernels benchmark <repo_id> --version <N>
	kernels benchmark <repo_id> --branch <name>
	kernels benchmark <repo_id>@v<N>
	kernels benchmark <repo_id>@<branch>
	```

	## Examples

	Benchmark a tagged kernel version:

	```bash
	kernels benchmark kernels-community/activation --version 1
	```

	Equivalent shorthand:

	```bash
	kernels benchmark kernels-community/activation@v1
	```

	Benchmark a branch:

	```bash
	kernels benchmark kernels-community/activation --branch main
	```

	Tune warmup and iteration count:

	```bash
	kernels benchmark kernels-community/activation@v1 --warmup 20 --iterations 200
	```

	Save results to a file (JSON):

	```bash
	kernels benchmark kernels-community/activation@v1 --output results.json
	```

	Benchmark a local kernel checkout (must contain `benchmarks/`):

	```bash
	kernels benchmark ./my_kernel
	```

	## Output

	- By default, a table is printed (timings in ms).
	- `--output <file>.json` writes a JSON payload to disk.

	## Writing Benchmark Scripts

	Benchmark scripts must live under `benchmarks/` in the kernel repository and match `benchmark*.py`.
	Each script should define one or more subclasses of `kernels.benchmark.Benchmark`.

	Minimal example (`benchmarks/benchmark_activation.py`):

	```python
	import torch

	from kernels.benchmark import Benchmark

	class ActivationBenchmark(Benchmark):
	seed = 0

	def setup(self):
	self.x = torch.randn(128, 1024, device=self.device, dtype=torch.float16)
	self.out = torch.empty(128, 512, device=self.device, dtype=torch.float16)

	def benchmark_silu_and_mul(self):
	self.kernel.silu_and_mul(self.out, self.x)

	def verify_silu_and_mul(self):
	# Return reference tensor; runner compares with self.out
	return torch.nn.functional.silu(self.x[..., :512]) * self.x[..., 512:]
	```

	The runner will:

	- Call `setup()` once per workload (or `setup_<workload>()` if present)
	- Warm up (`--warmup`)
	- Time `benchmark_<workload>()` for `--iterations`
	- If `verify_<workload>()` exists, check that outputs match (`torch.allclose(..., atol=1e-2)`) and show a speedup vs the reference computation

	## Troubleshooting

	- If the repo does not contain a `benchmarks/` directory (or no `benchmark*.py` files), the command exits with an error.
	- If a benchmark script defines no `Benchmark` subclasses, the command exits with an error.
	- If `verify_<workload>()` exists and the outputs do not match, the command exits with an error.

	### Kernels
	https://huggingface.co/docs/kernels/pr_607/index.md

	# Kernels

	The Kernel Hub allows Python libraries and applications to load compute
	kernels directly from the [Hub](https://huggingface.co/). Kernels are a first-class
	repository type on the Hub, with dedicated pages that surface supported
	hardware and versions. To support dynamic loading, Hub kernels differ from
	traditional Python kernel packages in that they are made to be:

	- Portable: a kernel can be loaded from paths outside `PYTHONPATH`.
	- Unique: multiple versions of the same kernel can be loaded in the
	same Python process.
	- Compatible: `kernels` must support all recent versions of Python and
	the different PyTorch build configurations (various CUDA versions
	and C++ ABIs). Furthermore, older C library versions must be supported.

	Browse available kernels at [huggingface.co/kernels](https://huggingface.co/kernels).

	The Kernels project is divided into two parts:

	- Builder: [`kernel-builder`](builder-cli) provides utilities to build, package, and distribute compute kernels in a way that is compatible with the Hugging Face Hub and `kernels`.
	- `kernels`: The [`kernels`](basic-usage) is a Python package that lets
	users load compatible compute kernels from the Hub. Refer to the [quickstart](basic-usage) to know more.

	If you're looking for a more involved "Why kernels?" answer, refer to
	[this page](./why_kernels).

	The [talks page](./talks) page has links to talks on the
	Kernels project.

	### FAQ
	https://huggingface.co/docs/kernels/pr_607/faq.md

	# FAQ

	## Kernel layers

	### Why is the kernelization step needed as a separate step?

	In earlier versions of `kernels`, a layer's `forward` method was replaced
	by [use_kernel_forward_from_hub()](/docs/kernels/pr_607/en/api/layers#kernels.use_kernel_forward_from_hub) and [replace_kernel_forward_from_hub()](/docs/kernels/pr_607/en/api/layers#kernels.replace_kernel_forward_from_hub).
	The new `forward` would dispatch to a kernel based on the device type,
	whether a model was training, etc. However, this approach was
	fundamentally incompatible with `torch.compile` since it relied
	on data-dependent branching.

	To avoid branching, we have to make dispatch decisions ahead of time,
	which is what the [kernelize()](/docs/kernels/pr_607/en/api/layers#kernels.kernelize) function does.

	### Why does kernelization only replace `forward` methods?

	There are some other possible approaches. The first is to completely
	replace existing layers by kernel layers. However, since this would
	permit free-form layer classes, it would be much harder to validate
	that layers are fully compatible with the layers that they are
	replacing. For instance, they could have completely different member
	variables. Besides that, we would also need to hold on to the original
	layers, in case we need to revert to the base layers when the model
	is [kernelize()](/docs/kernels/pr_607/en/api/layers#kernels.kernelize)d again with different options.

	A second approach would be to make an auxiliary layer that wraps the
	original layer and the kernel layer and dispatches to the kernel layer.
	This wouldn't have the issues of the first approach, because kernel layers
	could be similarly strict as they are now, and we would still have access
	to the original layers when [kernelize()](/docs/kernels/pr_607/en/api/layers#kernels.kernelize)-ing the model again. However,
	this would change the graph structure of the model and would break use
	cases where programs access the model internals (e.g.
	`model.layers[0].attention.query_weight`) or rely on the graph structure
	in other ways.

	The approach of `forward`-replacement is the least invasive, because
	it preserves the original model graph. It is also reversible, since
	even though the `forward` of a layer _instance_ might be replaced,
	the corresponding class still has the original `forward`.

	## Misc

	### How can I disable kernel reporting in the user-agent?

	By default, we collect telemetry when a call to [get_kernel()](/docs/kernels/pr_607/en/api/kernels#kernels.get_kernel) is made.
	This only includes the `kernels` version, `torch` version, and the build
	information for the kernel being requested.

	You can disable this by setting `export DISABLE_TELEMETRY=yes`.

	### kernels lock
	https://huggingface.co/docs/kernels/pr_607/cli-lock.md

	# kernels lock

	Use `kernels lock` to generate a `kernels.lock` file that pins kernel dependencies to specific revisions.

	## Usage

	```bash
	kernels lock <project_dir>
	```

	## What It Does

	- Reads kernel dependencies from `pyproject.toml` under `[tool.kernels.dependencies]`
	- Resolves each kernel to its current revision SHA
	- Writes a `kernels.lock` file with pinned versions and variant information

	## Examples

	Lock kernels in the current project:

	```bash
	kernels lock .
	```

	Lock kernels in a specific project:

	```bash
	kernels lock /path/to/my-project
	```

	## pyproject.toml Format

	Add your kernel dependencies to `pyproject.toml`:

	```toml
	[tool.kernels.dependencies]
	"kernels-community/activation" = 1
	```

	The version can be:

	- A version number (e.g., `1`, `2`)

	## kernels.lock Format

	The generated lock file contains:

	```json
	[
	{
	"repo_id": "kernels-community/activation",
	"sha": "ece277f908b9453112722d584fee4b5696f21c49",
	"variants": {
	"torch210-cu128-x86_64-windows": {
	"hash": "sha256-cbf085e1d297d990d9cb074fb5079ff48e9682c729f53a0899a36b5164a6fb45",
	"hash_type": "git_lfs_concat"
	},
	// ...
	"torch29-metal-aarch64-darwin": {
	"hash": "sha256-9f665b54a53246a7d3627422f8a0d41d7956dc5409043dbd14c4ec0327aea310",
	"hash_type": "git_lfs_concat"
	}
	}
	}
	]
	```

	## Workflow

	1. Add dependencies to `pyproject.toml`
	2. Run `kernels lock .` to generate the lock file
	3. Commit both `pyproject.toml` and `kernels.lock`
	4. Use `kernels download .` to install locked kernels

	## See Also

	- [kernels download](cli-download) - Download locked kernels
	- [kernels versions](cli-versions) - View available kernel versions

	### CLI reference for kernel-builder
	https://huggingface.co/docs/kernels/pr_607/builder-cli.md

	# CLI reference for kernel-builder

	This document contains the help content for the `kernel-builder` command-line program.

	Command Overview:

	* [`kernel-builder`↴](#kernel-builder)
	* [`kernel-builder completions`↴](#kernel-builder-completions)
	* [`kernel-builder init`↴](#kernel-builder-init)
	* [`kernel-builder build`↴](#kernel-builder-build)
	* [`kernel-builder build-and-copy`↴](#kernel-builder-build-and-copy)
	* [`kernel-builder build-and-upload`↴](#kernel-builder-build-and-upload)
	* [`kernel-builder upload`↴](#kernel-builder-upload)
	* [`kernel-builder check-config`↴](#kernel-builder-check-config)
	* [`kernel-builder check-abi`↴](#kernel-builder-check-abi)
	* [`kernel-builder check-builds`↴](#kernel-builder-check-builds)
	* [`kernel-builder create-pyproject`↴](#kernel-builder-create-pyproject)
	* [`kernel-builder devshell`↴](#kernel-builder-devshell)
	* [`kernel-builder list-variants`↴](#kernel-builder-list-variants)
	* [`kernel-builder testshell`↴](#kernel-builder-testshell)
	* [`kernel-builder update-build`↴](#kernel-builder-update-build)
	* [`kernel-builder skills`↴](#kernel-builder-skills)
	* [`kernel-builder skills add`↴](#kernel-builder-skills-add)
	* [`kernel-builder clean-pyproject`↴](#kernel-builder-clean-pyproject)

	## `kernel-builder`

	Build Hugging Face Hub kernels

	Usage: `kernel-builder <COMMAND>`

	###### Subcommands:

	* `completions` — Generate shell completions
	* `init` — Initialize a new kernel project from template
	* `build` — Build the kernel locally (alias for build-and-copy)
	* `build-and-copy` — Build the kernel and copy artifacts locally
	* `build-and-upload` — Build the kernel and upload to Hugging Face Hub
	* `upload` — Upload kernel build artifacts to the Hugging Face Hub
	* `check-config` — Validate the build.toml file
	* `check-abi` — Check the ABI compatibility of a kernel extension
	* `check-builds` — Validate kernel builds
	* `create-pyproject` — Generate CMake files for a kernel extension build
	* `devshell` — Spawn a kernel development shell
	* `list-variants` — List build variants
	* `testshell` — Spawn a kernel test shell
	* `update-build` — Update a `build.toml` to the current format
	* `skills` — Install skills for AI coding assistants (Claude, Codex, OpenCode)
	* `clean-pyproject` — Clean generated artifacts

	## `kernel-builder completions`

	Generate shell completions

	Usage: `kernel-builder completions <SHELL>`

	###### Arguments:

	* `<SHELL>`

	Possible values: `bash`, `elvish`, `fish`, `powershell`, `zsh`

	## `kernel-builder init`

	Initialize a new kernel project from template

	Usage: `kernel-builder init [OPTIONS] [PATH]`

	###### Arguments:

	* `<PATH>` — Directory to initialize (defaults to current directory)

	###### Options:

	* `--license <LICENSE>` — The kernel's license

	Default value: `Apache-2.0`
	* `--name <OWNER/REPO>` — Name of the kernel repo (e.g. `drbh/my-kernel`)
	* `--backends <BACKENDS>` — Backends to enable (`all`, `cpu`, `cuda`, `metal`, `neuron`, `rocm`, `xpu`)

	* `--overwrite` — Overwrite existing scaffold files (preserves other files)

	## `kernel-builder build`

	Build the kernel locally (alias for build-and-copy)

	Usage: `kernel-builder build [OPTIONS] [KERNEL_DIR]`

	###### Arguments:

	* `<KERNEL_DIR>` — Directory of the kernel project (defaults to current directory)

	###### Options:

	* `--variant <VARIANT>` — Build a specific variant
	* `--max-jobs <MAX_JOBS>` — Maximum number of parallel Nix build jobs
	* `--cores <CORES>` — Number of CPU cores to use for each build job
	* `-L`, `--print-build-logs` — Print full build logs on standard error

	## `kernel-builder build-and-copy`

	Build the kernel and copy artifacts locally

	Usage: `kernel-builder build-and-copy [OPTIONS] [KERNEL_DIR]`

	###### Arguments:

	* `<KERNEL_DIR>` — Directory of the kernel project (defaults to current directory)

	###### Options:

	* `--max-jobs <MAX_JOBS>` — Maximum number of parallel Nix build jobs
	* `--cores <CORES>` — Number of CPU cores to use for each build job
	* `-L`, `--print-build-logs` — Print full build logs on standard error

	## `kernel-builder build-and-upload`

	Build the kernel and upload to Hugging Face Hub

	Usage: `kernel-builder build-and-upload [OPTIONS] [KERNEL_DIR]`

	###### Arguments:

	* `<KERNEL_DIR>` — Directory of the kernel project (defaults to current directory)

	###### Options:

	* `--variant <VARIANT>` — Build a specific variant
	* `--max-jobs <MAX_JOBS>` — Maximum number of parallel Nix build jobs
	* `--cores <CORES>` — Number of CPU cores to use for each build job
	* `-L`, `--print-build-logs` — Print full build logs on standard error
	* `--repo-id <REPO_ID>` — Repository ID on the Hugging Face Hub (e.g. `user/my-kernel`)
	* `--branch <BRANCH>` — Upload to a specific branch (defaults to `v{version}` from metadata)
	* `--private` — Create the repository as private
	* `--repo-type <REPO_TYPE>` — Repository type on Hugging Face Hub (`kernel` by default, or `model` for legacy repos)

	Default value: `kernel`

	Possible values: `model`, `kernel`

	* `-q`, `--quiet` — Suppress progress output

	## `kernel-builder upload`

	Upload kernel build artifacts to the Hugging Face Hub

	Usage: `kernel-builder upload [OPTIONS] [KERNEL_DIR]`

	###### Arguments:

	* `<KERNEL_DIR>` — Directory of the kernel build (defaults to current directory)

	###### Options:

	* `--repo-id <REPO_ID>` — Repository ID on the Hugging Face Hub (e.g. `user/my-kernel`). Defaults to `general.hub.repo-id` from `build.toml`
	* `--branch <BRANCH>` — Upload to a specific branch (defaults to `v{version}` from metadata)
	* `--private` — Create the repository as private
	* `--repo-type <REPO_TYPE>` — Repository type on Hugging Face Hub (`kernel` by default, or `model` for legacy repos)

	Default value: `kernel`

	Possible values: `model`, `kernel`

	* `-q`, `--quiet` — Suppress progress output

	## `kernel-builder check-config`

	Validate the build.toml file

	Usage: `kernel-builder check-config [KERNEL_DIR]`

	###### Arguments:

	* `<KERNEL_DIR>`

	## `kernel-builder check-abi`

	Check the ABI compatibility of a kernel extension

	Usage: `kernel-builder check-abi [OPTIONS] [KERNEL_DIR]`

	###### Arguments:

	* `<KERNEL_DIR>` — Directory with kernels

	###### Options:

	* `-m`, `--manylinux <VERSION>` — Manylinux version

	Default value: `manylinux_2_28`
	* `--macos <VERSION>` — macOS version

	Default value: `15.0`
	* `-p`, `--python-abi <VERSION>` — Python ABI version

	Default value: `3.9`
	* `--torch-stable-abi <VERSION>` — Torch stable ABI version

	## `kernel-builder check-builds`

	Validate kernel builds

	Usage: `kernel-builder check-builds [KERNEL_DIR]`

	###### Arguments:

	* `<KERNEL_DIR>`

	## `kernel-builder create-pyproject`

	Generate CMake files for a kernel extension build

	Usage: `kernel-builder create-pyproject [OPTIONS] [KERNEL_DIR] [TARGET_DIR]`

	###### Arguments:

	* `<KERNEL_DIR>`
	* `<TARGET_DIR>` — The directory to write the generated files to (directory of `BUILD_TOML` when absent)

	###### Options:

	* `-f`, `--force` — Force-overwrite existing files
	* `--unique-id <UNIQUE_ID>` — This is an optional unique identifier that is suffixed to the kernel name to avoid name collisions. (e.g. Git SHA)

	## `kernel-builder devshell`

	Spawn a kernel development shell

	Usage: `kernel-builder devshell [OPTIONS] [KERNEL_DIR]`

	###### Arguments:

	* `<KERNEL_DIR>`

	###### Options:

	* `--variant <VARIANT>` — Use a specific variant
	* `--max-jobs <MAX_JOBS>` — Maximum number of parallel Nix build jobs
	* `--cores <CORES>` — Number of CPU cores to use for each build job
	* `-L`, `--print-build-logs` — Print full build logs on standard error

	## `kernel-builder list-variants`

	List build variants

	Usage: `kernel-builder list-variants [OPTIONS] [KERNEL_DIR]`

	###### Arguments:

	* `<KERNEL_DIR>`

	###### Options:

	* `--arch` — Only list variants for the current architecture

	## `kernel-builder testshell`

	Spawn a kernel test shell

	Usage: `kernel-builder testshell [OPTIONS] [KERNEL_DIR]`

	###### Arguments:

	* `<KERNEL_DIR>`

	###### Options:

	* `--variant <VARIANT>` — Use a specific variant
	* `--max-jobs <MAX_JOBS>` — Maximum number of parallel Nix build jobs
	* `--cores <CORES>` — Number of CPU cores to use for each build job
	* `-L`, `--print-build-logs` — Print full build logs on standard error

	## `kernel-builder update-build`

	Update a `build.toml` to the current format

	Usage: `kernel-builder update-build [KERNEL_DIR]`

	###### Arguments:

	* `<KERNEL_DIR>`

	## `kernel-builder skills`

	Install skills for AI coding assistants (Claude, Codex, OpenCode)

	Usage: `kernel-builder skills <COMMAND>`

	###### Subcommands:

	* `add` — Install a kernels skill for an AI assistant

	## `kernel-builder skills add`

	Install a kernels skill for an AI assistant

	Usage: `kernel-builder skills add [OPTIONS]`

	###### Options:

	* `--skill <SKILL>` — Skill to install

	Default value: `cuda-kernels`

	Possible values: `cuda-kernels`, `rocm-kernels`, `xpu-kernels`

	* `--claude` — Install for Claude
	* `--codex` — Install for Codex
	* `--opencode` — Install for OpenCode
	* `-g`, `--global` — Install globally (user-level) instead of in the current project directory
	* `--dest <DEST>` — Install into a custom destination (path to skills directory)
	* `--force` — Overwrite existing skills in the destination

	## `kernel-builder clean-pyproject`

	Clean generated artifacts

	Usage: `kernel-builder clean-pyproject [OPTIONS] [KERNEL_DIR] [TARGET_DIR]`

	###### Arguments:

	* `<KERNEL_DIR>`
	* `<TARGET_DIR>` — The directory to clean from (directory of `BUILD_TOML` when absent)

	###### Options:

	* `-d`, `--dry-run` — Show what would be deleted without actually deleting
	* `-f`, `--force` — Force deletion without confirmation
	* `--ops-id <OPS_ID>` — This is an optional unique identifier that is suffixed to the kernel name to avoid name collisions. (e.g. Git SHA)

	This document was generated automatically by
	clap-markdown.

	### Layers
	https://huggingface.co/docs/kernels/pr_607/layers.md

	# Layers

	A kernel can provide layers in addition to kernel functions. A layer from
	the Hub can replace the `forward` method of an existing layer for a certain
	device type. This makes it possible to provide more performant kernels for
	existing layers.

	See [Kernel requirements](kernel-requirements) for more information on the
	requirements of Hub layers.

	## Making a layer extensible with kernels from the hub

	### Using a decorator

	A layer can be made extensible with the [use_kernel_forward_from_hub()](/docs/kernels/pr_607/en/api/layers#kernels.use_kernel_forward_from_hub)
	decorator. For example:

	```python
	@use_kernel_forward_from_hub("SiluAndMul")
	class SiluAndMul(nn.Module):
	def forward(self, input: torch.Tensor) -> torch.Tensor:
	d = input.shape[-1] // 2
	return F.silu(input[..., :d]) * input[..., d:]
	```

	The decorator does not change the behavior of the class -- it annotates
	the class with the given name (here `SiluAndMul`). The [kernelize()](/docs/kernels/pr_607/en/api/layers#kernels.kernelize) function
	described below uses this name to look up kernels for the layer.

	### External layers

	An existing layer that does not (yet) have the [use_kernel_forward_from_hub()](/docs/kernels/pr_607/en/api/layers#kernels.use_kernel_forward_from_hub)
	decorator can be made extensible using the [replace_kernel_forward_from_hub()](/docs/kernels/pr_607/en/api/layers#kernels.replace_kernel_forward_from_hub)
	function:

	```python
	from somelibrary import SiluAndMul

	replace_kernel_forward_from_hub(SiluAndMul, "SiluAndMul")
	```

	Warning: we strongly recommend using layers with a decorator, since
	it signifies that the maintainer intends to keep the `forward` signature
	compatible with layers from the hub.

	### Using a function as a layer

	Sometimes it can be useful to make a function extensible, for example
	because the function cannot be replaced by a layer. In such cases, you
	can annotate the function with the [use_kernel_func_from_hub()](/docs/kernels/pr_607/en/api/layers#kernels.use_kernel_func_from_hub) decorator:

	```python
	@use_kernel_func_from_hub("silu_and_mul")
	def silu_and_mul(x: torch.Tensor) -> torch.Tensor:
	d = x.shape[-1] // 2
	return F.silu(x[..., :d]) * x[..., d:]
	```

	This will replace the function by an instantiated `torch.nn.Module`
	(singleton) that calls the function itself in its forward method.

	Note: for kernelization to see the function, it must be a member of
	another `torch.nn.Module` that is part of the model. For example:

	```python
	class FeedForward(nn.Module):
	def __init__(self, in_features: int, out_features: int):
	self.linear = nn.Linear(in_features, out_features)
	# Note: silu_and_mul is a Torch module.
	self.silu_and_mul = silu_and_mul

	def forward(self, x: torch.Tensor) -> torch.Tensor:
	return self.silu_and_mul(self.linear(x))
	```

	## Kernelizing a model

	A model will not use Hub kernels by default, even if it contains extensible
	layers. To enable the use of Hub kernels in the model, it needs to be
	'kernelized' using the [kernelize()](/docs/kernels/pr_607/en/api/layers#kernels.kernelize) function. This function traverses the
	model graph and replaces the `forward` methods of extensible layers for which
	Hub kernels are registered. [kernelize()](/docs/kernels/pr_607/en/api/layers#kernels.kernelize) can be used as follows:

	```python
	model = MyModel(...)
	model = kernelize(model, mode=Mode.INFERENCE)
	```

	The [kernelize()](/docs/kernels/pr_607/en/api/layers#kernels.kernelize) function modifies the model in-place, the model itself is
	returned as a convenience. The `mode` specifies that the model will be used
	in inference. Similarly, you can ask [kernelize()](/docs/kernels/pr_607/en/api/layers#kernels.kernelize) to prepare the model for
	training:

	```python
	model = MyModel(...)
	model = kernelize(model, mode=Mode.TRAINING)
	```

	A model that is kernelized for training can also be used for inference, but
	not the other way around. If you want to change the mode of the kernelized
	model, you can just run [kernelize()](/docs/kernels/pr_607/en/api/layers#kernels.kernelize) on the model again with the new mode.

	If you want to compile a model with `torch.compile`, this should be indicated
	in the mode as well. You can do this by combining `Mode.INFERENCE` or
	`Mode.TRAINING` with `Mode.TORCH_COMPILE` using the set union (`\|`) operator:

	```python
	model = MyModel(...)

	# Inference
	model = kernelize(model, mode=Mode.INFERENCE \| Mode.TORCH_COMPILE)

	# Training
	model = kernelize(model, mode=Mode.TRAINING \| Mode.TORCH_COMPILE)
	```

	### Kernel device

	Kernels can be registered per device type. For instance, separate `cuda` and
	`metal` kernels could be registered for the name `SiluAndMul`. By default,
	[kernelize()](/docs/kernels/pr_607/en/api/layers#kernels.kernelize) will try to infer the device type from the model's parameters.
	You can pass the device type to [kernelize()](/docs/kernels/pr_607/en/api/layers#kernels.kernelize) if the device type cannot be
	inferred (e.g. because the model has no parameters):

	```python
	model = MyModel(...)
	model = kernelize(model, device="cuda", mode=Mode.INFERENCE)
	```

	### Fallback `forward`

	If the `TRAINING` and/or `TORCH_COMPILE` modes are used, but a registered
	kernel does not support backward passes or `torch.compile` respectively,
	[kernelize()](/docs/kernels/pr_607/en/api/layers#kernels.kernelize) will fall back to the original, non-kernelized, layer. You
	can let [kernelize()](/docs/kernels/pr_607/en/api/layers#kernels.kernelize) raise an exception instead by using `use_fallback=False`:

	```python
	model = MyModel(...)
	model = kernelize(model, mode=Mode.INFERENCE \| Mode.TORCH_COMPILE, use_fallback=False)
	```

	This can be useful if you want to guarantee that Hub kernels are used.

	### Inspecting which kernels are used

	The kernels that are used are logged at the `INFO` level by [kernelize()](/docs/kernels/pr_607/en/api/layers#kernels.kernelize).
	See the [Python logging](https://docs.python.org/3/library/logging.html)
	documentation for information on how to configure logging.

	## Registering a hub kernel for a layer

	[kernelize()](/docs/kernels/pr_607/en/api/layers#kernels.kernelize) relies on kernel mappings to find Hub kernels for layers.
	Kernel mappings map a kernel name such as `SiluAndMul` to a kernel on
	the Hub. For example:

	```python
	kernel_layer_mapping = {
	"SiluAndMul": {
	"cuda": LayerRepository(
	repo_id="kernels-community/activation",
	layer_name="SiluAndMul",
	version=1,
	),
	"rocm": LayerRepository(
	repo_id="kernels-community/activation",
	layer_name="SiluAndMul",
	version=1,
	)
	}
	}
	```

	This uses version `1` of the `SiluAndMul` kernel layer from
	`kernels-community/activation` for the `cuda` and `rocm` backends. Kernel
	layers are versioned using a major version number. Using `version=1`
	will get the latest kernel build from the `v1` branch. Kernel layers
	within a version branch must never break the API or remove builds for
	older PyTorch versions. This ensures that your code will continue to
	work.
	Hub-backed [LayerRepository](/docs/kernels/pr_607/en/api/layers#kernels.LayerRepository) and [FuncRepository](/docs/kernels/pr_607/en/api/layers#kernels.FuncRepository) entries must specify
	either a `version` or an explicit `revision`.

	You can register a mapping, like the one above, using [register_kernel_mapping()](/docs/kernels/pr_607/en/api/layers#kernels.register_kernel_mapping):

	```python
	register_kernel_mapping(kernel_layer_mapping)
	```

	This will register the kernel mapping in the current context, which is
	normally global. It is recommended to scope the mapping to where it is
	used with the [use_kernel_mapping()](/docs/kernels/pr_607/en/api/layers#kernels.use_kernel_mapping) context manager:

	```python
	with use_kernel_mapping(kernel_layer_mapping):
	# Use the layer for which the mapping is applied.
	model = kernelize(model, mode=Mode.TRAINING \| Mode.TORCH_COMPILE)
	```

	This ensures that the mapping is not active anymore outside the
	`with`-scope.

	If the layer is stateless (it does not use member variables in its forward _or_ it was
	originally a function that was converted into a kernel layer with
	[use_kernel_func_from_hub()](/docs/kernels/pr_607/en/api/layers#kernels.use_kernel_func_from_hub)), it can also be mapped to a kernel function:

	```python
	kernel_layer_mapping = {
	"SiluAndMul": {
	"cuda": FuncRepository(
	repo_id="kernels-community/activation",
	func_name="silu_and_mul",
	version=1,
	),
	}
	}
	```

	### Registering kernels for specific modes

	You might want to register two different kernels for a particular layer,
	where one kernel is optimized for a specific mode. You can do so by
	registering layer repositories for specific modes. For example:

	```python
	kernel_layer_mapping = {
	"SiluAndMul": {
	"cuda": {
	Mode.INFERENCE: LayerRepository(
	repo_id="kernels-community/activation-inference-optimized",
	layer_name="SiluAndMul",
	version=1,
	),
	Mode.TRAINING \| Mode.TORCH_COMPILE: LayerRepository(
	repo_id="kernels-community/activation-training-optimized",
	layer_name="SiluAndMul",
	version=1,
	),
	}
	}
	}
	```

	The [kernelize()](/docs/kernels/pr_607/en/api/layers#kernels.kernelize) function will attempt to use the following registered
	kernels for a given mode:

	- `INFERENCE`: `INFERENCE` → `INFERENCE \| TORCH_COMPILE` → `TRAINING` →
	`TRAINING \| TORCH_COMPILE` → `FALLBACK`
	- `INFERENCE \| TORCH_COMPILE`: `INFERENCE \| TORCH_COMPILE` →
	`TRAINING \| TORCH_COMPILE` → `FALLBACK`
	- `TRAINING`: `TRAINING` → `TRAINING \| TORCH_COMPILE` → `FALLBACK`
	- `TRAINING \| TORCH_COMPILE`: `TRAINING \| TORCH_COMPILE` → `FALLBACK`

	`Mode.FALLBACK` is a special mode that is used when no other mode matches. It
	is also used when a kernel is registered without a mode, as described in the
	previous section.

	```python
	kernel_layer_mapping = {
	"SiluAndMul": {
	"cuda": {
	Mode.FALLBACK: LayerRepository(
	repo_id="kernels-community/activation",
	layer_name="SiluAndMul",
	version=1,
	),
	Mode.INFERENCE: LayerRepository(
	repo_id="kernels-community/activation-inference-optimized",
	layer_name="SiluAndMul",
	version=1,
	),
	Mode.TRAINING: LayerRepository(
	repo_id="kernels-community/activation-training-optimized",
	layer_name="SiluAndMul",
	version=1,
	),
	}
	}
	}
	```

	In this case, both `Mode.INFERENCE \| Mode.TORCH_COMPILE` and
	`Mode.TRAINING \| Mode.TORCH_COMPILE` will use the `Mode.FALLBACK` kernel,
	since the other kernels do not support `torch.compile`.

	### Registering kernels for specific CUDA capabilities

	Some kernels only work with newer CUDA architectures. For instance, some
	kernels require capability 9.0 for the TMA unit on Hopper GPUs. `kernels`
	supports registering layers for a range of CUDA capabilities. To do so,
	you need to register the layer for a [Device](/docs/kernels/pr_607/en/api/layers#kernels.Device) with type `cuda` and
	set the supported range of CUDA capabilities with using `CUDAProperties`:

	```python
	kernel_layer_mapping = {
	"SiluAndMul": {
	Device(
	type="cuda",
	properties=CUDAProperties(
	min_capability=75, max_capability=89
	),
	): LayerRepository(
	repo_id="kernels-community/activation",
	layer_name="SiluAndMul",
	version=1,
	),
	Device(
	type="cuda",
	properties=CUDAProperties(
	min_capability=90, max_capability=sys.maxsize
	),
	): LayerRepository(
	repo_id="kernels-community/activation-hopper",
	layer_name="SiluAndMul",
	version=1,
	),
	}
	}
	```

	Capabilities behave as follows:

	- The minimum and maximum capabilities are inclusive.
	- When a new kernel is registered with the same min/max capabilities as
	an existing kernel, the new kernel will replace the old kernel.
	- When there are multiple kernels that support a capability, the kernel
	with the smaller capability interval will be used. E.g. given:
	- `KernelA` with `min_capability=80` and `max_capability=89`;
	- `KernelB` with `min_capability=75` and `max_capability=89`;
	- [kernelize()](/docs/kernels/pr_607/en/api/layers#kernels.kernelize) runs on a system with capability 8.6.

	Then `KernelA` will be used because the interval 80..89 is smaller
	than 75..89. The motivation is that kernels with smaller ranges
	tend to be more optimized for a specific set of GPUs. **This behavior
	might still change in the future.**

	### Registering kernels for specific ROCm capabilities

	Registering kernels for the ROCm architecture follows the exact same
	pattern as CUDA kernels, using `min_capability` and `max_capability` to restrict
	a kernel to a range of ROCm capabilities.

	### Loading from a local repository for testing

	The [LocalLayerRepository](/docs/kernels/pr_607/en/api/layers#kernels.LocalLayerRepository) class is provided to load a repository from
	a local directory. For example:

	```python
	with use_kernel_mapping(
	{
	"SiluAndMul": {
	"cuda": LocalLayerRepository(
	repo_path="/home/daniel/kernels/activation",
	package_name="activation",
	layer_name="SiluAndMul",
	)
	}
	},
	inherit_mapping=False,
	):
	kernelize(linear, mode=Mode.INFERENCE)
	```

	Similarly, the [LocalFuncRepository](/docs/kernels/pr_607/en/api/layers#kernels.LocalFuncRepository) class can be used to load a kernel
	function from a local directory:

	```python
	with use_kernel_mapping(
	{
	"silu_and_mul": {
	"cuda": LocalFuncRepository(
	repo_path="/home/daniel/kernels/activation",
	package_name="activation",
	func_name="silu_and_mul",
	)
	}
	},
	inherit_mapping=False,
	):
	kernelize(model, mode=Mode.INFERENCE)
	```

	### Kernel requirements
	https://huggingface.co/docs/kernels/pr_607/kernel-requirements.md

	# Kernel requirements

	Kernels on the Hub must fulfill the requirements outlined on this page. By
	ensuring kernels are compliant, they can be used on a wide range of Linux
	systems and Torch builds.

	[Join us on Discord](https://discord.gg/H6Tkmd88N3) for questions and discussions
	about building kernels!

	## Repository type

	Compliant kernels are published as `kernel`-type repositories on the Hub
	(the first-class kernel repository type). New uploads via `kernel-builder`
	default to this type; see the [migration guide](migration) if you
	maintain an older `model`-type kernel repository.

	## Trusted publishers

	`kernels` only loads kernels from a curated set of trusted publishers by
	default. Loading from any other publisher raises an error unless the caller
	opts in with `trust_remote_code=True`:

	```python
	# Trusted publisher: works without opt-in.
	get_kernel("kernels-community/activation", version=1)

	# Untrusted publisher: must opt in explicitly.
	get_kernel("some-other-org/my-kernel", version=1, trust_remote_code=True)
	```

	The Hub also exposes a `trustedKernelPublisher` flag on the kernel API and
	displays a corresponding badge in the UI.

	## Directory layout

	A kernel repository on the Hub must contain a `build` directory. This
	directory contains build variants of a kernel in the form of directories
	following the template
	`<framework><version>-cxx<abiver>-<cu><cudaver>-<arch>-<os>`.
	For example `build/torch26-cxx98-cu118-x86_64-linux`.

	The kernel is in the build variant directory and must contain a
	`__init__.py` file. For compatibility with older versions of the
	`kernels` package, each variant directory must also contain a single
	directory with the same name as the repository (replacing `-` by `_`).
	For instance, kernels in the `kernels-community/activation` repository
	have a directory like `build/<variant>/activation`. This directory
	must contain an `__init__.py` file that exports the same symbols as
	`__init__.py` in the build variant directory `build/<variant>`.
	[This example](https://huggingface.co/kernels-test/flattened-build/blob/main/build/torch-universal/flattened_build/__init__.py)
	shows how this can be done. This compatibility directory is
	automatically created by `kernel-builder`.

	## Build variants

	A kernel can be compliant for a specific compute framework (e.g. CUDA) or
	architecture (e.g. x86_64). For compliance with a compute framework and
	architecture combination, all the variants from the [build variant list](builder/build-variants)
	must be available for that combination.

	## Kernel metadata

	The build variant directory must contain a `metadata.json` file with kernel
	metadata. Currently the following top-level keys are supported:

	- `id` (`str`, required): a unique identifier for the kernel. This
	identifier must also be a valid Python module name. If the kernel
	registers Torch ops, they must be registered as `torch.ops.<id>`
	- `name` (`str`, required): then name of the kernel. Replacing dashes
	by underscores should result in the module name of the kernel.
	- `version` (`int`, required): the kernel version number.
	- `license` (`str`, required): the kernel license in. Refer to the
	list of [supported license identifiers](https://huggingface.co/docs/hub/repositories-licenses).
	- `backend` (`dict`, required): information about the compute backend that
	this build variant supports.
	- `python-depends` (`list[str]`, optional): list of Python dependencies
	from a curated set of Python dependencies.

	Example `metadata.json`:

	```json
	{
	"name": "mykernel",
	"id": "_mykernel_cuda_7a4e5a7",
	"version": 1,
	"license": "Apache-2.0",
	"python-depends": ["einops"],
	"backend": {
	"type": "cuda",
	"archs": ["7.0", "7.2", "7.5", "8.0", "8.6", "8.7", "8.9", "9.0+PTX"]
	}
	}
	```

	The `metadata.json` file is generated automatically by `kernel-builder`.

	## Backend

	The `backend` specifies a dictionary of the following form:

	```json
	{
	# ...
	"backend": {
	"type": "cuda",
	"archs": ["7.0", "7.2", "7.5", "8.0", "8.6", "8.7", "8.9", "9.0+PTX"]
	}
	}
	```

	The backend `type` must be one of `cann`, `cpu`, `cuda`, `metal`, `neuron`,
	`rocm`, or `xpu`. For CUDA and ROCm, the supported architectures must
	be specified in the `archs` field.

	### Python dependencies

	You can specify Python dependencies that your kernel requires. Dependencies can be either general (required for all backends) or backend-specific (required only for certain compute backends like CUDA, ROCm, XPU, Metal, or CPU).

	#### General dependencies

	For dependencies required regardless of the backend, use the `python-depends` field:

	```json
	{
	"python-depends": ["einops"]
	}
	```

	#### Backend-specific dependencies

	For dependencies that are only needed for specific backends, use the `python-depends-backends` field:

	```json
	{
	"python-depends-backends": {
	"cuda": ["nvidia-cutlass-dsl"],
	"xpu": ["onednn"]
	}
	}
	```

	#### Combined example

	You can specify both general and backend-specific dependencies:

	```json
	{
	"python-depends": ["einops"],
	"python-depends-backends": {
	"cuda": ["nvidia-cutlass-dsl"]
	},
	"version": 1
	}
	```

	#### Allowed dependencies

	The following dependencies are currently allowed:

	General dependencies:

	- `einops`

	Backend-specific dependencies:

	- CUDA: `nvidia-cutlass-dsl`
	- XPU: `onednn`

	Dependencies are validated based on the backend being used. When a kernel is loaded, only the dependencies relevant to the active backend are checked.

	## Versioning

	Kernels are versioned using a major version. The kernel revisions of a
	version are stored in a branch of the form `v<version>`. Each build
	variant will also have the kernel version in `metadata.json`.

	The version must be bumped in the following cases:

	- The kernel API is changed in an incompatible way.
	- The API is extended in a compatible way, but not all build variants
	receive the extension (e.g. because they are for older Torch versions
	that are not supported by `kernel-builder` anymore).

	In both cases, build variants that are not updated must be removed from
	the new version's branch.

	## Native Python module

	Kernels will typically contain a native Python module with precompiled
	compute kernels and bindings. This module must fulfill the requirements
	outlined in this section. For all operating systems, a kernel must not
	have dynamic library dependencies outside:

	- Torch;
	- CUDA/ROCm libraries installed as dependencies of Torch.

	## Compatibility with torch.compile

	The Kernel Hub also encourages to write the kernels in a `torch.compile`
	compliant way. This helps to ensure that the kernels are compatible with
	`torch.compile` without introducing any graph breaks and triggering
	recompilation which can limit the benefits of compilation.

	[Here](https://github.com/huggingface/kernels/blob/f83b4da6b7f6b171b47bb9bf96271ae2273bc9d3/builder/examples/relu-backprop-compile/tests/test_relu.py#L162)
	is a simple test example which checks for graph breaks and
	recompilation triggers during `torch.compile`.

	### Linux

	- Use [ABI3/Limited API](https://docs.python.org/3/c-api/stable.html#stable-application-binary-interface)
	for compatibility with Python 3.9 and later.
	- Compatible with [`manylinux_2_28`](https://github.com/pypa/manylinux?tab=readme-ov-file#manylinux_2_28-almalinux-8-based).
	This means that the extension must not use symbols versions higher than:
	- GLIBC 2.28
	- GLIBCXX 3.4.24
	- CXXABI 1.3.11
	- GCC 7.0.0

	These requirements can be checked with the ABI checker (see below).

	### macOS

	- Use [ABI3/Limited API](https://docs.python.org/3/c-api/stable.html#stable-application-binary-interface)
	for compatibility with Python 3.9 and later.
	- macOS deployment target 15.0.
	- Metal 3.0 (`-std=metal3.0`).

	The ABI3 requirement can be checked with the ABI checker (see below).

	### ABI checker

	The manylinux_2_28 and Python ABI 3.9 version requirements can be checked with
	`kernel-builder check-abi`:

	```bash
	$ kernel-builder check-abi examples/kernels/relu
	🐍 Checking for compatibility with manylinux_2_28 and Python ABI version 3.9: /home/daniel/git/kernels/examples/kernels/relu/result/torch211-cpu-x86_64-linux/_relu_cpu_30dc0ae_dirty.abi3.so
	✅ No compatibility issues found
	🐍 Checking for compatibility with manylinux_2_28 and Python ABI version 3.9: /home/daniel/git/kernels/examples/kernels/relu/result/torch211-cu126-x86_64-linux/_relu_cuda_30dc0ae_dirty.abi3.so
	✅ No compatibility issues found
	🐍 Checking for compatibility with manylinux_2_28 and Python ABI version 3.9: /home/daniel/git/kernels/examples/kernels/relu/result/torch211-cu128-x86_64-linux/_relu_cuda_30dc0ae_dirty.abi3.so
	✅ No compatibility issues found
	🐍 Checking for compatibility with manylinux_2_28 and Python ABI version 3.9: /home/daniel/git/kernels/examples/kernels/relu/result/torch211-cu130-x86_64-linux/_relu_cuda_30dc0ae_dirty.abi3.so
	✅ No compatibility issues found
	[...]
	```

	## Torch extension

	Torch native extension functions must be [registered](https://pytorch.org/tutorials/advanced/cpp_custom_ops.html#cpp-custom-ops-tutorial)
	in `torch.ops.<namespace>`. Since we allow loading of multiple versions of
	a module in the same Python process, `namespace` must be unique for each
	version of a kernel. Failing to do so will create clashes when different
	versions of the same kernel are loaded. Two suggested ways of doing this
	are:

	- Appending a truncated SHA-1 hash of the git commit that the kernel was
	built from to the name of the extension.
	- Appending random material to the name of the extension.

	Note: we recommend against appending a version number or git tag.
	Version numbers are typically not bumped on each commit, so users
	might use two different commits that happen to have the same version
	number. Git tags are not stable, so they do not provide a good way
	of guaranteeing uniqueness of the namespace.

	## Layers

	A kernel can provide layers in addition to kernel functions. A layer from
	the Hub can replace the `forward` method of an existing layer for a certain
	device type. This makes it possible to provide more performant kernels for
	existing layers. See the [layers documentation](layers) for more information
	on how to use layers.

	### Writing layers

	To make the extension of layers safe, the layers must fulfill the following
	requirements:

	- The layers are subclasses of `torch.nn.Module`.
	- The layers are pure, meaning that they do not have their own state. This
	means that:
	- The layer must not define its own constructor.
	- The layer must not use class variables.
	- No other methods must be defined than `forward`.
	- The `forward` method has a signature that is compatible with the
	`forward` method that it is extending.

	There are two exceptions to the _no class variables rule_:

	1. The `has_backward` variable can be used to indicate whether the layer has
	a backward pass implemented (`True` when absent).
	2. The `can_torch_compile` variable can be used to indicate whether the layer
	supports `torch.compile` (`False` when absent).

	This is an example of a pure layer:

	```python
	class SiluAndMul(nn.Module):
	# This layer does not implement backward.
	has_backward: bool = False

	def forward(self, x: torch.Tensor):
	d = x.shape[-1] // 2
	output_shape = x.shape[:-1] + (d,)
	out = torch.empty(output_shape, dtype=x.dtype, device=x.device)
	ops.silu_and_mul(out, x)
	return out
	```

	For some layers, the `forward` method has to use state from the adopting class.
	In these cases, we recommend to use type annotations to indicate what member
	variables are expected. For instance:

	```python
	class LlamaRMSNorm(nn.Module):
	weight: torch.Tensor
	variance_epsilon: float

	def forward(self, hidden_states: torch.Tensor) -> torch.Tensor:
	return rms_norm_fn(
	hidden_states,
	self.weight,
	bias=None,
	residual=None,
	eps=self.variance_epsilon,
	dropout_p=0.0,
	prenorm=False,
	residual_in_fp32=False,
	)
	```

	This layer expects the adopting layer to have `weight` and `variance_epsilon`
	member variables and uses them in the `forward` method.

	### Exporting layers

	To accommodate portable loading, `layers` must be defined in the main
	`__init__.py` file. For example:

	```python
	from . import layers

	__all__ = [
	# ...
	"layers"
	# ...
	]
	```

	## Python requirements

	- Python code must be compatible with Python 3.9 and later.
	- All Python code imports from the kernel itself must be relative. So,
	for instance if in the example kernel `example`,
	`module_b` needs a function from `module_a`, import as:

	```python
	from .module_a import foo
	```

	Never use:

	```python
	# DO NOT DO THIS!

	from example.module_a import foo
	```

	The latter would import from the module `example` that is in Python's
	global module dict. However, since we allow loading multiple versions
	of a module, we uniquely name the module.

	- Only modules from the Python standard library, Torch, or the kernel itself
	can be imported.

	### Locking kernel/layer versions
	https://huggingface.co/docs/kernels/pr_607/locking.md

	# Locking kernel/layer versions

	Projects that use `setuptools` can lock the kernel versions that should be
	used. First specify the accepted versions in `pyproject.toml` and make
	sure that `kernels` is a build dependency:

	```toml
	[build-system]
	requires = ["kernels", "setuptools"]
	build-backend = "setuptools.build_meta"

	[tool.kernels.dependencies]
	"kernels-community/activation" = 1
	```

	Then run `kernels lock .` in the project directory. This generates a `kernels.lock` file with
	the locked revisions. The locked revision will be used when loading a kernel with
	[get_locked_kernel()](/docs/kernels/pr_607/en/api/kernels#kernels.get_locked_kernel):

	```python
	from kernels import get_locked_kernel

	activation = get_locked_kernel("kernels-community/activation")
	```

	Note: the lock file is included in the package metadata, so it will only be visible
	to `kernels` after doing an (editable or regular) installation of your project.

	## Locked kernel layers

	Locking is also supported for kernel layers. To use locked layers, register them
	with the [LockedLayerRepository](/docs/kernels/pr_607/en/api/layers#kernels.LockedLayerRepository) class:

	```python
	kernel_layer_mapping = {
	"SiluAndMul": {
	"cuda": LockedLayerRepository(
	repo_id="kernels-community/activation",
	layer_name="SiluAndMul",
	)
	}
	}

	register_kernel_mapping(kernel_layer_mapping)
	```

	Similarly, you can use the [LockedFuncRepository](/docs/kernels/pr_607/en/api/layers#kernels.LockedFuncRepository) class to lock kernel function
	versions:

	```python
	kernel_layer_mapping = {
	"silu_and_mul": {
	"cuda": LockedFuncRepository(
	repo_id="kernels-community/activation",
	func_name="silu_and_mul",
	)
	}
	}

	register_kernel_mapping(kernel_layer_mapping)
	```

	## Pre-downloading locked kernels

	Locked kernels can be pre-downloaded by running `kernels download .` in your
	project directory. This will download the kernels to your local Hugging Face
	Hub cache.

	The pre-downloaded kernels are used by the [get_locked_kernel()](/docs/kernels/pr_607/en/api/kernels#kernels.get_locked_kernel) function.
	[get_locked_kernel()](/docs/kernels/pr_607/en/api/kernels#kernels.get_locked_kernel) will download a kernel when it is not pre-downloaded. If you
	want kernel loading to error when a kernel is not pre-downloaded, you can use
	the [load_kernel()](/docs/kernels/pr_607/en/api/kernels#kernels.load_kernel) function instead:

	```python
	from kernels import load_kernel

	activation = load_kernel("kernels-community/activation")
	```

	### Integrating kernels
	https://huggingface.co/docs/kernels/pr_607/integrating-kernels.md

	# Integrating kernels

	This page shows how different projects use `kernels`.

	## autoresearch

	[karpathy/autoresearch](https://github.com/karpathy/autoresearch) [uses](https://github.com/karpathy/autoresearch/blob/c2450add72cc80317be1fe8111974b892da10944/train.py#L23) `kernels` to
	integrate Flash-Attention 3 through the [get_kernel()](/docs/kernels/pr_607/en/api/kernels#kernels.get_kernel) method.

	## AReaL

	[inclusionAI/AReaL](https://github.com/inclusionAI/AReaL) uses `kernels` in an opt-in manner to integrate
	optimized attention mechanisms.

	## transformers

	[huggingface/transformers](https://github.com/huggingface/transformers/) primarily
	depends on `kernels` for all optimizations related to optimized kernels, including
	optimized attention implementations, MoE blocks, and quantization. Besides
	[get_kernel()](/docs/kernels/pr_607/en/api/kernels#kernels.get_kernel), it also uses [kernel layers](./layers) to optimize the forward passes
	of common layers involved in the modeling blocks. Some references are available
	[here]()
	and [here](https://github.com/search?q=repo%3Ahuggingface%2Ftransformers+use_kernel_forward_from_hub&type=code).

	Refer to the following posts to know more:

	* [Tricks from OpenAI gpt-oss YOU 🫵 can use with transformers](https://huggingface.co/blog/faster-transformers)
	* [Mixture of Experts (MoEs) in Transformers](https://huggingface.co/blog/moe-transformers)

	## diffusers

	Similar to `transformers`, [huggingface/diffusers](https://github.com/huggingface/diffusers/) uses
	`kernels` for integrating optimized kernels to [compute attention](https://github.com/huggingface/diffusers/blob/e5aa719241f9b74d6700be3320a777799bfab70a/src/diffusers/models/attention_dispatch.py).

	Besides leveraging pre-built compute kernels, different projects
	rely on `kernels` to also package, build, and distribute their
	kernels on the Hugging Face Hub platform. This is made possible by the
	["builder" component of `kernels`](./builder/writing-kernels).
	Visit [huggingface.co/kernels](https://huggingface.co/kernels) to browse
	the pre-built compute kernels available on the Hub.

	Feel free to open a PR enlisting your project to show how `kernels`
	is leveraged there.

	### kernels versions
	https://huggingface.co/docs/kernels/pr_607/cli-versions.md

	# kernels versions

	Use `kernels versions` to list all available versions of a kernel on the Hub
	and marks compatible versions.

	## Usage

	```bash
	kernels versions <repo_id>
	```

	## Examples

	List versions of a kernel:

	```bash
	kernels versions kernels-community/activation
	```

	## Example Output

	```text
	Version 1: torch210-metal-aarch64-darwin, torch28-cxx11-cu126-aarch64-linux, torch28-cxx11-cu129-aarch64-linux, torch28-cxx11-cu128-aarch64-linux, torch29-cxx11-cu130-x86_64-linux, torch27-cxx11-cu118-x86_64-linux, torch210-cxx11-cu130-x86_64-linux, torch29-cxx11-cu128-aarch64-linux, torch29-cxx11-cu130-aarch64-linux, torch27-cxx11-cu126-x86_64-linux, ✅ torch29-cxx11-cu126-x86_64-linux (compatible), torch27-cxx11-cu128-x86_64-linux, torch210-cxx11-cu126-x86_64-linux, torch29-metal-aarch64-darwin, torch27-cxx11-cu128-aarch64-linux, torch210-cu128-x86_64-windows, torch28-cxx11-cu128-x86_64-linux, torch28-cxx11-cu126-x86_64-linux, torch210-cxx11-cu128-x86_64-linux, torch29-cxx11-cu126-aarch64-linux, ✅ torch29-cxx11-cu128-x86_64-linux (preferred), torch28-cxx11-cu129-x86_64-linux
	```

	## Use Cases

	- Check which versions are available before locking dependencies
	- Find the latest version of a kernel
	- Identify version SHAs for pinning in `pyproject.toml`

	## See Also

	- [kernels lock](cli-lock) - Lock kernel versions in your project
	- [kernels download](cli-download) - Download locked kernels

	### kernel-builder skills add
	https://huggingface.co/docs/kernels/pr_607/cli-skills.md

	### kernel-builder skills add

	Use `kernel-builder skills add` to install the skills for AI coding assistants like Claude, Codex, and OpenCode.
	Supported skills include:
	- `cuda-kernels` (default)
	- `rocm-kernels`
	- `xpu-kernels`

	Skill files are downloaded from the `huggingface/kernels` directory in this [repository](https://github.com/huggingface/kernels/tree/main/kernel-builder/skills).

	Skills instruct agents how to deal with hardware-specific optimizations, integrate with libraries like diffusers and transformers, and benchmark kernel performance in consistent ways.

	Examples:

	```bash
	<CopyLLMTxtMenu containerStyle="float: right; margin-left: 10px; display: inline-flex; position: relative; z-index: 10;"></CopyLLMTxtMenu>

	# install for Claude in the current project
	kernel-builder skills add --claude

	# install ROCm kernels skill for Codex
	kernel-builder skills add --skill rocm-kernels --codex

	# install globally for Codex
	kernel-builder skills add --codex --global

	# install for multiple assistants
	kernel-builder skills add --claude --codex --opencode

	# install to a custom destination and overwrite if already present
	kernel-builder skills add --dest ~/my-skills --force
	```

	### Talks
	https://huggingface.co/docs/kernels/pr_607/talks.md

	# Talks

	This page lists talks on the Kernels project, delivered at
	different events:

	* [Lecture 106: Hugging Face Kernels (GPU Mode)](https://www.youtube.com/watch?v=Ok8vi6JemVQ)
	* [PyTorch Day India 2026 Beyond the Brrr: Building a Unified Ecosystem for Optimized Kernels](https://www.youtube.com/watch?v=q0GfzJmuaUM)
	* [Build a PyTorch ReLU Kernel with Hugging Face Kernels (CPU + Metal)](https://youtu.be/wQR-QC7pbqQ?is=JY4pvjXpkz8ghgwS)

	This page will be kept updated.

	### Migrating from older versions
	https://huggingface.co/docs/kernels/pr_607/migration.md

	# Migrating from older versions

	## 0.12

	### Adopting kernel versions

	Before `kernels` 0.12, kernels could be pulled from a repository
	without specifying a version. This is deprecated in kernels 0.12
	and is an error in kernels 0.15. Instead, use of a kernel should
	always specify a version or revision (except for local kernels).

	Kernels only use a major version. The kernel maintainer is responsible
	for never breaking a kernel within a major version and should bump up
	the major version if the kernel API changes and/or when support for
	older Torch versions is removed.

	You can find the versions that are supported by a kernel using the
	`kernels versions command`. For example:

	```bash
	$ kernels versions kernels-community/activation
	Version 1: torch210-cxx11-cu126-x86_64-linux, torch210-cxx11-cu128-x86_64-linux, torch210-cxx11-cu130-x86_64-linux, torch27-cxx11-cu118-x86_64-linux, torch27-cxx11-cu126-x86_64-linux, torch27-cxx11-cu128-aarch64-linux, torch27-cxx11-cu128-x86_64-linux ✅, torch28-cxx11-cu126-aarch64-linux, torch28-cxx11-cu126-x86_64-linux, torch28-cxx11-cu128-aarch64-linux, torch28-cxx11-cu128-x86_64-linux, torch28-cxx11-cu129-aarch64-linux, torch28-cxx11-cu129-x86_64-linux, torch29-cxx11-cu126-aarch64-linux, torch29-cxx11-cu126-x86_64-linux, torch29-cxx11-cu128-aarch64-linux, torch29-cxx11-cu128-x86_64-linux, torch29-cxx11-cu130-aarch64-linux, torch29-cxx11-cu130-x86_64-linux
	```

	The command lists all available versions (here only version 1) with
	all the variants that are supported. A check mark is printed after
	the variant that is compatible with your current environment.

	Code that uses a kernel can be updated as follows:

	```python
	# Old:
	activation = get_kernel("kernels-community/activation")
	activation = get_kernel("kernels-community/activation", version=">=0.0.2 && <0.1.0")

	# New:
	activation = get_kernel("kernels-community/activation", version=1)

	# Old:
	kernel_layer_mapping = {
	"SiluAndMul": {
	"cuda": LayerRepository(
	repo_id="kernels-community/activation",
	layer_name="SiluAndMul",
	),
	}
	}
	kernel_layer_mapping = {
	"SiluAndMul": {
	"cuda": LayerRepository(
	repo_id="kernels-community/activation",
	layer_name="SiluAndMul",
	version=">=0.0.2 && <0.1.0",
	),
	}
	}

	# New:
	kernel_layer_mapping = {
	"SiluAndMul": {
	"cuda": LayerRepository(
	repo_id="kernels-community/activation",
	layer_name="SiluAndMul",
	version=1,
	),
	}
	}
	```

	## 0.14

	### `kernel` repo type on the Hub

	Kernels are now a first-class repository type on the Hugging Face Hub, and
	`kernels` 0.14 loads kernels exclusively from `kernel`-type repositories.
	`model`-type kernel repositories are no longer supported by the loader.

	New uploads via `kernel-builder build-and-upload` default to
	`--repo-type kernel`. To publish, the owning user or org must have
	kernel-creation access. Request it from
	[huggingface.co/settings/account](https://huggingface.co/settings/account)
	("Request Kernels Creation").

	To migrate an existing `model`-type kernel repository:

	1. Make sure the publishing org has been granted kernel-creation access
	(see above).
	2. Re-upload with `kernel-builder build-and-upload` to a `kernel`-type
	repository. Either keep the same `repo-id` in `build.toml` if the
	repository has been migrated to the new type, or point it at a newly
	created `kernel`-type repository.
	3. Update consumers' [get_kernel()](/docs/kernels/pr_607/en/api/kernels#kernels.get_kernel) and [LayerRepository](/docs/kernels/pr_607/en/api/layers#kernels.LayerRepository) calls
	to reference the new repository if the `repo-id` changed.

	### kernels download
	https://huggingface.co/docs/kernels/pr_607/cli-download.md

	# kernels download

	Use `kernels download` to download kernels that have been locked in a project's `kernels.lock` file.

	## Usage

	```bash
	kernels download <project_dir> [--all-variants]
	```

	## What It Does

	- Reads the `kernels.lock` file from the specified project directory
	- Downloads each locked kernel at its pinned revision (SHA)
	- Installs the appropriate variant for your platform (or all variants with `--all-variants`)

	## Examples

	Download kernels for the current project:

	```bash
	kernels download .
	```

	Download all build variants (useful for CI or multi-platform builds):

	```bash
	kernels download . --all-variants
	```

	Download kernels for a specific project:

	```bash
	kernels download /path/to/my-project
	```

	## Options

	\| Option \| Description \|
	\| ---------------- \| ----------------------------------------------------------------------------------------- \|
	\| `--all-variants` \| Download all build variants of each kernel instead of just the current platform's variant \|

	## Prerequisites

	Your project directory must contain a `kernels.lock` file. Generate one using [`kernels lock`](cli-lock).

	## See Also

	- [kernels lock](cli-lock) - Generate the lock file
	- [kernels versions](cli-versions) - View available kernel versions

	### Kernels API Reference
	https://huggingface.co/docs/kernels/pr_607/api/kernels.md

	# Kernels API Reference

	## Main Functions

	### get_kernel[[kernels.get_kernel]]

	#### kernels.get_kernel[[kernels.get_kernel]]

	[Source](https://github.com/huggingface/kernels/blob/vr_607/kernels/src/kernels/utils.py#L415)

	Load a kernel from the kernel hub.

	This function downloads a kernel to the local Hugging Face Hub cache directory (if it was not downloaded before)
	and then loads the kernel.

	Example:
	```python
	import torch
	from kernels import get_kernel

	activation = get_kernel("kernels-community/relu", version=1)
	x = torch.randn(10, 20, device="cuda")
	out = torch.empty_like(x)
	result = activation.relu(out, x)
	```

	Parameters:

	repo_id (str) : The Hub repository containing the kernel.

	revision (str, optional) : The specific revision (branch, tag, or commit) to download. Cannot be used together with version.

	version (int, optional) : The kernel version to download. Cannot be used together with revision. Either version or revision must be specified.

	backend (str, optional) : The backend to load the kernel for. Can only be cpu or the backend that Torch is compiled for. The backend will be detected automatically if not provided.

	user_agent (Union[str, dict], optional) : The user_agent info to pass to snapshot_download() for internal telemetry.

	trust_remote_code (bool \| list[str], optional, defaults to False) : Whether to allow loading kernels from untrusted organisations. When `False`, only kernels from trusted organisations are allowed. When `True`, all repositories are allowed. A list of strings will be used to verify signing identities in a future release; for now it emits a warning and falls back to the default trust check.

	Returns:

	`ModuleType`

	The imported kernel module.

	### get_local_kernel[[kernels.get_local_kernel]]

	#### kernels.get_local_kernel[[kernels.get_local_kernel]]

	[Source](https://github.com/huggingface/kernels/blob/vr_607/kernels/src/kernels/utils.py#L483)

	Import a kernel from a local kernel repository path.

	Parameters:

	repo_path (`Path`) : The local path to the kernel repository.

	backend (`str`, optional) : The backend to load the kernel for. Can only be `cpu` or the backend that Torch is compiled for. The backend will be detected automatically if not provided.

	Returns:

	``ModuleType``

	The imported kernel module.

	### has_kernel[[kernels.has_kernel]]

	#### kernels.has_kernel[[kernels.has_kernel]]

	[Source](https://github.com/huggingface/kernels/blob/vr_607/kernels/src/kernels/utils.py#L516)

	Check whether a kernel build exists for the current environment (Torch version and compute framework).

	Parameters:

	repo_id (`str`) : The Hub repository containing the kernel.

	revision (`str`, optional) : The specific revision (branch, tag, or commit) to download. Cannot be used together with `version`.

	version (`int`, optional) : The kernel version to download. Cannot be used together with `revision`. Either `version` or `revision` must be specified.

	backend (`str`, optional) : The backend to load the kernel for. Can only be `cpu` or the backend that Torch is compiled for. The backend will be detected automatically if not provided.

	Returns:

	``bool``

	`True` if a kernel is available for the current environment.

	### get_kernel_variants[[kernels.get_kernel_variants]]

	#### kernels.get_kernel_variants[[kernels.get_kernel_variants]]

	[Source](https://github.com/huggingface/kernels/blob/vr_607/kernels/src/kernels/utils.py#L557)

	Resolve all build variants of a kernel against the current environment.

	The decisions are sorted with compatible variants first, the most preferred
	variant leading.

	Parameters:

	repo_id (`str`) : The Hub repository containing the kernel.

	revision (`str`, optional) : The specific revision (branch, tag, or commit) to inspect. Cannot be used together with `version`.

	version (`int`, optional) : The kernel version to inspect. Cannot be used together with `revision`. Either `version` or `revision` must be specified.

	backend (`str`, optional) : The backend to resolve variants for. Can only be `cpu` or the backend that Torch is compiled for. The backend will be detected automatically if not provided.

	Returns:

	``list[Decision]``

	One `VariantAccepted` or `VariantRejected` per build variant
	in the repository, compatible variants first.

	### get_loaded_kernels[[kernels.get_loaded_kernels]]

	#### kernels.get_loaded_kernels[[kernels.get_loaded_kernels]]

	[Source](https://github.com/huggingface/kernels/blob/vr_607/kernels/src/kernels/utils.py#L142)

	Return a snapshot of every kernel that has been loaded into the current process.

	The returned list is a new list; mutating it does not affect the registry.

	Example:
	```python
	from kernels import get_kernel, get_loaded_kernels

	get_kernel("kernels-community/activation", version=1)
	for loaded in get_loaded_kernels():
	print(loaded.metadata.name, loaded.repo_info)
	```

	Returns:

	``list[LoadedKernel]``

	One [LoadedKernel](/docs/kernels/pr_607/en/api/kernels#kernels.LoadedKernel) per distinct kernel variant path
	loaded in this process.

	## Loading locked kernels

	### load_kernel[[kernels.load_kernel]]

	#### kernels.load_kernel[[kernels.load_kernel]]

	[Source](https://github.com/huggingface/kernels/blob/vr_607/kernels/src/kernels/utils.py#L593)

	Get a pre-downloaded, locked kernel.

	If `lockfile` is not specified, the lockfile will be loaded from the caller's package metadata.

	Parameters:

	repo_id (`str`) : The Hub repository containing the kernel.

	lockfile (`Path`, optional) : Path to the lockfile. If not provided, the lockfile will be loaded from the caller's package metadata.

	backend (`str`, optional) : The backend to load the kernel for. Can only be `cpu` or the backend that Torch is compiled for. The backend will be detected automatically if not provided.

	revision (`str`, optional) : The specific revision (branch, tag, or commit) to download. Cannot be used together with `version`.

	Returns:

	``ModuleType``

	The imported kernel module.

	### get_locked_kernel[[kernels.get_locked_kernel]]

	#### kernels.get_locked_kernel[[kernels.get_locked_kernel]]

	[Source](https://github.com/huggingface/kernels/blob/vr_607/kernels/src/kernels/utils.py#L649)

	Get a kernel using a lock file.

	Parameters:

	repo_id (`str`) : The Hub repository containing the kernel.

	local_files_only (`bool`, optional, defaults to `False`) : Whether to only use local files and not download from the Hub.

	Returns:

	``ModuleType``

	The imported kernel module.

	## Classes

	### LoadedKernel[[kernels.LoadedKernel]]

	#### kernels.LoadedKernel[[kernels.LoadedKernel]]

	[Source](https://github.com/huggingface/kernels/blob/vr_607/kernels/src/kernels/utils.py#L112)

	This dataclass provides information about a loaded kernel:

	- `metadata` (`Metadata`): kernel metadata.
	- `module` (`ModuleType`): the imported kernel module.
	- `repo_info` (`kernels.utils.RepoInfo \| None`): populated only for
	kernels loaded via `get_kernel`. Loaders that work from a local path
	(`get_local_kernel`) or a lockfile (`get_locked_kernel`, `load_kernel`)
	leave this as `None`.

	The metadata includes the following properties that describe a kernel:

	- `id` (`str`): kernel identifier that is unique to the kernel version + backend.
	- `name` (`str`): the name of the kernel.
	- `version` (`int`): the version of the kernel.
	- `license` (`str`): the license of the kernel.
	- `upstream` (`str \| None`): the upstream repository of the kernel.
	- `python_depends` (`list[str]`): required Python dependencies.
	- `backend`: information about the kernel's backend.

	### RepoInfo[[kernels.RepoInfo]]

	#### kernels.RepoInfo[[kernels.RepoInfo]]

	[Source](https://github.com/huggingface/kernels/blob/vr_607/kernels/src/kernels/utils.py#L97)

	This dataclass stores the origin of the kernel.

	The following fields are available:

	- `repo_id` (`str`): the Hub repository containing the kernel.
	- `revision` (`str`): the specific revision of the kernel.

	### Layers API Reference
	https://huggingface.co/docs/kernels/pr_607/api/layers.md

	# Layers API Reference

	## Making layers kernel-aware

	### use_kernel_forward_from_hub[[kernels.use_kernel_forward_from_hub]]

	#### kernels.use_kernel_forward_from_hub[[kernels.use_kernel_forward_from_hub]]

	[Source](https://github.com/huggingface/kernels/blob/vr_607/kernels/src/kernels/layer/layer.py#L269)

	Decorator factory that makes a layer extensible using the specified layer name.

	This is a decorator factory that returns a decorator which prepares a layer class to use kernels from the
	Hugging Face Hub.

	Example:
	```python
	import torch
	import torch.nn as nn

	from kernels import use_kernel_forward_from_hub
	from kernels import Mode, kernelize

	@use_kernel_forward_from_hub("MyCustomLayer")
	class MyCustomLayer(nn.Module):
	def __init__(self, hidden_size):
	super().__init__()
	self.hidden_size = hidden_size

	def forward(self, x: torch.Tensor):
	# original implementation
	return x

	model = MyCustomLayer(768)

	# The layer can now be kernelized:
	# model = kernelize(model, mode=Mode.TRAINING \| Mode.TORCH_COMPILE, device="cuda")
	```

	Parameters:

	layer_name (`str`) : The name of the layer to use for kernel lookup in registered mappings.

	Returns:

	``Callable``

	A decorator function that can be applied to layer classes.

	### use_kernel_func_from_hub[[kernels.use_kernel_func_from_hub]]

	#### kernels.use_kernel_func_from_hub[[kernels.use_kernel_func_from_hub]]

	[Source](https://github.com/huggingface/kernels/blob/vr_607/kernels/src/kernels/layer/func.py#L167)

	Decorator that makes a function extensible using the specified function name.

	This is a decorator factory that returns a decorator which prepares a function to use kernels from the
	Hugging Face Hub.

	The function will be exposed as an instance of `torch.nn.Module` in which
	the function is called in `forward`. For the function to be properly
	kernelized, it must be a member of another `torch.nn.Module` that is
	part of the model (see the example).

	Example:
	```python
	import torch
	import torch.nn as nn

	from kernels import use_kernel_func_from_hub
	from kernels import Mode, kernelize

	@use_kernel_func_from_hub("my_custom_func")
	def my_custom_func(x: torch.Tensor):
	# Original implementation
	return x

	class MyModel(torch.nn.Module):
	def __init__(self):
	super().__init__()
	self.fn = my_custom_func

	def forward(self, x):
	return self.fn(x)

	model = MyModel()

	# The layer can now be kernelized:
	# model = kernelize(model, mode=Mode.TRAINING \| Mode.TORCH_COMPILE, device="cuda")
	```

	Parameters:

	func_name (`str`) : The name of the function name to use for kernel lookup in registered mappings.

	Returns:

	``Callable``

	A decorator function that can be applied to layer classes.

	### replace_kernel_forward_from_hub[[kernels.replace_kernel_forward_from_hub]]

	#### kernels.replace_kernel_forward_from_hub[[kernels.replace_kernel_forward_from_hub]]

	[Source](https://github.com/huggingface/kernels/blob/vr_607/kernels/src/kernels/layer/layer.py#L246)

	Function that prepares a layer class to use kernels from the Hugging Face Hub.

	It is recommended to use [use_kernel_forward_from_hub()](/docs/kernels/pr_607/en/api/layers#kernels.use_kernel_forward_from_hub) decorator instead.
	This function should only be used as a last resort to extend third-party layers,
	it is inherently fragile since the member variables and `forward` signature
	of such a layer can change.

	Example:
	```python
	from kernels import replace_kernel_forward_from_hub
	import torch.nn as nn

	replace_kernel_forward_from_hub(nn.LayerNorm, "LayerNorm")
	```

	## Registering kernel mappings

	### use_kernel_mapping[[kernels.use_kernel_mapping]]

	#### kernels.use_kernel_mapping[[kernels.use_kernel_mapping]]

	[Source](https://github.com/huggingface/kernels/blob/vr_607/kernels/src/kernels/layer/kernelize.py#L17)

	Context manager that sets a kernel mapping for the duration of the context.

	This function allows temporary kernel mappings to be applied within a specific context, enabling different
	kernel configurations for different parts of your code.

	Example:
	```python
	import torch
	import torch.nn as nn
	from torch.nn import functional as F

	from kernels import use_kernel_forward_from_hub
	from kernels import use_kernel_mapping, LayerRepository, Device
	from kernels import Mode, kernelize

	# Define a mapping
	mapping = {
	"SiluAndMul": {
	"cuda": LayerRepository(
	repo_id="kernels-community/activation",
	layer_name="SiluAndMul",
	version=1
	)
	}
	}

	@use_kernel_forward_from_hub("SiluAndMul")
	class SiluAndMul(nn.Module):
	def forward(self, x: torch.Tensor) -> torch.Tensor:
	d = x.shape[-1] // 2
	return F.silu(x[..., :d]) * x[..., d:]

	model = SiluAndMul()

	# Use the mapping for the duration of the context.
	with use_kernel_mapping(mapping):
	# kernelize uses the temporary mapping
	model = kernelize(model, mode=Mode.TRAINING \| Mode.TORCH_COMPILE, device="cuda")

	# Outside the context, original mappings are restored
	```

	Parameters:

	mapping (`dict[str, dict[Union[Device, str], Union[LayerRepositoryProtocol, dict[Mode, LayerRepositoryProtocol]]]]`) : The kernel mapping to apply. Maps layer names to device-specific kernel configurations.

	inherit_mapping (`bool`, optional, defaults to `True`) : When `True`, the current mapping will be extended by `mapping` inside the context. When `False`, only `mapping` is used inside the context.

	Returns:

	Context manager that handles the temporary kernel mapping.

	### register_kernel_mapping[[kernels.register_kernel_mapping]]

	#### kernels.register_kernel_mapping[[kernels.register_kernel_mapping]]

	[Source](https://github.com/huggingface/kernels/blob/vr_607/kernels/src/kernels/layer/kernelize.py#L97)

	Register a global mapping between layer names and their corresponding kernel implementations.

	This function allows you to register a mapping between a layer name and the corresponding kernel(s) to use,
	depending on the device and mode. This should be used in conjunction with [kernelize()](/docs/kernels/pr_607/en/api/layers#kernels.kernelize).

	Example:
	```python
	from kernels import LayerRepository, register_kernel_mapping, Mode

	# Simple mapping for a single kernel per device
	kernel_layer_mapping = {
	"LlamaRMSNorm": {
	"cuda": LayerRepository(
	repo_id="kernels-community/layer_norm",
	layer_name="LlamaRMSNorm",
	version=1,
	),
	},
	}
	register_kernel_mapping(kernel_layer_mapping)

	# Advanced mapping with mode-specific kernels
	advanced_mapping = {
	"MultiHeadAttention": {
	"cuda": {
	Mode.TRAINING: LayerRepository(
	repo_id="kernels-community/training-kernels",
	layer_name="TrainingAttention",
	version=1,
	),
	Mode.INFERENCE: LayerRepository(
	repo_id="kernels-community/inference-kernels",
	layer_name="FastAttention",
	version=1,
	),
	}
	}
	}
	register_kernel_mapping(advanced_mapping)
	```

	Parameters:

	mapping (`dict[str, dict[Union[Device, str], Union[RepositoryProtocol, dict[Mode, RepositoryProtocol]]]]`) : The kernel mapping to register globally. Maps layer names to device-specific kernels. The mapping can specify different kernels for different modes (training, inference, etc.).

	inherit_mapping (`bool`, optional, defaults to `True`) : When `True`, the current mapping will be extended by `mapping`. When `False`, the existing mappings are erased before adding `mapping`.

	## Kernelizing a model

	### kernelize[[kernels.kernelize]]

	#### kernels.kernelize[[kernels.kernelize]]

	[Source](https://github.com/huggingface/kernels/blob/vr_607/kernels/src/kernels/layer/kernelize.py#L175)

	Replace layer forward methods with optimized kernel implementations.

	This function iterates over all modules in the model and replaces the `forward` method of extensible layers
	for which kernels are registered using [register_kernel_mapping()](/docs/kernels/pr_607/en/api/layers#kernels.register_kernel_mapping) or [use_kernel_mapping()](/docs/kernels/pr_607/en/api/layers#kernels.use_kernel_mapping).

	Example:
	```python
	import torch
	import torch.nn as nn

	from kernels import kernelize, Mode, use_kernel_mapping, LayerRepository
	from kernels import use_kernel_forward_from_hub

	@use_kernel_forward_from_hub("SiluAndMul")
	class SiluAndMul(nn.Module):
	def forward(self, x: torch.Tensor) -> torch.Tensor:
	d = x.shape[-1] // 2
	return F.silu(x[..., :d]) * x[..., d:]

	mapping = {
	"SiluAndMul": {
	"cuda": LayerRepository(
	repo_id="kernels-community/activation",
	layer_name="SiluAndMul",
	version=1,
	)
	}
	}

	# Create and kernelize a model
	model = nn.Sequential(
	nn.Linear(1024, 2048, device="cuda"),
	SiluAndMul(),
	)

	# Kernelize for inference
	with use_kernel_mapping(mapping):
	kernelized_model = kernelize(model, mode=Mode.TRAINING \| Mode.TORCH_COMPILE)
	```

	Parameters:

	model (`nn.Module`) : The PyTorch model to kernelize.

	mode ([Mode](/docs/kernels/pr_607/en/api/layers#kernels.Mode)) : The mode that the kernel is going to be used in. For example, `Mode.TRAINING \| Mode.TORCH_COMPILE` kernelizes the model for training with `torch.compile`.

	device (`Union[str, torch.device]`, optional) : The device type to load kernels for. Supported device types are: "cuda", "mps", "npu", "rocm", "xpu". The device type will be inferred from the model parameters when not provided.

	use_fallback (`bool`, optional, defaults to `True`) : Whether to use the original forward method of modules when no compatible kernel could be found. If set to `False`, an exception will be raised in such cases.

	Returns:

	``nn.Module``

	The kernelized model with optimized kernel implementations.

	## Classes

	### Device[[kernels.Device]]

	#### kernels.Device[[kernels.Device]]

	[Source](https://github.com/huggingface/kernels/blob/vr_607/kernels/src/kernels/layer/device.py#L106)

	Represents a compute device with optional properties.

	This class encapsulates device information including device type and optional device-specific properties
	like CUDA capabilities.

	Example:
	```python
	from kernels import Device, CUDAProperties

	# Basic CUDA device
	cuda_device = Device(type="cuda")

	# CUDA device with specific capability requirements
	cuda_device_with_props = Device(
	type="cuda",
	properties=CUDAProperties(min_capability=75, max_capability=90)
	)

	# MPS device for Apple Silicon
	mps_device = Device(type="mps")

	# XPU device (e.g., Intel(R) Data Center GPU Max 1550)
	xpu_device = Device(type="xpu")

	# NPU device (Huawei Ascend)
	npu_device = Device(type="npu")
	```

	validatekernels.Device.validatehttps://github.com/huggingface/kernels/blob/vr_607/kernels/src/huggingface_hub/dataclasses.py#L247[]
	Run class validators on the instance.

	Parameters:

	type (`str`) : The device type (e.g., "cuda", "mps", "npu", "rocm", "xpu").

	properties (`CUDAProperties`, optional) : Device-specific properties. Currently only `CUDAProperties` is supported for CUDA devices.

	### Mode[[kernels.Mode]]

	#### kernels.Mode[[kernels.Mode]]

	[Source](https://github.com/huggingface/kernels/blob/vr_607/kernels/src/kernels/layer/mode.py#L4)

	Kernelize mode

	The `Mode` flag is used by [kernelize()](/docs/kernels/pr_607/en/api/layers#kernels.kernelize) to select kernels for the given mode. Mappings can be registered for
	specific modes.

	Note:
	Different modes can be combined. For instance, `INFERENCE \| TORCH_COMPILE` should be used for layers that
	are used for inference with `torch.compile`.

	Parameters:

	INFERENCE : The kernel is used for inference.

	TRAINING : The kernel is used for training.

	TORCH_COMPILE : The kernel is used with `torch.compile`.

	FALLBACK : In a kernel mapping, this kernel is used when no other mode matches.

	### FuncRepository[[kernels.FuncRepository]]

	#### kernels.FuncRepository[[kernels.FuncRepository]]

	[Source](https://github.com/huggingface/kernels/blob/vr_607/kernels/src/kernels/layer/func.py#L27)

	Repository and name of a function for kernel mapping.

	Example:
	```python
	from kernels import FuncRepository

	# Reference a specific layer by revision
	layer_repo = FuncRepository(
	repo_id="kernels-community/activation",
	func_name="silu_and_mul",
	revision="main",
	)

	# Reference a layer by version
	layer_repo_versioned = FuncRepository(
	repo_id="kernels-community/relu",
	func_name="relu",
	version=1
	)
	```

	Parameters:

	repo_id (`str`) : The Hub repository containing the layer.

	func_name (`str`) : The name of the function within the kernel repository.

	revision (`str`, optional) : The specific revision (branch, tag, or commit) to download. Cannot be used together with `version`.

	version (`int`, optional) : The kernel version to download. Cannot be used together with `revision`. Either `version` or `revision` must be specified.

	### LayerRepository[[kernels.LayerRepository]]

	#### kernels.LayerRepository[[kernels.LayerRepository]]

	[Source](https://github.com/huggingface/kernels/blob/vr_607/kernels/src/kernels/layer/layer.py#L32)

	Repository and name of a layer for kernel mapping.

	Example:
	```python
	from kernels import LayerRepository

	# Reference a specific layer by version
	layer_repo = LayerRepository(
	repo_id="kernels-community/activation",
	layer_name="SiluAndMul",
	version=1,
	)
	```

	Parameters:

	repo_id (`str`) : The Hub repository containing the layer.

	layer_name (`str`) : The name of the layer within the kernel repository.

	revision (`str`, optional) : The specific revision (branch, tag, or commit) to download. Cannot be used together with `version`.

	version (`int`, optional) : The kernel version to download. Cannot be used together with `revision`. Either `version` or `revision` must be specified.

	trust_remote_code (`bool \| list[str]`, optional, defaults to `False`) : Whether to allow loading kernels from untrusted organisations. A list of signing identities can be provided for future verification support; until then it warns and falls back to the default trust check.

	### LocalFuncRepository[[kernels.LocalFuncRepository]]

	#### kernels.LocalFuncRepository[[kernels.LocalFuncRepository]]

	[Source](https://github.com/huggingface/kernels/blob/vr_607/kernels/src/kernels/layer/func.py#L116)

	Repository and function name from a local directory for kernel mapping.

	Example:
	```python
	from pathlib import Path

	from kernels import LocalFuncRepository

	# Reference a specific layer by revision
	layer_repo = LocalFuncRepository(
	repo_path=Path("/home/daniel/kernels/activation"),
	func_name="silu_and_mul",
	)
	```

	Parameters:

	repo_path (`Path`) : The local repository containing the layer.

	func_name (`str`) : The name of the function within the kernel repository.

	### LocalLayerRepository[[kernels.LocalLayerRepository]]

	#### kernels.LocalLayerRepository[[kernels.LocalLayerRepository]]

	[Source](https://github.com/huggingface/kernels/blob/vr_607/kernels/src/kernels/layer/layer.py#L128)

	Repository from a local directory for kernel mapping.

	Example:
	```python
	from pathlib import Path

	from kernels import LocalLayerRepository

	# Reference a specific layer by revision
	layer_repo = LocalLayerRepository(
	repo_path=Path("/home/daniel/kernels/activation"),
	layer_name="SiluAndMul",
	)
	```

	Parameters:

	repo_path (`Path`) : The local repository containing the layer.

	layer_name (`str`) : The name of the layer within the kernel repository.

	### LockedFuncRepository[[kernels.LockedFuncRepository]]

	#### kernels.LockedFuncRepository[[kernels.LockedFuncRepository]]

	[Source](https://github.com/huggingface/kernels/blob/vr_607/kernels/src/kernels/layer/func.py#L222)

	Repository and name of a function.

	In contrast to `FuncRepository`, this class uses repositories that
	are locked inside a project.

	### LockedLayerRepository[[kernels.LockedLayerRepository]]

	#### kernels.LockedLayerRepository[[kernels.LockedLayerRepository]]

	[Source](https://github.com/huggingface/kernels/blob/vr_607/kernels/src/kernels/layer/layer.py#L179)

	Repository and name of a layer.

	In contrast to `LayerRepository`, this class uses repositories that
	are locked inside a project.

	### IDE setup with direnv and the kernel devshell
	https://huggingface.co/docs/kernels/pr_607/builder/ide-setup.md

	# IDE setup with direnv and the kernel devshell

	## Introduction

	Language servers do not interpret `build.toml`, so IDE completion for
	CUDA, ROCm, framework headers, and the kernel's Python wrapper does not
	work out of the box. This guide shows how to configure VS Code so that
	completion resolves against the same toolchain `kernel-builder`
	uses.

	The setup has three pieces:

	- `kernel-builder create-pyproject` to emit CMake and setuptools files
	the IDE can read (see [Local Development](./local-dev)).
	- The kernel-builder devshell, which provides the toolchain (CUDA, ROCm,
	Torch headers, etc.) from the Nix store.
	- `direnv` to activate the devshell on `cd`, so VS Code inherits the
	environment through the shell.

	Pinning the toolchain through Nix keeps IDE completion aligned with
	the build. It also makes switching between CUDA, ROCm, or XPU a
	one-line change in `.envrc`.

	## Installing direnv and nix-direnv

	On non-NixOS systems, install both via `nix profile`:

	```bash
	$ nix profile install nixpkgs#nix-direnv
	```

	Add the direnv hook to your shell rc (`~/.bashrc` or
	`~/.zshrc`, for example):

	```bash
	eval "$(direnv hook bash)" # or: direnv hook zsh
	```

	Source the rc file (or open a new shell) so the hook is
	active in the current session:

	```bash
	$ source ~/.bashrc # or: source ~/.zshrc
	```

	Wire `nix-direnv` into direnv:

	```bash
	$ mkdir -p ~/.config/direnv
	$ echo 'source $HOME/.nix-profile/share/nix-direnv/direnvrc' \
	>> ~/.config/direnv/direnvrc
	```

	On [NixOS](https://github.com/nix-community/nix-direnv#via-system-configuration-on-nixos)
	or with [home-manager](https://github.com/nix-community/nix-direnv#via-home-manager),
	enable `programs.direnv` with
	`nix-direnv` instead. See
	[`terraform/nixos-configuration.nix`](https://github.com/huggingface/kernels/tree/main/terraform/nixos-configuration.nix)
	for a working example.

	## Activating the devshell with direnv

	From the kernel root directory (the one containing `flake.nix` and
	`build.toml`), tell direnv to use the flake's default devshell:

	```bash
	$ echo 'use flake' > .envrc
	$ direnv allow
	```

	`direnv` now activates the default devshell whenever you `cd` into the
	project. The devshell's `shellHook` creates and activates a `.venv` on
	first entry. Confirm it picked up the toolchain and venv:

	```bash
	$ which nvcc
	/nix/store/.../bin/nvcc
	$ ls -ld .venv
	drwxr-xr-x ... .venv
	$ which python
	/path/to/kernel/.venv/bin/python
	```

	If `.venv` is missing, re-run `direnv reload` and check the output for
	the `Creating new venv environment in path: './.venv'` line from the
	`shellHook`.

	To pin a non-default build variant, name it explicitly:

	```bash
	$ echo 'use flake .#devShells.torch211-cxx11-rocm71-x86_64-linux' > .envrc
	$ direnv allow
	```

	See [Build Variants](./build-variants) for the variant list.

	## Generating IDE-facing project files

	direnv puts the toolchain on `PATH`, but the C++ language server still
	needs a CMake-derived `compile_commands.json` to resolve per-file
	include paths. Generate the CMake/setuptools project and the file:

	```bash
	$ kernel-builder create-pyproject -f
	$ cmake -B build-ext -DCMAKE_EXPORT_COMPILE_COMMANDS=ON
	$ ln -sf build-ext/compile_commands.json compile_commands.json
	```

	`-DCMAKE_EXPORT_COMPILE_COMMANDS=ON` is required: the generated CMake
	does not set it. The symlink lets the language server find the file
	at the project root.

	As noted in [Local Development](./local-dev), do not commit the
	generated files.

	## Configuring VS Code

	Install the [`mkhl.direnv`](https://github.com/direnv/direnv-vscode)
	extension. It activates the project's `.envrc` when VS Code opens
	the workspace, so language servers and the integrated terminal see
	the devshell environment without launching `code` from a shell.

	Alternatively, skip the extension and open the project from a
	direnv-activated shell — VS Code inherits the environment that way
	too:

	```bash
	$ cd path/to/kernel
	$ code .
	```

	Install one of the following first-party extensions for C++/CUDA
	completion:

	- `llvm-vs-code-extensions.vscode-clangd` (recommended for CUDA).
	- `ms-vscode.cpptools` (Microsoft C/C++).

	Add `.vscode/settings.json` (do not commit):

	```jsonc
	{
	"python.defaultInterpreterPath": "${workspaceFolder}/.venv/bin/python",

	// clangd
	"clangd.arguments": ["--compile-commands-dir=${workspaceFolder}"],

	// Microsoft C/C++ extension
	"C_Cpp.default.compileCommands": "${workspaceFolder}/compile_commands.json"
	}
	```

	Depending on the extension being used, the configuration above behaves
	differently:

	- With `clangd`, the `clangd.arguments` line is optional. clangd already
	looks in the parent directories of each source file for
	`compile_commands.json` and will find the workspace-root symlink on its
	own ([clangd docs](https://clangd.llvm.org/installation#project-setup)).
	Setting it explicitly does no harm.
	- With the Microsoft C/C++ extension, the `C_Cpp.default.compileCommands`
	line is required. The extension does not pick up
	`compile_commands.json` from the workspace root on its own, unless
	another extension (such as CMake Tools) tells it where to look.

	To verify, open `torch-ext/torch_binding.cpp` and hover an
	`#include <torch/torch.h>` directive. The resolved path should point
	into `/nix/store/...`, not a system path.

	## Remote development

	Use the VS Code Remote-SSH extension and put the direnv hook in the
	remote shell's rc. The remote integrated terminal activates the
	devshell on `cd`, and VS Code's language servers — which run on the
	remote — inherit that environment. The
	[`terraform/`](https://github.com/huggingface/kernels/tree/main/terraform)
	setup is already configured this way.

	## Switching toolchains

	Change the `use flake` line in `.envrc` to point at a different
	variant. For example:

	```bash
	# CUDA 13.0
	use flake .#devShells.torch211-cxx11-cu130-x86_64-linux

	# ROCm 7.1
	use flake .#devShells.torch211-cxx11-rocm71-x86_64-linux

	# XPU
	use flake .#devShells.torch211-cxx11-xpu20253-x86_64-linux
	```

	Remove `.venv/` first if it was created against a different variant,
	then reload direnv to recreate it via the new devshell's `shellHook`:

	```bash
	$ rm -rf .venv
	$ direnv reload
	```

	## noarch kernels

	For Python-only (noarch) kernels, skip the CMake step in "Generating
	IDE-facing project files" and the C++ portions of the VS Code
	configuration. The `direnv` setup and `python.defaultInterpreterPath`
	are all that is needed.

	### Writing custom kernels with code agents
	https://huggingface.co/docs/kernels/pr_607/builder/agents-guide.md

	# Writing custom kernels with code agents

	Code agents are a good fit to build custom kernels because the hard part is not just writing in Domain Specific Language (DSLs) like CUDA. You also need the right project layout, PyTorch bindings, architecture-specific choices, model-specific integration, and trustworthy benchmarks.

	Kernels on Hugging Face are compatible with agents via skills and the `hf` CLI. The `cuda-kernels`, `rocm-kernels`, and `xpu-kernels` skills contain knowledge so an agent can generate and publish a complete kernel project, instead of isolated snippets.

	This guide is for authoring new kernels. If you only want to load an existing precompiled kernel, use `get_kernel()` instead.

	## Before you start

	You need:

	- a coding agent that supports skills, such as Claude Code, Codex, Cursor, or OpenCode
	- a clear target: library, model, operation, GPU, dtype, and representative shapes

	The skill currently focuses on NVIDIA GPUs such as H100, A100, and T4, and on integration patterns for transformers and diffusers.

	Install the skill into your agent. If you need the latest version from `main`, use:

	```shell
	cargo install --git https://github.com/huggingface/kernels hf-kernel-builder

	# Install your skills. Use --claude, --codex, or --opencode
	kernel-builder skills add --claude
	```

	> [!NOTE]
	> Check [this example](https://github.com/burtenshaw/kernel-skill/tree/main/examples/ltx_video) to see what generated kernels look like.

	## 1. Give the agent a precise task prompt

	Writing kernels is a hard problem, so be specific to agents. A robust prompt will declare all core attributes, including:

	- the library, for example `transformers` or `diffusers`
	- the model id, for example `Qwen3-8B` or `LTX-Video`
	- the operation, for example `RMSNorm`, attention, RoPE, `GEGLU`, or `AdaLN`
	- the target GPU, for example `H100`, `A100`, or `T4`
	- the dtype, for example `bfloat16`, `float16`, or `float32`
	- the outputs you expect: kernel code, bindings, tests, and benchmarks

	In practice, you can often skip some of these and the agent will infer based on common practice, but if you know a detail declare it.

	For example:

	```
	Build a vectorized RMSNorm kernel for H100 targeting Qwen3-8B in transformers.
	Create the full kernel-builder project, PyTorch bindings, correctness tests, and benchmark scripts.
	```

	Or for diffusers:

	```
	Build an H100 RMSNorm kernel for LTX-Video in diffusers.
	Patch the pipeline correctly, benchmark it against the PyTorch baseline, and report end-to-end impact.
	```

	If you prefer, you can first scaffold a project with `kernel-builder init --name <org>/<kernel>` and then ask the agent to fill in the implementation.

	## 2. Verify that the agent produces a complete kernel project

	A useful result is a full `kernel-builder` project, not just a `.cu` file. The exact layout can vary, but it should include at least:

	```
	examples/your_model/
	├── kernel_src/
	│ └── rmsnorm.cu # Vectorized CUDA kernel
	├── torch-ext/
	│ ├── your_kernels/__init__.py
	│ └── torch_binding.cpp # PyTorch C++ bindings
	├── benchmark_rmsnorm.py # Micro-benchmark script
	├── build.toml # kernel-builder config
	├── setup.py # pip install -e .
	└── pyproject.toml
	```

	The agent skills contain example scipts to help you verify the project. So you can briefly test it yourself by running:

	```
	Verify the kernel project works with a transformers example.
	```

	## 3. Review the generated files

	Let's dive deeper into the generated files, and explore how to validate the project.

	### `build.toml`

	This is the main configuration file for `kernel-builder`. It tells `kernel-builder` what to build and how so it should contain all the core information about your kernel project.

	```
	[general]
	name = "your_kernels"
	backends = ["cuda"]
	version = 1

	[torch]
	src = ["torch-ext/torch_binding.cpp"]

	[kernel.rmsnorm]
	backend = "cuda"
	src = ["kernel_src/rmsnorm.cu"]
	depends = ["torch"]
	cuda-capabilities = ["9.0"] # H100
	```

	First check that:

	- `backends = ["cuda"]` is correct for your project
	- the kernel source files are listed correctly
	- the Torch binding sources are included under `[torch]`
	- `cuda-capabilities` is only set when the kernel truly targets specific architectures

	For architecture-specific kernels, typical capability values are:

	- H100: `9.0`
	- A100: `8.0`
	- T4: `7.5`

	If the kernel does not require a specific capability, the kernels docs recommend leaving `cuda-capabilities` unset so the builder can target all supported capabilities. In practice, you can prompt your agent to review the `build.toml` for excessive definitions. Agents have a tendency to over-specify capabilities.

	### Torch bindings

	The kernel should be registered as Torch ops in `torch-ext/torch_binding.cpp`, with declarations in a header and a small Python wrapper in `torch-ext/<name>/__init__.py`. This is what makes the kernel callable from Python and is the right foundation for `torch.compile` compatibility.

	### Model integration code

	Make sure the integration matches the library:

	- transformers: patch the target modules directly, often RMSNorm modules whose class name contains `RMSNorm`
	- diffusers: inspect the actual pipeline structure before patching, because modules and attention processors can differ across pipelines

	> [!NOTE]
	> One common issue is that the agent will not integrate the kernel at all. Typically because the project's context is so long.

	A few patterns matter in practice for the integration code:

	- In transformers, RMSNorm modules generally have weights, but epsilon may be exposed as `variance_epsilon` or `eps` depending on the model.
	- In diffusers, some RMSNorm modules may have `weight=None`, so the integration code needs to handle both weighted and unweighted cases.
	- In diffusers, checking `type(module).__name__` is often more reliable than `isinstance(...)` for matching RMSNorm modules across implementations.
	- If a diffusers pipeline uses CPU offloading, inject custom kernels before enabling offload.

	For attention, prefer the model library's existing optimized path when one already exists. For example, in `transformers`, Flash Attention 2 is usually the right baseline for attention, while custom kernels are especially useful for operations like RMSNorm and other targeted hotspots.

	## 5. Build and test, and benchmark

	Kernel Hub kernels must support all recent PyTorch and CUDA configurations. The kernel-builder Nix flake handles this automatically. Copy the [example `flake.nix`](https://github.com/huggingface/kernels/blob/main/builder/examples/relu/flake.nix) into your project and run:

	```shell
	nix flake update
	nix run .#build-and-copy -L
	```

	This builds the kernel for every required PyTorch/CUDA variant and places the results in `build/`. For faster builds, enable the HuggingFace Nix cache:

	```shell
	nix run nixpkgs#cachix -- use huggingface
	```

	## 6. Benchmark

	There are two main benchmarks to consider:

	1. an isolated kernel micro-benchmark
	2. an end-to-end benchmark in the real model or pipeline

	The agent will generate both benchmarks based on the agent skills examples. Typically as a script called `benchmark_example.py`. If you have access to the target hardware, you can run it to verify the kernel works. For example, the agent will generat a table like this:

	```markdown
	\| Shape \| Custom (ms) \| PyTorch (ms) \| Speedup \|
	\| :---- \| :---: \| :---: \| :---: \|
	\| [1x128x4096] \| 0.040 \| 0.062 \| 1.58x \|
	\| [1x512x4096] \| 0.038 \| 0.064 \| 1.69x \|
	\| [1x1024x4096] \| 0.037 \| 0.071 \| 1.90x \|
	\| [1x2048x4096] \| 0.045 \| 0.091 \| 2.03x \|
	\| [1x4096x4096] \| 0.071 \| 0.150 \| 2.12x \|
	\| [4x512x4096] \| 0.056 \| 0.093 \| 1.67x \|
	\| [8x256x4096] \| 0.045 \| 0.092 \| 2.06x \|
	\| [1x8192x4096] \| 0.109 \| 0.269 \| 2.47x \|
	```

	Interpret the results carefully. A kernel can show a large isolated speedup but only a modest end-to-end gain if that operation is a small fraction of total runtime. In the LTX-Video example from [the blog we wrote](https://huggingface.co/blog/custom-cuda-kernels-agent-skills), the generated RMSNorm kernel improved the isolated benchmark by about 1.88x on average, but end-to-end video generation improved by about 6%, which matched the fact that RMSNorm accounted for only a small share of total compute.

	## 7. Publish to the Hub

	Once the project is correct and benchmarked, you can build Hub-compatible artifacts and upload them. For this, you should first push to the Hub using the `hf` CLI tool:

	```shell
	# install the hf CLI tool
	hf skills add

	# Authenticate
	hf auth login

	# Push to the Hub
	<agent-prompt>
	Push the kernel to the Hub.
	</agent-prompt>
	```

	Or, you can manually create the repository and upload the artifacts:

	```shell
	# Create the repository
	hf repo create your-org/your-kernel --type model

	# Upload the artifacts
	# Run inside the main kernel directory, where build/ is.
	kernel-builder upload
	```

	After pushing to the Hub, users can load the kernel without compiling:

	```py
	from kernels import get_kernel

	kernel = get_kernel("your-org/your-kernel", version=1)
	```

	Well done! You have now built a custom kernel and published it to the Hub.

	### Writing Hub kernels with kernel-builder
	https://huggingface.co/docs/kernels/pr_607/builder/writing-kernels.md

	# Writing Hub kernels with kernel-builder

	## Introduction

	The Kernel Hub allows Python libraries and applications to load compute
	kernels directly from the [Hub](https://hf.co/). To support this kind
	of dynamic loading, Hub kernels differ from traditional Python kernel
	packages in that they are made to be:

	- Portable: a kernel can be loaded from paths outside `PYTHONPATH`.
	- Unique: multiple versions of the same kernel can be loaded in the
	same Python process.
	- Compatible: kernels must support all recent versions of Python and
	the different PyTorch build configurations (various CUDA versions
	and C++ ABIs). Furthermore, older C library versions must be supported.

	`kernel-builder` is a set of tools that can build conforming kernels. It
	takes care of:

	- Building kernels for all supported PyTorch configurations (C++98/11 and
	different CUDA versions).
	- Compatibility with old glibc and libstdc++ versions, so that kernels also
	work on older Linux distributions.
	- Registering Torch ops, such that multiple versions the same kernel can be
	loaded without namespace conflicts.

	`kernel-builder` builds are configured through a `build.toml` file.
	`build.toml` is a simple format that does not require intricate knowledge
	of CMake or setuptools.

	This page describes the directory layout of a kernel-builder project, the
	format of the `build.toml` file, and some additional Python glue that
	`kernel-builder` provides. We will use a [simple ReLU kernel](https://github.com/huggingface/kernels/tree/main/examples/kernels/relu)
	as the running example. After reading this page, you may also want to have
	a look at the more realistic [ReLU kernel with backprop and `torch.compile`](https://github.com/huggingface/kernels/tree/main/examples/kernels/relu-backprop-compile)
	support.

	> [!TIP]
	> We maintain a set of conforming kernels in the
	> [kernels-community repository](https://github.com/huggingface/kernels-community).
	> We try to keep these kernels synced with upstream as much as possible.

	## Setting up environment

	### Quick install

	The fastest way to get started is to run the install script. This
	installs [Determinate Nix](https://docs.determinate.systems/determinate-nix/)
	and `kernel-builder` in a single command:

	```bash
	curl -fsSL https://raw.githubusercontent.com/huggingface/kernels/main/install.sh \| bash
	```

	This will:

	1. Install Determinate Nix (if not already installed).
	2. Configure the Hugging Face binary cache (to avoid building dependencies from
	source).
	3. Install `kernel-builder` via `nix profile install`.

	To update `kernel-builder` later:

	```bash
	nix profile upgrade --all
	```

	For a step-by-step breakdown of what the script does, see
	[Using the kernel builder with Nix](nix).

	### Cloud environment

	In the [`terraform`](https://github.com/huggingface/kernels/tree/main/terraform) directory, we provide an
	example of programatically spinning up an EC2 instance that is ready
	with everything needed for you to start developing and building
	kernels.

	If you use a different provider, the Terraform bridges should be
	similar and straightforward to modify.

	## Starting a new kernel

	The easiest way to start a new kernel is by using the `init` subcommand
	of `kernel-builder`. This creates a minimal, compilable kernel:

	```bash
	$ kernel-builder init --name myorg/mykernel
	Initialized `myorg/mykernel` at /home/daniel/git/kernels/examples/kernels/mykernel
	```

	This creates a kernel named `mykernel` in the directory `mykernel`. The
	kernel is configured to upload to the `myorg/mykernel` Hub
	repository when an upload command is used.

	By default, the `init` subcommand creates a CUDA kernel. You can specify
	another backend with the `--backends` option:

	```bash
	$ kernel-builder init --name myorg/mykernel --backends xpu
	```

	You can also make a multi-backend kernel by adding all the backends
	that you would like to support as arguments to `--backends`:

	```bash
	$ kernel-builder init --name myorg/mykernel --backends cuda xpu
	Initialized `myorg/mykernel` at /home/daniel/git/kernels/examples/kernels/mykernel
	```

	Finally, if you want to create a kernel for all supported backends, you
	can use `--backends all`.

	## Kernel project layout

	Kernel projects follow this general directory layout:

	```text
	mykernel
	├── benchmarks
	│ └── benchmark.py
	├── build.toml
	├── CARD.md
	├── example.py
	├── flake.nix
	├── mykernel_cuda
	│ └── mykernel.cu
	├── tests
	│ ├── __init__.py
	│ └── test_mykernel.py
	└── torch-ext
	├── mykernel
	│ └── __init__.py
	├── torch_binding.cpp
	└── torch_binding.h
	```

	In this example we can find:

	- The build configuration in `build.toml`.
	- One or more top-level directories containing kernels (`mykernel_cuda`).
	- The `torch-ext` directory, which contains:
	- `torch_binding.h`: contains declarations for kernel entry points
	(from `kernel_a` and `kernel_b`).
	- `torch_binding.cpp`: registers the entry points as Torch ops.
	- `torch_ext/mykernel`: contains any Python wrapping the kernel needs. At the
	bare minimum, it should contain an `__init__.py` file.
	- Kernel tests in the directory `tests`.
	- Benchmarks in the directory `benchmarks`.
	- A kernel card template in `CARD.md`. This placeholders in the card are filled
	during the kernel build.
	- The Nix flake configuration in `flake.nix`.
	- An example script that uses the kernel in `example.py`.

	## `build.toml`

	`build.toml` tells `kernel-builder` what to build and how. It looks as
	follows for the `mykernel` kernel:

	```toml
	[general]
	backends = [
	"cuda",
	]
	name = "mykernel"
	version = 1

	[general.hub]
	repo-id = "myorg/mykernel"

	[torch]
	src = [
	"torch-ext/torch_binding.cpp",
	"torch-ext/torch_binding.h",
	]

	[kernel.mykernel]
	backend = "cuda"
	depends = ["torch"]
	src = ["mykernel_cuda/mykernel.cu"]
	# If the kernel is only supported on specific capabilities, set the
	# cuda-capabilities option:
	#
	# cuda-capabilities = [ "9.0", "10.0", "12.0" ]
	```

	The following sections enumerate all supported options for `build.toml`.

	### `general`

	- `name` (required): the name of the kernel. The Python code for a Torch
	extension must be stored in `torch-ext/<name>`.
	- `version` (int): the major version of the kernel.
	The version is written to the kernel's `metadata.json` and is used
	by the `kernels upload` command to upload the kernel to a version
	branch named `v<version>`.
	- `backends` (required): a list of supported backends. Must be one or
	more of `cpu`, `cuda`, `metal`, `rocm`, or `xpu`.
	- `python-depends` (experimental): a list of additional Python dependencies
	that the kernel requires. The only supported dependencies are `einops`
	and `nvidia-cutlass-dsl`.

	### `general.hub`

	- `repo-id`: the Hub repository to upload the kernel to when the `upload` or
	`build-and-upload` subcommands of `kernel-builder` are used.

	### `general.cuda`

	- `maxver`: the maximum CUDA toolkit version (inclusive). This option
	_must not_ be set under normal circumstances, since it can exclude Torch
	build variants that are [required for compliant kernels](../kernel-requirements).
	This option is provided for kernels that cause compiler errors on
	newer CUDA toolkit versions.
	- `minver`: the minimum required CUDA toolkit version. This option
	_must not_ be set under normal circumstances, since it can exclude Torch
	build variants that are [required for compliant kernels](../kernel-requirements).
	This option is provided for kernels that require functionality only
	provided by newer CUDA toolkits.

	### `torch`

	This section describes the Torch extension. In the future, there may be
	similar sections for other frameworks. This section has the following
	options:

	- `src` (required): a list of source files and headers.
	- `pyext` (optional): the list of extensions for Python files. Default:
	`["py", "pyi"]`.
	- `include` (optional): include directories relative to the project root.
	Default: `[]`.
	- `maxver` (optional): only build for this Torch version and earlier. Use cautiously, since this option produces
	non-compliant kernels if the version range does not correspond to the [required variants](build-variants).
	- `minver` (optional): only build for this Torch version and later. Use cautiously, since this option produces
	non-compliant kernels if the version range does not correspond to the [required variants](build-variants).
	- `stable-abi` (experimental): when set to a Torch version (e.g.
	`"2.11"`), the kernel is built using the Torch stable ABI. This
	requires that the kernel itself only use
	[stable ABI headers](https://docs.pytorch.org/docs/2.12/notes/libtorch_stable_abi.html).
	For an example, see the [`relu-torch-stable-abi`](https://github.com/huggingface/kernels/tree/main/examples/kernels/relu-torch-stable-abi)
	example kernel.

	### `kernel.<name>`

	Specification of a kernel with the name `<name>`. Multiple `kernel.<name>`
	sections can be defined in the same `build.toml`.
	See for example [`kernels-community/quantization`](https://huggingface.co/kernels-community/quantization/)
	for an example with multiple kernel sections.

	The following options can be set for a kernel:

	- `backend` (required): the compute backend of the kernel. The currently
	supported backends are `cpu`, `cuda`, `metal`, `rocm`, and `xpu`.
	The `cpu` backend is currently experimental and might still change.
	- `depends` (required): a list of dependencies. The supported dependencies
	are listed in [`deps.nix`](https://github.com/huggingface/kernels/blob/main/builder/lib/deps.nix).
	- `src` (required): a list of source files and headers.
	- `include` (optional): include directories relative to the project root.
	Default: `[]`.

	Besides these shared options, the following backend-specific options
	are available:

	#### cuda

	- `cuda-capabilities` (optional): a list of CUDA capabilities that the
	kernel should be compiled for. When absent, the kernel will be built
	using all capabilities that the builder supports. The effective
	capabilities are the intersection of this list and the capabilities
	supported by the CUDA compiler. It is recommended to leave this option
	unspecified unless a kernel requires specific capabilities.
	- `cuda-flags` (optional): additional flags to be passed to `nvcc`.
	Warning: this option should only be used in exceptional circumstances.
	Custom compile flags can interfere with the build process or break
	compatibility requirements.

	#### rocm

	- `rocm-archs`: a list of ROCm architectures that the kernel should be
	compiled for.

	#### xpu

	- `sycl-flags`: a list of additional flags to be passed to the SYCL
	compiler.

	### cpu

	- `cxx-flags`: a list of additional flags to be passed to the C++
	compiler.

	## Torch bindings

	### Defining bindings

	Torch bindings are defined in C++, kernels commonly use two files:

	- `torch_binding.h` containing function declarations.
	- `torch_binding.cpp` registering the functions as Torch ops.

	For instance, the `mykernel` kernel discussed above has the following
	declaration in `torch_binding.h`:

	```cpp
	#pragma once

	#include <torch/torch.h>

	void mykernel(torch::Tensor &out, torch::Tensor const &input);
	```

	This function is then registered as a Torch op in `torch_binding.cpp`:

	```cpp
	#include <torch/library.h>

	#include "registration.h"
	#include "torch_binding.h"

	TORCH_LIBRARY_EXPAND(TORCH_EXTENSION_NAME, ops) {
	ops.def("mykernel(Tensor! out, Tensor input) -> ()");
	#if defined(CUDA_KERNEL) \|\| defined(ROCM_KERNEL)
	ops.impl("mykernel", torch::kCUDA, &mykernel);
	#endif
	}

	REGISTER_EXTENSION(TORCH_EXTENSION_NAME)
	```

	This snippet uses macros from `registration.h` to register the function.
	`registration.h` is generated by `kernel-builder` itself. A function
	is registered through the `def`/`ops` methods. `ops` specifies the
	function signature following the [function schema](https://github.com/pytorch/pytorch/blob/main/aten/src/ATen/native/README.md#func).
	`impl` associates the function name with the C/C++ function and
	the applicable device.

	## Using kernel functions from Python

	The bindings are typically wrapped in Python code in `torch_ext/<name>`.
	The native code is exposed under the `torch.ops` namespace. However,
	we add some unique material to the name of the extension to ensure that
	different versions of the same extension can be loaded at the same time.
	As a result, the extension is registered as
	`torch.ops.<name>_<unique_material>`.

	To deal with this uniqueness, `kernel_builder` generates a Python module
	named `_ops` that contains an alias for the name. This can be used to
	refer to the correct `torch.ops` module. For example:

	```python
	from typing import Optional

	import torch

	from ._ops import ops

	def mykernel(x: torch.Tensor, out: Optional[torch.Tensor] = None) -> torch.Tensor:
	if out is None:
	out = torch.empty_like(x)
	ops.mykernel(out, x)
	return out
	```

	## Registering Torch operators

	You may want to register Torch ops from your kernel's Python code or
	register fake ops for `torch.compile` support. It is important to register
	such ops in the namespace that kernel-builder makes for your kernel
	build. This is required for compliant kernels to ensure that multiple
	versions of the same kernel can be loaded at the same time without
	namespace conflicts.

	You can use the `add_op_namespace_prefix` to prefix an op name with the
	correct prefix. So for instance, replace

	```python
	@torch.library.register_fake("relu::relu_fwd")
	def relu_fwd_fake(input: torch.Tensor) -> torch.Tensor:
	return torch.empty_like(input)
	```

	by

	```python
	from ._ops import add_op_namespace_prefix

	@torch.library.register_fake(add_op_namespace_prefix("relu_fwd"))
	def relu_fwd_fake(input: torch.Tensor) -> torch.Tensor:
	return torch.empty_like(input)
	```

	As mentioned in the above, the `_ops` module is generated by kernel-builder.

	kernel-builder uses a hook to reject incorrect usage of Torch op registration
	functions. However, it can only catch direct use of certain `torch.library`
	decorators. For instance, the hook would not reject the following decorator,
	so it should be seen as a last-resort check if human review failed:

	```python
	@some_indirection_for_register_fake("relu::relu_fwd")
	def relu_fwd_fake(input: torch.Tensor) -> torch.Tensor:
	return torch.empty_like(input)
	```

	## Kernel tests

	Kernel tests are stored in the `tests` directory. Since running all
	kernel tests in CI may be prohibitively expensive, the `pyproject.toml`
	generated by the builder adds support for the special `kernels_ci`
	PyTest marker that can be used as follows:

	```python
	import pytest

	@pytest.mark.kernels_ci
	def test_mykernel():
	...
	```

	We recommend that you to pick tests that together would catch most error
	cases while running within 60 seconds.

	You can run the tests (e.g. in CI) using:

	```bash
	$ nix run .#ci-test
	```

	If the kernel supports multiple backends, it will run the test for the
	first supported backend that was found, obeying the following order: CUDA,
	ROCm, XPU, Metal, CPU. If you would like to the tests for a specific build
	variant, you can use `nix run .#ciTests.<variant>`. For instance:

	```bash
	$ nix run .#ciTests.torch210-cxx11-cpu-x86_64-linux
	```

	When running the tests on a non-NixOS systems, make sure that
	[the CUDA driver library can be found](https://danieldk.eu/Software/Nix/Nix-CUDA-on-non-NixOS-systems#solutions).

	## Kernel docs

	We provide a utility to generate a system card for a given kernel, utilizing
	information from its `build.toml` and metadata. This system card provides a
	reasonable starting point and is meant to be edited afterward by the kernel
	developer.

	The template card is generated as a part of `kernel-builder init`
	command and is serialized in the root directory of the kernel.

	The card will be filled automatically by the builder when using the
	`build-and-upload` or `build-and-copy` command. It will be serialized
	to the `build` sub-directory inside the main kernel directory. It
	will be uploaded as `README.md` to the Hub.

	### Using the kernel builder with Nix
	https://huggingface.co/docs/kernels/pr_607/builder/build.md

	# Using the kernel builder with Nix

	## Installation

	> [!NOTE]
	> The [install script](writing-kernels#quick-install) automates
	> the Nix and kernel-builder setup described below. Use these manual
	> instructions if you prefer step-by-step control.

	### Installing Nix

	The kernel builder uses Nix for building kernels. You can build or
	run the kernels directly if you have Nix installed on your system.
	We recommend installing Nix in the following way:

	- Linux: use the [official Nix installer](https://nixos.org/download/).
	- macOS: use the [Determinate Nix installer](https://docs.determinate.systems/determinate-nix/).
	In addition, Xcode 16.x is currently required to build kernels.

	### Using the Hugging Face binary cache

	Since the kernel builder depends on many packages (e.g. every supported
	PyTorch version), it is recommended to enable the huggingface cache
	to avoid expensive rebuilds.

	To use the cache, you can either install cachix and configure it:

	```bash
	# Install cachix and configure the cache
	cachix use huggingface
	```

	Or run it once without installing cachix permanently:

	```bash
	# Use cachix without installing it
	nix run nixpkgs#cachix -- use huggingface
	```

	### GPU library configuration

	The kernel builder also provides Nix development shells with all Torch
	and CUDA/ROCm dependencies needed to develop kernels (see below). If
	you want to test your kernels inside a Nix development shell and you
	are not using NixOS, [make sure that the CUDA driver is visible](https://danieldk.eu/Nix-CUDA-on-non-NixOS-systems#make-runopengl-driverlib-and-symlink-the-driver-library) to Torch.

	## Getting started

	The easiest way to start a new kernel is using the `kernel-builder init`
	subcommand, which is discussed in [Writing Kernels](writing-kernels).
	The commands discussed in the following sections will also work on
	existing kernel sources that have `build.toml`/`flake.nix`.

	## Building a kernel

	A kernel can be built with the `kernel-builder build-and-copy` command.
	For example:

	```bash
	cd examples/relu
	kernel-builder build-and-copy -L
	```

	The `-L` option prints out build logs in the terminal, which can be handy
	for monitoring the build. The compiled kernel will then be in the local
	`build/` directory.

	## Shell for local development

	`kernel-builder` provides shells for developing kernels. In such a shell,
	all required dependencies are available, as well as `kernel-builder` for generating
	project files. For example:

	```bash
	$ kernel-builder devshell
	# A devshell is opened in which you can run the following commands:
	$ kernel-builder create-pyproject
	$ cmake -B build-ext
	$ cmake --build build-ext
	```

	If you want to test the kernel as a Python package, you can do so.
	`kernel-builder devshell` will automatically create a virtual environment in
	the `.venv` and activate it. You can install the kernel as a regular
	Python package in this virtual environment:

	```bash
	$ kernel-builder devshell
	$ kernel-builder create-pyproject
	$ pip install --no-build-isolation -e .
	```

	Development shells are available for every build configuration. For
	instance, you can get a Torch 2.11 development shell for ROCm kernels
	using:

	```bash
	$ rm -rf .venv # Remove existing venv if any.
	$ kernel-builder devshell --variant torch211-cxx11-rocm71-x86_64-linux
	```

	For an editor-driven workflow with `direnv` activating the devshell on
	`cd`, see [IDE Setup](./ide-setup).

	You can list the variants that the kernel supports with the `list-variants`
	subcommand:

	```bash
	$ kernel-builder list-variants
	torch29-cxx11-cu129-x86_64-linux
	torch210-cxx11-cu126-x86_64-linux
	torch210-cxx11-cu128-x86_64-linux
	torch210-cxx11-cu130-x86_64-linux
	torch210-cxx11-rocm70-x86_64-linux
	torch210-cxx11-rocm71-x86_64-linux
	torch210-cxx11-cpu-x86_64-linux
	torch210-cxx11-xpu20253-x86_64-linux
	torch211-cxx11-cpu-x86_64-linux
	torch211-cxx11-cu126-x86_64-linux
	torch211-cxx11-cu128-x86_64-linux
	torch211-cxx11-cu130-x86_64-linux
	torch211-cxx11-rocm71-x86_64-linux
	torch211-cxx11-rocm72-x86_64-linux
	torch211-cxx11-xpu20253-x86_64-linux
	```

	## Shell for testing a kernel

	You can also start a test shell. This will give you a Python interpreter
	with the kernel in Python's search path. This makes it more convenient to run
	tests:

	```bash
	cd examples/relu
	kernel-builder testshell
	python -m pytest tests
	```

	`testshell` also supports the `--variant` option, so you can test a particular
	kernel variant.

	## Adding test dependencies to development shells

	You can add test dependencies to a development or testing shell. Adapt
	the kernel's `flake.nix` to use the `pythonCheckInputs` option:

	```nix
	{
	description = "Flake for my kernel";

	inputs = {
	builder.url = "github:huggingface/kernels";
	};

	outputs =
	{
	self,
	builder,
	}:
	builder.lib.genKernelFlakeOutputs {
	inherit self;
	path = ./.;

	# The einops and numpy test dependencies are added here:
	pythonCheckInputs = pkgs: with pkgs; [ numpy ];
	};
	}
	```

	The available packages can be found on [search.nixos.org](https://search.nixos.org/packages?channel=25.05&query=python312Packages).

	Keep in mind that these additional dependencies will only be available to
	the Nix shells, not the final kernel uploaded to the Hub.

	## Uploading your kernel to the Hub

	Finally, when you are ready to make a kernel release, you can build and
	upload a kernel to the Hub:

	```bash
	$ cd mykernel
	$ kernel-builder build-and-upload
	```

	> [!NOTE]
	> Uploads go to a `kernel`-type Hub repository (the first-class kernel
	> repository type). The owning user or org must have kernel-creation
	> access. Request it from
	> [huggingface.co/settings/account](https://huggingface.co/settings/account)
	> ("Request Kernels Creation").

	Aside from building and uploading the kernel itself, this will also fill
	the card template and upload it as `README.md` to the Hub if the card
	template is provided in the source repository as `CARD.md`.

	The repository to upload to is determined by the `repo-id` and `version`
	fields in `build.toml`. For example, with the following `build.toml`, the
	kernel will be uploaded to the repository `kernels-community/flash-attn4`
	in the `v1` version branch:

	```toml
	[general]
	# ...
	version = 1

	[general.hub]
	repo-id = "kernels-community/flash-attn4"
	```

	See [Writing Kernels](writing-kernels) for more details on the `build.toml`
	format.

	## Updating the kernel build toolchain

	The kernel's dependencies are fully pinned down in the `flake.lock` that
	is shipped with the kernel. We periodically release new versions of the
	build toolchain that includes bug fixes and supports newer Torch and compute backend
	versions. To update the kernel build toolchain, run `nix flake update`
	in the kernel directory:

	```bash
	❯ nix flake update
	• Added input 'kernel-builder':
	'github:huggingface/kernels/8ad8a5094f1b3c425f70900699ed690d65d878c3?narHash=sha256-m8tBntCIlH/rY4BcIv5X5%2BdtgSS1yQi883Co%2Bj5cudI%3D' (2026-04-09)
	• Added input 'kernel-builder/flake-compat':
	'github:edolstra/flake-compat/5edf11c44bc78a0d334f6334cdaf7d60d732daab?narHash=sha256-vNpUSpF5Nuw8xvDLj2KCwwksIbjua2LZCqhV1LNRDns%3D' (2025-12-29)
	• Added input 'kernel-builder/flake-utils':
	'github:numtide/flake-utils/11707dc2f618dd54ca8739b309ec4fc024de578b?narHash=sha256-l0KFg5HjrsfsO/JpG%2Br7fRrqm12kzFHyUHqHCVpMMbI%3D' (2024-11-13)
	• Added input 'kernel-builder/flake-utils/systems':
	'github:nix-systems/default/da67096a3b9bf56a91d16901293e51ba5b49a27e?narHash=sha256-Vy1rq5AaRuLzOxct8nz4T6wlgyUR7zLU309k9mBC768%3D' (2023-04-09)
	• Added input 'kernel-builder/nixpkgs':
	'github:NixOS/nixpkgs/2f4fd5e1abf9bac8c1d22750c701a7a5e6b524c6?narHash=sha256-Mh6bLcYAcENBAZk3RoMPMFCGGMZmfaGMERE4siZOgP4%3D' (2026-03-31)
	• Added input 'kernel-builder/rust-overlay':
	'github:oxalica/rust-overlay/962a0934d0e32f42d1b5e49186f9595f9b178d2d?narHash=sha256-JMdDYn0F%2BswYBILlpCeHDbCSyzqkeSGNxZ/Q5J584jM%3D' (2026-03-31)
	• Added input 'kernel-builder/rust-overlay/nixpkgs':
	follows 'kernel-builder/nixpkgs'
	```

	## Skipping the `get_kernel` check

	`kernel-builder` verifies that a kernel can be
	imported with the [`kernels`](https://github.com/huggingface/kernels)
	package. This check can be disabled by passing `doGetKernelCheck = false`
	to `genKernelFlakeOutputs`. Warning: it is strongly recommended to keep
	this check enabled, as it is one of the checks that validates that a kernel
	is compliant. This option is primarily intended for kernels with
	`triton.autotune` decorators, which can fail because there is no GPU available
	in the build sandbox.

	### Design overview
	https://huggingface.co/docs/kernels/pr_607/builder/design-overview.md

	# Design overview

	The Kernel Builder is a Nix flake that combines two components:

	![Kernel Builder architecture](https://huggingface.co/datasets/huggingface/documentation-images/resolve/main/kernels/kernel_design_overview.jpg)

	- `kernel-builder/`: a Rust-based CLI
	that parses and validates `build.toml`, scaffolds new kernel projects, and drives the
	build. It does not compile kernels itself: every build subcommand executes
	`nix build`/`nix run`/`nix develop` against the kernel project's `flake.nix`.
	- `nix-builder`: the Nix builder drives the build itself. It generates a build matrix using the supported build configurations
	(see `nix-builder/versions.nix`) and the kernel's `build.toml`. The build matrix is the cartesian product of
	the backend (one or more of CPU/CUDA/Metal/ROCm/XPU), backend versions, and framework versions.
	Nix builder uses `kernel-builder` to generate the CMake files that drive the build. To do so,
	`kernel-builder` itself is packaged in `nix-builder/pkgs`.

	A kernel author uses the Nix builder through the `lib.genKernelFlakeOutputs` that is exposed through the top-level `flake.nix`. This generates the Nix flake outputs for building and developing kernels, such as `bundle` for bundle builds and `devShells` for development shells.

	For a deeper look at the design of the Nix builder, see
	[Nix Builder design](./design-nix-builder).

	### Build variants
	https://huggingface.co/docs/kernels/pr_607/builder/build-variants.md

	# Build variants

	A kernel can be compliant for a specific compute framework (e.g. CUDA) or
	architecture (e.g. x86_64). For compliance with a compute framework and
	architecture combination, all the build variants listed below must be
	available. This list will be updated as new PyTorch versions are released.

	## CPU aarch64-darwin

	- `torch211-cpu-aarch64-darwin`
	- `torch212-cpu-aarch64-darwin`

	## Metal aarch64-darwin

	- `torch211-metal-aarch64-darwin`
	- `torch212-metal-aarch64-darwin`

	## CPU aarch64-linux

	- `torch211-cxx11-cpu-aarch64-linux`
	- `torch212-cxx11-cpu-aarch64-linux`

	## CUDA aarch64-linux

	- `torch211-cxx11-cu126-aarch64-linux`
	- `torch211-cxx11-cu128-aarch64-linux`
	- `torch211-cxx11-cu130-aarch64-linux`
	- `torch212-cxx11-cu126-aarch64-linux`
	- `torch212-cxx11-cu130-aarch64-linux`
	- `torch212-cxx11-cu132-aarch64-linux`

	## CPU x86_64-linux

	- `torch211-cxx11-cpu-x86_64-linux`
	- `torch212-cxx11-cpu-x86_64-linux`

	## CUDA x86_64-linux

	- `torch211-cxx11-cu126-x86_64-linux`
	- `torch211-cxx11-cu128-x86_64-linux`
	- `torch211-cxx11-cu130-x86_64-linux`
	- `torch212-cxx11-cu126-x86_64-linux`
	- `torch212-cxx11-cu130-x86_64-linux`
	- `torch212-cxx11-cu132-x86_64-linux`

	## ROCm x86_64-linux

	- `torch211-cxx11-rocm71-x86_64-linux`
	- `torch211-cxx11-rocm72-x86_64-linux`
	- `torch212-cxx11-rocm71-x86_64-linux`
	- `torch212-cxx11-rocm72-x86_64-linux`

	## XPU x86_64-linux

	- `torch211-cxx11-xpu20253-x86_64-linux`
	- `torch212-cxx11-xpu20253-x86_64-linux`

	## Python-only kernels

	Kernels that are in pure Python (e.g. Triton kernels) only need to provide
	one or more of the following variants:

	- `torch-cpu`
	- `torch-cuda`
	- `torch-metal`
	- `torch-rocm`
	- `torch-xpu`

	### Why Nix?
	https://huggingface.co/docs/kernels/pr_607/builder/why-nix.md

	# Why Nix?

	The Kernel Builder project uses Nix to build custom kernels designed specifically for PyTorch.

	Here’s why we chose Nix and why it's particularly suited to our workflow:

	## 1. Consistent and Reproducible Builds

	Nix guarantees deterministic evaluation, ensuring that every kernel is built identically, regardless of the host environment. This consistency prevents "it works on my machine" problems, making debugging and deployment straightforward.

	## 2. Simplified Dependency Management

	Compiling PyTorch kernels often involves complex dependencies such as CUDA versions, PyTorch APIs, and various C++ toolchains. Nix explicitly defines and manages these dependencies, eliminating version conflicts and making maintenance easier.

	## 3. Declarative Configuration

	Nix’s declarative approach clearly specifies exactly what each kernel build needs. This transparency aids collaboration, speeds up troubleshooting, and makes it easy to document the build process.

	## 4. Isolated, Reliable Builds

	Each kernel build with Nix runs in a fully isolated sandbox, removing any uncertainty about external state. This isolation ensures clean builds, free of unexpected side effects.

	## 5. Efficient Caching and CI Integration

	Kernel compilation can be resource-intensive. Nix leverages efficient caching of build artifacts, significantly reducing build times and optimizing continuous integration workflows.

	## 6. Easy Experimentation and Rollbacks

	Nix allows you to experiment with different kernel configurations, PyTorch versions, or CUDA toolkits easily. If a change introduces an issue, reverting to a previous state is quick and effortless.

	Overall, Nix streamlines the Kernel Builder workflow, allowing us to efficiently and reliably manage complex machine learning kernel builds.

	## Commonly Asked Questions

	Q. Why not use Docker or other containerization tools instead of Nix?

	While Docker provides isolation and consistent runtime environments, it doesn't guarantee fully reproducible builds. Factors like base image changes or implicit dependencies can still introduce variability.

	Nix focuses on reproducibility through deterministic builds, ensuring the same inputs always produce identical outputs. Its declarative configuration, precise dependency management, and efficient caching also make it well-suited for complex environments and CI/CD workflows.

	---

	If you want to learn more about Nix, check out the following resources:

	## References

	- The Official Nix Manual:
	- The definitive source for all things Nix, providing comprehensive coverage of its features, commands, and ecosystem.
	- Link: [Nix Manual (nixos.org)](https://nixos.org/manual/nix/stable/)
	- Nix Pills:
	- A series of blog posts breaking down complex Nix concepts into digestible pieces, ideal for a structured, tutorial-style approach.
	- Link: [Nix Pills (nixos.org)](https://nixos.org/guides/nix-pills/)
	- nix.dev:
	- Home of official documentation for the Nix ecosystem.
	- Link [nix.dev](https://nix.dev/)
	- NixOS Wiki:
	- A community-driven wiki with a wealth of information, including tips, tricks, and tutorials, covering a wide range of topics, including NixOS-specific information.
	- Link: [NixOS Wiki](https://nixos.wiki/wiki/Main_Page)

	### Nix Builder design
	https://huggingface.co/docs/kernels/pr_607/builder/design-nix-builder.md

	# Nix Builder design

	## Introduction

	kernel-builder uses a Nix-based builder that orchestrates the build. The Nix
	builder provides:

	- Reproducible evaluation. The same Nix builder version will always produce
	the same derivations (build recipes).
	- Largely reproducible builds by using a build sandbox that only has the
	dependencies specified in a derivation.
	- Seamless creation of different build environments (e.g. different Torch
	and CUDA combinations).

	## Kernel build steps

	A kernel derivation builds a kernel in the following steps:

	1. Generate CMake files for the kernel using
	`kernel-builder create-pyproject`.
	2. Generate Ninja build files using CMake.
	3. Build the kernel using Ninja.
	4. Perform various checks on the compiled kernel, such as:
	- Verify that the kernel only uses ABI3/`manylinux_2_28` symbols.
	- Verify that the kernel can be loaded by the `kernels` Python package.
	5. Strip runpaths (ELF-embedded library directories) from kernel binaries
	to make the kernel distribution-independent.

	## manylinux_2_28 compatibility

	To achieve `manylinux_2_28` compatibility, kernels are built using a
	toolchain similar to the `manylinux_2_28` Docker images. This toolchain
	is based on the gcc toolsets from AlmaLinux 8. `manylinux_2_28` [uses
	AlmaLinux 8 as its base](https://github.com/pypa/manylinux#manylinux_2_28-almalinux-8-based),
	so we have to compile against the same glibc/libstdc++ versions to
	ensure compatibility.

	We repackage the AlmaLinux 8 toolsets and libstdc++ as Nix derivations (see
	the `nix-builder/packages/manylinux_2_28` source directory). Then we merge
	various toolset packages to an unwrapped gcc that resembles unwrapped gcc in
	nixpkgs. Finally, we wrap binutils and gcc to combine them into a stdenv.

	The stdenv does not reuse glibc from AlmaLinux, since its dynamic loader has
	hardcoded FHS paths (`/lib64` etc.) that are not valid in Nix. Using this
	dynamic loader results in linking errors, since the paths in the dynamic
	loader are used as a last resort (to link glibc libraries). So, instead we
	build our own glibc 2.28 package
	(see `nix-builder/pkgs/manylinux_2_28/stdenv.nix`) and use that.

	## The package set pattern

	We repackage various existing package sets as Nix derivations. For instance,
	this is done for ROCm, XPU, and manylinux_2_28 packages. We do this because
	we want these libraries to be as close as what the user would install. This
	avoids compatibility issues between the kernels and the official vendor
	packages. For instance, suppose that we built a ROCm library as a shared
	library and ROCm provides the same library as a static library, then compiled
	kernels could use symbols that cannot be resolved when installing the official
	ROCm packages. Similarly, using the official packages allows us to test
	against the official upstream packages.

	These package sets all follow the same pattern:

	```nix
	{
	lib,
	callPackage,
	newScope,
	pkgs,
	}:

	{
	packageMetadata,
	}:

	let
	inherit (lib.fixedPoints) extends composeManyExtensions;

	fixedPoint = final: {
	inherit lib;
	};
	composed = lib.composeManyExtensions [
	# Base package set.
	(import ./components.nix { inherit packageMetadata; })

	# Package-specific overrides.
	(import ./overrides.nix)

	# Additional overlays that extend the package set.
	(import ./some-overlay.nix)
	];
	in
	lib.makeScope newScope (lib.extends composed fixedPoint)
	```

	We use a fixed point to build up the package set as a list of
	[overlays](https://nixos.org/manual/nixpkgs/stable/#sec-overlays-definition).
	This has various benefits. For instance, it allows us to refine the
	package set incrementally and we can refer to the final versions of
	packages in intermediate overlays.

	The package sets all use a similar list of overlays:

	- An initial overlay (`components.nix`) that applies a generic builder
	to the package set metadata. The metadata typically comes from a Yum/DNF
	repository that contains RPM packages.The generic builder will extract the
	RPMs and move binaries, libraries, and headers to the right location. This
	results in a set of Nix derivations that may or may not build.
	- The next overlay (`overrides.nix`) fixes up derivations generated by the
	generic builder in the previous overlay that do not build. Fixing the
	derivations typically consists of adding missing dependencies and changing
	embedded FHS paths to Nix store paths.
	- Additional overlays with derivations that combine outputs from previous
	overlays. One typical example are derivations that construct a full compiler
	toolchain (e.g. `nix-builder/pkgs/manylinux_2_28/gcc-unwrapped.nix`).

	### Metal kernels 🤘
	https://huggingface.co/docs/kernels/pr_607/builder/metal.md

	# Metal kernels 🤘

	Instructions on this page assume that you installed Nix with the
	[Determinate Nix installer](https://docs.determinate.systems/determinate-nix/).

	## Targeted macOS versions

	Since new macOS versions get [adopted quickly](https://telemetrydeck.com/survey/apple/macOS/versions/),
	we only support the latest major macOS version except for the first weeks
	after a release, when we also support the previous major version.

	We currently support macOS 26.0 and later on ARM64 (Apple silicon).

	## Requirements

	To build a Metal kernel, the following requirements must be met:

	- Xcode 26.x must be available on the build machine.
	- `xcode-select -p` must point to the Xcode 26 installation, typically
	`/Applications/Xcode.app/Contents/Developer`. If this is not the case,
	you can set the path with:
	`sudo xcode-select -s /path/to/Xcode.app/Contents/Developer`
	- The Metal Toolchain must be installed. Starting with macOS 26, this is
	a separate download from Xcode. You can install it with:
	`xcodebuild -downloadComponent MetalToolchain`
	- The Nix sandbox should be set to `relaxed`, because the Nix derivation
	that builds the kernel must have access to Xcode and the Metal Toolchain.
	You can verify this by checking that `/etc/nix/nix.custom.conf` contains
	the line:

	```
	sandbox = relaxed
	```

	If you had to add the line, make sure to restart the Nix daemon:

	```
	sudo launchctl kickstart -k system/systems.determinate.nix-daemon
	```

	You can check these requirements as follows. First, you can check the Xcode
	version as follows:

	```bash
	$ xcodebuild -version
	Xcode 26.1
	Build version 17B55
	```

	The reported version must be 26.0 or newer. Then you can validate that the
	Metal Toolchain is installed with:

	```bash
	$ xcodebuild -showComponent metalToolchain
	Asset Path: /System/Library/AssetsV2/com_apple_MobileAsset_MetalToolchain/68d8db6212b48d387d071ff7b905df796658e713.asset/AssetData
	Build Version: 17B54
	Status: installed
	Toolchain Identifier: com.apple.dt.toolchain.Metal.32023
	Toolchain Search Path: /private/var/run/com.apple.security.cryptexd/mnt/com.apple.MobileAsset.MetalToolchain-v17.2.54.0.mDxgz0
	```

	### Security
	https://huggingface.co/docs/kernels/pr_607/builder/security.md

	# Security

	## Introduction

	As a kernel builder, you provide code that might be run on thousands or
	even millions of machines. This comes with the responsibility of ensuring
	no malicious code is distributed.

	Below, we provide guidelines to help avoid common attack vectors. These
	are _in addition to_ common-sense security practices, such as keeping
	your credentials/tokens safe, being vigilant against machine compromise,
	and doing proper code reviews.

	## Handling pull requests

	The Hugging Face Hub allows users to submit pull requests to your
	repositories. Never merge a pull request that contains a `build/`
	directory. The binaries inside the `build/` directory might be compromised
	even when the source code looks fine. When a pull request includes
	`build/`, ask the submitter to re-submit it without builds. Build the
	kernel on your own trusted infrastructure after the PR is merged.

	When a PR does not contain build outputs and is ready to review, _carefully_
	review every changed line, also taking security into account. Even if the
	PR is from a trusted party, review it as if their credentials might have
	been compromised.

	## Build hygiene

	If possible, do builds on a dedicated build machine/VM that is only used
	for sandboxed builds (non-macOS kernel-builder builds are sandboxed as
	well). Specialized machines are less likely to be compromised, especially
	when they are accessed with a hardware-stored SSH key that requires user
	interaction.

	## Supporting reproducibility

	Reproducible builds are very helpful to verify that no malicious code has
	slipped into a kernel. If a kernel build is reproducible, then anyone can
	rebuild a kernel and verify the binaries match the distributed binaries.
	Full reproducibility is a goal we are working toward in `kernel-builder`.

	However, this also requires assistance from the kernel builder. This section
	describes what you need to do to make reproducible builds possible.

	### Only build kernels with Nix sandboxing enabled.

	Nix can be used with sandboxing disabled to support systems that do not
	support sandboxing (e.g. Linux systems that are configured to disable
	mount/network namespaces). Never build kernels with sandboxing disabled.
	Not only can this cause stray system dependencies to be picked up, but
	it can also cause other impurities to slip into the build, making it
	impossible to reproduce the build. You can verify that sandboxing is enabled
	using `nix-info`:

	```bash
	$ nix-shell -p nix-info --run "nix-info -m"
	- system: `"x86_64-linux"`
	- host os: `Linux 6.12.39, NixOS, 25.11 (Xantusia), 25.11.20250723.1744f3d`
	- multi-user?: `yes`
	- sandbox: `yes`
	- version: `nix-env (Nix) 2.28.4`
	- nixpkgs: `/nix/store/fqwc3ghi5qfdmzklpwssbamxcqj1vgl3-source`
	```

	### Do not build from dirty Git trees

	Before building a kernel, ensure that all changes are committed. This
	makes it possible to reproduce a build from exactly the same source code.
	We bake the git shorthash into the ops name, so that it is clear from
	which git hash a kernel was built.

	### Local development of kernels
	https://huggingface.co/docs/kernels/pr_607/builder/local-dev.md

	# Local development of kernels

	## Introduction

	`kernel-builder` builds kernels in a sandbox. This has various benefits,
	such as building kernels for a wide range of Torch versions, compatibility
	with older C library versions and avoiding accidental dependencies.

	However, this is not ideal during kernel development, since language
	servers and IDEs do not interpret the `build.toml` file. As a result,
	code completion will typically not work. `kernel-builder` provides the
	`kernel-builder` utility to generate CMake files to build native code and
	setuptools files for building the kernel as a regular Python package.
	Since CMake and setuptools are widely supported by IDEs, this provides
	a much-improved development experience.

	## Generating a Python project with `kernel-builder`

	`kernel-builder` can create CMake/Python project files for a kernel with
	a [`build.toml`](./writing-kernels) file. The `create-pyproject`
	command will create the files for the kernel in the current directory:

	```bash
	$ kernel-builder create-pyproject -f
	```

	The `-f` flag is optional and instructs `kernel-builder` to overwrite
	existing files.

	It is recommended to do an editable install of the generated project into
	your Python virtual environment for development:

	```bash
	$ pip install wheel # Needed once to enable bdist_wheel.
	$ pip install --no-build-isolation -e .
	```

	You can also create a Python project for a kernel in another directory:

	```bash
	$ kernel-builder create-pyproject -f path/to/kernel
	```

	Warnings:

	- Kernels built in this way should not be published on the Kernel
	Hub. They do not fulfill the [kernel requirements](../kernel-requirements).
	- Do not add the generated files to Git. `kernel-builder` has regular updates
	and you generally want to use files generated by the latest version.

	See [IDE Setup](./ide-setup) for wiring the generated project into
	VS Code with direnv.

	## Testing kernel builds before publishing

	Once you have built a kernel with kernel builder, you may want to test it
	locally with software that uses `get_kernel` or `LayerRepository` before
	publishing. This can be done using the `LOCAL_KERNELS` variable, which
	maps a repository ID to a local kernel directory. For example, you could
	use the kernel in `devel/activation` for any use of the
	`kernels-community/activation` repository with:

	```bash
	$ LOCAL_KERNELS="kernels-community/activation=devel/activation" \
	python my_app.py
	```

	It is also possible to map multiple repositories to local kernel directories
	by separating the entries with a colon (`:`):

	```bash
	$ LOCAL_KERNELS="kernels-community/activation=devel/activation:kernels-community/flash-attn2=devel/flash-attn2" \
	python my_app.py
	```

Xet Storage Details

Size:: 160 kB
Xet hash:: a67920d6028781c449e5cdfc6b4f1878fabf2842612f804286dfa6a1cfc2bef9

Xet efficiently stores files, intelligently splitting them into unique chunks and accelerating uploads and downloads. More info.