diff --git a/docs/source/builder/build.md b/docs/source/builder/build.md
index 71beb224a..9cd7b8b10 100644
--- a/docs/source/builder/build.md
+++ b/docs/source/builder/build.md
@@ -274,3 +274,7 @@ 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.
+
+Generating a kernel's public API symbols (`symbols.json`) also requires
+importing the kernel. So, when the `get_kernel` check is disabled, the build
+variants will not contain a `symbols.json` file.
diff --git a/docs/source/kernel-requirements.md b/docs/source/kernel-requirements.md
index 761520510..3da858adc 100644
--- a/docs/source/kernel-requirements.md
+++ b/docs/source/kernel-requirements.md
@@ -486,6 +486,16 @@ __all__ = [
 > [versioning guarantees](#versioning) apply to, so be sure to export
 > every function, class, and `layers` module you want to expose.
 
+The same applies to the `layers` module itself: only the layers listed in
+the `__all__` of `layers` are part of the public API. For example:
+
+```python
+class SiluAndMul(nn.Module):
+    # ...
+
+__all__ = ["SiluAndMul"]
+```
+
 ## Python requirements
 
 - Python code must be compatible with Python 3.9 and later.
diff --git a/examples/kernels/flake.nix b/examples/kernels/flake.nix
index 0232bd91d..b0085be49 100644
--- a/examples/kernels/flake.nix
+++ b/examples/kernels/flake.nix
@@ -21,6 +21,12 @@
       torchVersion = "213";
       tvmFfiVersion = "01";
 
+      # Expected public symbols of the relu example kernels.
+      reluSymbols = {
+        functions = [ "relu" ];
+        layers = [ "ReLU" ];
+      };
+
       # All example kernels to build in CI.
       #
       # - name: name in the output path
@@ -34,6 +40,9 @@
       # - checkCudaCapabilities: optional list of CUDA capabilities (e.g. "9.0").
       #        When set, the kernel dylib must contain exactly this set of
       #        capabilities.
+      # - checkSymbols: optional attrset `{ functions = [ ... ]; layers = [ ... ]; }`.
+      #        When set, every build variant must contain a symbols.json with
+      #        exactly these public function and layer names.
       ciKernels = [
         {
           name = "cpp20-symbols-kernel";
@@ -45,6 +54,7 @@
           name = "relu-kernel";
           path = ./relu;
           drv = sys: out: out.packages.${sys}.redistributable.${"torch${torchVersion}-${cudaVersion}-${sys}"};
+          checkSymbols = reluSymbols;
           checkCudaCapabilities = [
             "7.0"
             "7.2"
@@ -97,6 +107,13 @@
           path = ./relu-tvm-ffi;
           drv =
             sys: out: out.packages.${sys}.redistributable.${"tvm-ffi${tvmFfiVersion}-${cudaVersion}-${sys}"};
+          checkSymbols = {
+            functions = [
+              "relu"
+              "relu_jax"
+            ];
+            layers = [ ];
+          };
         }
         {
           name = "relu-tvm-ffi-compiler-flags-kernel";
@@ -293,6 +310,7 @@
           name = "relu-triton-kernel";
           path = ./relu-triton;
           drv = sys: out: out.packages.${sys}.redistributable.torch-xpu;
+          checkSymbols = reluSymbols;
         }
         {
           name = "gemm-triton-autotune-kernel";
@@ -303,12 +321,20 @@
           name = "relu-kernel";
           path = ./relu;
           drv = sys: out: out.packages.${sys}.redistributable.${"torch${torchVersion}-${xpuVersion}-${sys}"};
+          checkSymbols = reluSymbols;
         }
         {
           name = "relu-tvm-ffi-kernel";
           path = ./relu-tvm-ffi;
           drv =
             sys: out: out.packages.${sys}.redistributable.${"tvm-ffi${tvmFfiVersion}-${xpuVersion}-${sys}"};
+          checkSymbols = {
+            functions = [
+              "relu"
+              "relu_jax"
+            ];
+            layers = [ ];
+          };
         }
         {
           name = "relu-tvm-ffi-compiler-flags-kernel";
@@ -345,6 +371,7 @@
           name = "relu-kernel";
           path = ./relu;
           drv = sys: out: out.packages.${sys}.redistributable.${"torch${torchVersion}-metal-${sys}"};
+          checkSymbols = reluSymbols;
         }
         {
           name = "relu-metal-cpp-kernel";
@@ -456,6 +483,41 @@
                 ln -s ${drv} $out
               '';
 
+          # Check that every build variant in the output of `drv` contains a
+          # symbols.json with exactly the expected public function and layer
+          # names. On success, the output is a symlink to `drv`.
+          checkSymbols =
+            drv: expected:
+            pkgs.runCommand "${drv.name}-check-symbols"
+              {
+                nativeBuildInputs = [ pkgs.jq ];
+                expected = builtins.toJSON expected;
+                passAsFile = [ "expected" ];
+              }
+              ''
+                variants=$(find ${drv}/ -mindepth 2 -maxdepth 2 -name metadata.json)
+                if [ -z "$variants" ]; then
+                  echo "no build variants found in ${drv}" >&2
+                  exit 1
+                fi
+
+                jq -S . "$expectedPath" > expected
+                for metadata in $variants; do
+                  symbols="$(dirname "$metadata")/symbols.json"
+                  if [ ! -f "$symbols" ]; then
+                    echo "missing $symbols" >&2
+                    exit 1
+                  fi
+                  jq -S '{functions: [.functions[].name], layers: [.layers[].name]}' "$symbols" > actual
+                  if ! diff -u expected actual; then
+                    echo "unexpected symbols in $symbols" >&2
+                    exit 1
+                  fi
+                done
+
+                ln -s ${drv} $out
+              '';
+
           checkRocmArchs =
             drv: expectedArchs:
             checkKernelArchs drv expectedArchs
@@ -484,18 +546,23 @@
               drv =
                 let
                   baseDrv = kernel.drv system kernel.outputs;
+                  archsDrv =
+                    if kernel ? checkRocmArchs then
+                      checkRocmArchs baseDrv kernel.checkRocmArchs
+                    else if kernel ? checkCudaCapabilities then
+                      checkCudaCapabilities baseDrv kernel.checkCudaCapabilities
+                    else
+                      baseDrv;
                 in
                 if kernel.assertFail or false then
                   pkgs.testers.testBuildFailure' {
                     drv = baseDrv;
                     expectedBuilderLogEntries = kernel.assertFailLogs or [ ];
                   }
-                else if kernel ? checkRocmArchs then
-                  checkRocmArchs baseDrv kernel.checkRocmArchs
-                else if kernel ? checkCudaCapabilities then
-                  checkCudaCapabilities baseDrv kernel.checkCudaCapabilities
+                else if kernel ? checkSymbols then
+                  checkSymbols archsDrv kernel.checkSymbols
                 else
-                  baseDrv;
+                  archsDrv;
             }) kernelOutputsList;
 
           mkCiBuild =
diff --git a/examples/kernels/relu-nki/torch-ext/relu_nki/layers/__init__.py b/examples/kernels/relu-nki/torch-ext/relu_nki/layers/__init__.py
index 7800a2ef7..d632e1af4 100644
--- a/examples/kernels/relu-nki/torch-ext/relu_nki/layers/__init__.py
+++ b/examples/kernels/relu-nki/torch-ext/relu_nki/layers/__init__.py
@@ -7,3 +7,6 @@
 class ReLU(nn.Module):
     def forward(self, x: torch.Tensor) -> torch.Tensor:
         return relu(x)
+
+
+__all__ = ["ReLU"]
diff --git a/examples/kernels/relu-torch-stable-abi/torch-ext/relu_torch_stable_abi/layers/__init__.py b/examples/kernels/relu-torch-stable-abi/torch-ext/relu_torch_stable_abi/layers/__init__.py
index 6105a1910..ee2ebd82b 100644
--- a/examples/kernels/relu-torch-stable-abi/torch-ext/relu_torch_stable_abi/layers/__init__.py
+++ b/examples/kernels/relu-torch-stable-abi/torch-ext/relu_torch_stable_abi/layers/__init__.py
@@ -9,3 +9,6 @@ def forward(self, x: torch.Tensor) -> torch.Tensor:
         out = torch.empty_like(x)
         ops.relu(out, x)
         return out
+
+
+__all__ = ["ReLU"]
diff --git a/examples/kernels/relu-tpu/torch-ext/relu_tpu/layers/__init__.py b/examples/kernels/relu-tpu/torch-ext/relu_tpu/layers/__init__.py
index 7800a2ef7..d632e1af4 100644
--- a/examples/kernels/relu-tpu/torch-ext/relu_tpu/layers/__init__.py
+++ b/examples/kernels/relu-tpu/torch-ext/relu_tpu/layers/__init__.py
@@ -7,3 +7,6 @@
 class ReLU(nn.Module):
     def forward(self, x: torch.Tensor) -> torch.Tensor:
         return relu(x)
+
+
+__all__ = ["ReLU"]
diff --git a/examples/kernels/relu-triton/torch-ext/relu_triton/layers.py b/examples/kernels/relu-triton/torch-ext/relu_triton/layers.py
index 4b7dfacff..d25ab409b 100644
--- a/examples/kernels/relu-triton/torch-ext/relu_triton/layers.py
+++ b/examples/kernels/relu-triton/torch-ext/relu_triton/layers.py
@@ -9,3 +9,6 @@ def forward(self, x: torch.Tensor) -> torch.Tensor:
         out = torch.empty_like(x)
         ops.relu(out, x)
         return out
+
+
+__all__ = ["ReLU"]
diff --git a/examples/kernels/relu-tvm-ffi-compiler-flags/tvm-ffi-ext/relu_tvm_ffi_compiler_flags/layers/__init__.py b/examples/kernels/relu-tvm-ffi-compiler-flags/tvm-ffi-ext/relu_tvm_ffi_compiler_flags/layers/__init__.py
index 8409908f5..102e708c4 100644
--- a/examples/kernels/relu-tvm-ffi-compiler-flags/tvm-ffi-ext/relu_tvm_ffi_compiler_flags/layers/__init__.py
+++ b/examples/kernels/relu-tvm-ffi-compiler-flags/tvm-ffi-ext/relu_tvm_ffi_compiler_flags/layers/__init__.py
@@ -10,5 +10,6 @@ def forward(self, x: torch.Tensor) -> torch.Tensor:
             relu(x, out)
             return out
 
+    __all__ = ["ReLU"]
 except ImportError:
-    pass
+    __all__ = []
diff --git a/examples/kernels/relu-tvm-ffi/tvm-ffi-ext/relu_tvm_ffi/layers/__init__.py b/examples/kernels/relu-tvm-ffi/tvm-ffi-ext/relu_tvm_ffi/layers/__init__.py
index 8409908f5..102e708c4 100644
--- a/examples/kernels/relu-tvm-ffi/tvm-ffi-ext/relu_tvm_ffi/layers/__init__.py
+++ b/examples/kernels/relu-tvm-ffi/tvm-ffi-ext/relu_tvm_ffi/layers/__init__.py
@@ -10,5 +10,6 @@ def forward(self, x: torch.Tensor) -> torch.Tensor:
             relu(x, out)
             return out
 
+    __all__ = ["ReLU"]
 except ImportError:
-    pass
+    __all__ = []
diff --git a/examples/kernels/relu/torch-ext/relu/layers/__init__.py b/examples/kernels/relu/torch-ext/relu/layers/__init__.py
index 6105a1910..ee2ebd82b 100644
--- a/examples/kernels/relu/torch-ext/relu/layers/__init__.py
+++ b/examples/kernels/relu/torch-ext/relu/layers/__init__.py
@@ -9,3 +9,6 @@ def forward(self, x: torch.Tensor) -> torch.Tensor:
         out = torch.empty_like(x)
         ops.relu(out, x)
         return out
+
+
+__all__ = ["ReLU"]
diff --git a/nix-builder/lib/extension/torch/arch.nix b/nix-builder/lib/extension/torch/arch.nix
index 3522227bd..e9fbb322e 100644
--- a/nix-builder/lib/extension/torch/arch.nix
+++ b/nix-builder/lib/extension/torch/arch.nix
@@ -13,6 +13,7 @@
   cmake,
   cmakeNvccThreadsHook,
   cuda_nvcc,
+  generate-symbols-hook,
   get-kernel-check,
   hash-kernel-hook,
   kernel-layout-check,
@@ -188,6 +189,10 @@ stdenv.mkDerivation (prevAttrs: {
       python3 = python3.withPackages (ps: dependencies);
       kernels = overrideTorch python3.pkgs.kernels;
     })
+    (generate-symbols-hook.override {
+      python3 = python3.withPackages (ps: dependencies);
+      kernels = overrideTorch python3.pkgs.kernels;
+    })
   ]
   ++ lib.optionals cudaSupport [
     cmakeNvccThreadsHook
diff --git a/nix-builder/lib/extension/torch/no-arch.nix b/nix-builder/lib/extension/torch/no-arch.nix
index ae3e365cd..a1934d173 100644
--- a/nix-builder/lib/extension/torch/no-arch.nix
+++ b/nix-builder/lib/extension/torch/no-arch.nix
@@ -8,6 +8,7 @@
   stdenv,
 
   kernel-builder,
+  generate-symbols-hook,
   get-kernel-check,
   hash-kernel-hook,
   kernel-layout-check,
@@ -116,6 +117,10 @@ stdenv.mkDerivation (prevAttrs: {
       python3 = python3.withPackages (_: dependencies);
       kernels = overrideTorch python3.pkgs.kernels;
     })
+    (generate-symbols-hook.override {
+      python3 = python3.withPackages (_: dependencies);
+      kernels = overrideTorch python3.pkgs.kernels;
+    })
   ];
 
   buildPhase = ''
diff --git a/nix-builder/lib/extension/tvm-ffi/arch.nix b/nix-builder/lib/extension/tvm-ffi/arch.nix
index ed1aca28a..3f07c337d 100644
--- a/nix-builder/lib/extension/tvm-ffi/arch.nix
+++ b/nix-builder/lib/extension/tvm-ffi/arch.nix
@@ -14,6 +14,7 @@
   cmake,
   cmakeNvccThreadsHook,
   cuda_nvcc,
+  generate-symbols-hook,
   get-kernel-check,
   hash-kernel-hook,
   kernel-layout-check,
@@ -136,6 +137,30 @@ let
 
   rustSupport = cargoLock != null;
 
+  # rpaths are stripped from kernels to make them portable, but that
+  # also means that in a Nix environment the CUDA/oneAPI dependencies
+  # cannot be located anymore, so pass them to hooks that load the kernel.
+  libraryPath = lib.makeLibraryPath (
+    map lib.getLib (
+      lib.optionals cudaSupport (
+        with cudaPackages;
+        [
+          cuda_cudart
+          libcublas
+          libcusolver
+          libcusparse
+        ]
+      )
+      ++ lib.optionals xpuSupport (
+        with xpuPackages;
+        [
+          intel-oneapi-compiler-dpcpp-cpp-runtime
+          intel-oneapi-compiler-shared-runtime
+        ]
+      )
+    )
+  );
+
   provenanceFlags = import ../provenance-flags.nix { inherit lib kernelProvenance; };
 
 in
@@ -203,29 +228,12 @@ stdenv.mkDerivation (
       (get-kernel-check.override {
         python3 = python3.withPackages (ps: dependencies);
         kernels = python3.pkgs.kernels.override { withTorch = false; };
-        # rpaths are stripped from kernels to make them portable, but that
-        # also means that in a Nix environment the CUDA/oneAPI dependencies
-        # cannot be located anymore, so pass them to get-kernel-check.
-        libraryPath = lib.makeLibraryPath (
-          map lib.getLib (
-            lib.optionals cudaSupport (
-              with cudaPackages;
-              [
-                cuda_cudart
-                libcublas
-                libcusolver
-                libcusparse
-              ]
-            )
-            ++ lib.optionals xpuSupport (
-              with xpuPackages;
-              [
-                intel-oneapi-compiler-dpcpp-cpp-runtime
-                intel-oneapi-compiler-shared-runtime
-              ]
-            )
-          )
-        );
+        inherit libraryPath;
+      })
+      (generate-symbols-hook.override {
+        python3 = python3.withPackages (ps: dependencies);
+        kernels = python3.pkgs.kernels.override { withTorch = false; };
+        inherit libraryPath;
       })
     ]
     ++ lib.optionals cudaSupport [
diff --git a/nix-builder/overlay.nix b/nix-builder/overlay.nix
index 4cf33e2ef..909173ba4 100644
--- a/nix-builder/overlay.nix
+++ b/nix-builder/overlay.nix
@@ -18,6 +18,10 @@ final: prev:
 
   fetchKernelDeps = final.callPackage ./pkgs/fetch-kernel-deps { };
 
+  generate-symbols = final.python3.pkgs.callPackage ./pkgs/generate-symbols { };
+
+  generate-symbols-hook = final.callPackage ./pkgs/generate-symbols/hook.nix { };
+
   get-kernel-check = final.callPackage ./pkgs/get-kernel-check { };
 
   hash-kernel-hook = final.callPackage ./pkgs/hash-kernel-hook { };
diff --git a/nix-builder/pkgs/generate-symbols/default.nix b/nix-builder/pkgs/generate-symbols/default.nix
new file mode 100644
index 000000000..363918f6c
--- /dev/null
+++ b/nix-builder/pkgs/generate-symbols/default.nix
@@ -0,0 +1,40 @@
+{
+  lib,
+  buildPythonPackage,
+  pytestCheckHook,
+  setuptools,
+
+  kernels,
+}:
+
+let
+  version = (builtins.fromTOML (builtins.readFile ./pyproject.toml)).project.version;
+in
+buildPythonPackage {
+  pname = "generate-symbols";
+  inherit version;
+  pyproject = true;
+
+  src = lib.fileset.toSource {
+    root = ./.;
+    fileset = lib.fileset.unions [
+      ./pyproject.toml
+      ./src
+      ./tests
+    ];
+  };
+
+  build-system = [ setuptools ];
+
+  dependencies = [ kernels ];
+
+  nativeCheckInputs = [ pytestCheckHook ];
+
+  pythonImportsCheck = [ "generate_symbols" ];
+
+  meta = {
+    description = "Generate public API symbols for built kernels";
+    license = lib.licenses.asl20;
+    mainProgram = "generate-symbols";
+  };
+}
diff --git a/nix-builder/pkgs/generate-symbols/generate-symbols-hook.sh b/nix-builder/pkgs/generate-symbols/generate-symbols-hook.sh
new file mode 100644
index 000000000..b3364411b
--- /dev/null
+++ b/nix-builder/pkgs/generate-symbols/generate-symbols-hook.sh
@@ -0,0 +1,62 @@
+#!/bin/sh
+
+echo "Sourcing generate-symbols-hook.sh"
+
+_generateSymbolsHook() {
+  echo "Generating kernel symbols"
+
+  if [ -z ${kernelDeps+x} ]; then
+    echo "kernelDeps must be set in derivation"
+    exit 1
+  fi
+
+  if [ -z ${variant+x} ]; then
+    echo "variant must be set in derivation"
+    exit 1
+  fi
+
+  # We strip the full library paths from the extension. Unfortunately,
+  # in a Nix environment, the library dependencies cannot be found
+  # anymore. So we have to add the Torch library directory to the
+  # dynamic linker path to get it to pick it up.
+  if [ $(uname -s) == "Darwin" ]; then
+    TORCH_DIR=$(python -c "from pathlib import Path; import torch; print(Path(torch.__file__).parent)")
+    export DYLD_LIBRARY_PATH="${TORCH_DIR}/lib:${DYLD_LIBRARY_PATH}"
+  fi
+
+  HOME=$(mktemp -d -t test.XXXXXX) || exit 1
+  trap "rm -rf '$HOME'" EXIT
+
+  # Prepare fake /sys for tcmalloc. Without this path, tcmalloc will crash:
+  #
+  # https://github.com/google/tcmalloc/issues/245
+  #
+  # tcmalloc is used by the TPU libraries.
+  local prootCmd=""
+  if [[ -n "@useFakeSys@" ]]; then
+      echo "Faking /sys for tcmalloc"
+      local fakeSys
+      fakeSys="$(mktemp -d)"
+      trap 'rm -rf -- "${fakeSys}"' EXIT
+      mkdir -p "${fakeSys}/devices/system/cpu"
+      echo "0-1" > "${fakeSys}/devices/system/cpu/possible"
+      prootCmd="@proot@ -b ${fakeSys}:/sys"
+  fi
+
+  # Avoid adding an empty entry to the library path, which the dynamic
+  # linker interprets as the current directory.
+  local libraryPath="${LD_LIBRARY_PATH-}"
+  if [[ -n "@libraryPath@" ]]; then
+    libraryPath="@libraryPath@${libraryPath:+:${libraryPath}}"
+  fi
+
+  LD_LIBRARY_PATH="${libraryPath}" \
+  PYTHONPATH="@pythonPath@" \
+    ${prootCmd} \
+      @python3@ -m generate_symbols \
+        --kernel-deps "${kernelDeps}" \
+        --output "${out}/${variant}/symbols.json" \
+        "${out}/${variant}"
+}
+
+postInstallCheckHooks+=(_generateSymbolsHook)
diff --git a/nix-builder/pkgs/generate-symbols/hook.nix b/nix-builder/pkgs/generate-symbols/hook.nix
new file mode 100644
index 000000000..2eec66c9e
--- /dev/null
+++ b/nix-builder/pkgs/generate-symbols/hook.nix
@@ -0,0 +1,31 @@
+{
+  config,
+  lib,
+  makeSetupHook,
+
+  generate-symbols,
+  proot,
+  python3,
+
+  kernels ? python3.pkgs.kernels,
+
+  # rpaths are stripped from kernels to make them portable, but that
+  # also means that in a Nix environment the CUDA dependencies cannot
+  # be located anymore. This argument is used to provide additional
+  # library directories to be provided to the dynamic loader.
+  libraryPath ? "",
+}:
+
+let
+  useFakeSys = config.tpuSupport or false;
+in
+makeSetupHook {
+  name = "generate-symbols-hook";
+  substitutions = {
+    python3 = "${python3}/bin/python";
+    pythonPath = python3.pkgs.makePythonPath [ (generate-symbols.override { inherit kernels; }) ];
+    inherit libraryPath;
+    proot = lib.optionalString useFakeSys "${proot}/bin/proot";
+    useFakeSys = lib.optionalString useFakeSys "1";
+  };
+} ./generate-symbols-hook.sh
diff --git a/nix-builder/pkgs/generate-symbols/src/generate_symbols.py b/nix-builder/pkgs/generate-symbols/src/generate_symbols.py
index 2b220ec10..0cab47630 100644
--- a/nix-builder/pkgs/generate-symbols/src/generate_symbols.py
+++ b/nix-builder/pkgs/generate-symbols/src/generate_symbols.py
@@ -112,6 +112,58 @@ def generate_symbols(module: ModuleType) -> dict[str, Any]:
     }
 
 
+def load_kernel_with_paths(
+    kernel_dir: Path, kernel_deps: Path, *, backend: str | None = None
+) -> ModuleType:
+    """Load a kernel, resolving its kernel dependencies from local paths.
+
+    This is used in the Nix build sandbox, since there is no network access to
+    download dependencies.
+    """
+    from kernels._rust import KernelDependency, KernelPaths, KernelVersion
+    from kernels.hf_hub import _get_hf_api
+    from kernels.load import get_kernel_with_resolver
+    from kernels.resolver import (
+        KernelPathsResolver,
+        RepoPathsResolver,
+        SequentialResolver,
+    )
+    from kernels.validate import (
+        AllKernelValidator,
+        AllMetadataValidator,
+        default_metadata_validators,
+    )
+
+    kernel_paths = KernelPaths.from_json(kernel_deps.read_text(encoding="utf-8"))
+    repo_id = str(kernel_dir)
+    resolvers = [
+        RepoPathsResolver(local_kernels={repo_id: kernel_dir}),
+        KernelPathsResolver(kernel_paths=kernel_paths),
+    ]
+
+    return get_kernel_with_resolver(
+        api=_get_hf_api(),
+        backend=backend,
+        kernel=KernelDependency(repo_id=repo_id, version=KernelVersion.Version(0)),
+        resolver=SequentialResolver(resolvers=resolvers),
+        kernel_validator=AllKernelValidator(validators=[]),
+        metadata_validator=AllMetadataValidator(
+            validators=default_metadata_validators()
+        ),
+    )
+
+
+def load_kernel(
+    kernel_dir: Path, *, backend: str | None = None, kernel_deps: Path | None = None
+) -> ModuleType:
+    if kernel_deps is not None:
+        return load_kernel_with_paths(kernel_dir, kernel_deps, backend=backend)
+
+    from kernels import get_local_kernel
+
+    return get_local_kernel(kernel_dir, backend=backend)
+
+
 def main() -> None:
     parser = argparse.ArgumentParser(description=__doc__)
     parser.add_argument(
@@ -122,14 +174,19 @@ def main() -> None:
     parser.add_argument(
         "--backend", help="Backend to select when loading a local kernel repository"
     )
+    parser.add_argument(
+        "--kernel-deps",
+        type=Path,
+        help=("JSON file mapping kernel dependencies to local paths."),
+    )
     parser.add_argument(
         "--output", type=Path, required=True, help="Destination symbols.json"
     )
     args = parser.parse_args()
 
-    from kernels import get_local_kernel
-
-    module = get_local_kernel(args.kernel_dir, backend=args.backend)
+    module = load_kernel(
+        args.kernel_dir, backend=args.backend, kernel_deps=args.kernel_deps
+    )
     # Finish inspection and serialization before touching an existing output file.
     content = json.dumps(generate_symbols(module), indent=2, ensure_ascii=False) + "\n"
     args.output.write_text(content, encoding="utf-8")
diff --git a/nix-builder/pkgs/generate-symbols/tests/test_generate_symbols.py b/nix-builder/pkgs/generate-symbols/tests/test_generate_symbols.py
index d211e4de8..9609d8d85 100644
--- a/nix-builder/pkgs/generate-symbols/tests/test_generate_symbols.py
+++ b/nix-builder/pkgs/generate-symbols/tests/test_generate_symbols.py
@@ -229,3 +229,30 @@ def test_result_can_be_written_as_json(kernel):
     assert result["schema_version"] == 1
     assert result["module"] == "example"
     assert json.loads(json.dumps(result)) == result
+
+
+def test_kernel_deps_selects_path_based_loading(kernel, monkeypatch, tmp_path):
+    calls = []
+
+    def load_kernel_with_paths(kernel_dir, kernel_deps, *, backend=None):
+        calls.append((kernel_dir, kernel_deps, backend))
+        return kernel
+
+    monkeypatch.setattr(generator, "load_kernel_with_paths", load_kernel_with_paths)
+    output = tmp_path / "symbols.json"
+    monkeypatch.setattr(
+        "sys.argv",
+        [
+            "generate-symbols",
+            "--kernel-deps",
+            str(tmp_path / "deps.json"),
+            "--output",
+            str(output),
+            str(tmp_path / "variant"),
+        ],
+    )
+
+    generator.main()
+
+    assert calls == [(tmp_path / "variant", tmp_path / "deps.json", None)]
+    assert json.loads(output.read_text()) == generator.generate_symbols(kernel)