feat(kyojin): add Kyojin ExLlamaV3 engine for AMD Strix Halo

Kyojin (Yamz-Labs/kyojin) is a fork of ExLlamaV3 that runs 100 GB-class
EXL3 MoE packs on one 128 GB Ryzen AI Max box. It ships a torch C++
extension that torch's cpp_extension cross-compiles to HIP, and upstream
supports gfx1151 only, so the kernels are built for gfx1151 alone and
the package is marked Linux-only. Pinned to the v1.3 release.

The extension must be compiled against a ROCm build of torch. The flake's
nixos-unstable carries torch 2.13, whose ROCm build fails in nixpkgs (the
CK SDPA configure step runs a script through /bin/bash) and has no binary
on the cache, so the build would compile PyTorch from source and die.
python3Packages.torchWithRocm from the already pinned nixpkgs-torch211
input gives torch 2.11 with ROCm 7.2.2, gfx1151 in its target list and a
cached binary, so default.nix builds through that input the way freetoken
does.

nixpkgs splits the single devel tree the AMD wheels ship, so two
symlinkJoins stand in for it: one as ROCM_HOME (hipcc, headers, amdgcn
bitcode) and one handed to setup.py as EXL3_ROCM_DEV_INCLUDE, which wants
hipsparse/, rocsparse/, rocrand/, thrust/ and pybind11 (nixpkgs torch no
longer exports pybind11 headers). hipsolver is in there because torch's
own HIPContextLight.h includes it.

setup.py gets one patch in postPatch. It imports exllamav3 to reach
build_config, that runs the package __init__, which imports ext.py, which
JIT-compiles the entire extension whenever no precompiled exllamav3_ext is
importable, meaning every wheel build, into a $HOME the sandbox does not
grant. Registering a stub package keeps the two leaf modules importable
without that detour.

doCheck = false: tests/ drives a real gfx1151 GPU and the published
packs. Inference on a card is not verified here, this builder has no
/dev/kfd, so the wheel-only LD_PRELOAD workaround for torch's bundled HSA
runtime is untested. It should not be needed, the store torch links the
store rocm-runtime.

Verified with nix build .#kyojin at v1.3: exllamav3_ext loads, torch
reports hip 7.2.53211, the closure holds one torch (ROCm), and
kyojin-serve-qwen plus kyojin-serve-glm print their usage.

Refs nix-overlay-du9

README.md

@@ -31,6 +31,7 @@ A custom Nix overlay and flake providing additional packages not found in upstre
 | `shardr` | Decentralized LLM repository: content-addressed model store, BitTorrent-based sync, OpenAI-compatible serving via llama-server | AI Inference |
 | `skillsmcp` | MCP server that exposes Agent Skills to AI agents via the Model Context Protocol | MCP Servers |
 | `kubernetes-mcp-server` | Model Context Protocol (MCP) server for Kubernetes and OpenShift | MCP Servers |
+| `kyojin` | ExLlamaV3-based inference engine with speculative decoding for 100 GB-class MoE packs on one 128 GB AMD Strix Halo (gfx1151, ROCm 7) | AI Inference |
 | `llama-cpp` | llama.cpp pinned to build b11439, built for Strix Halo (gfx1151) with ROCm + Vulkan backends | AI Inference |
 | `loop` | Corporate messenger for your team | Communication |
 | `radar` | Modern Kubernetes visibility — topology, event timeline, service traffic, resource browsing, Helm management, and GitOps support | Kubernetes |
@@ -199,6 +200,7 @@ nix-overlay/
 │   ├── haivemind/              # Multi-model AI consensus runner
 │   ├── hipengine/              # ROCm-native LLM inference engine for AMD GPUs
 │   ├── kubernetes-mcp-server/  # MCP server for Kubernetes and OpenShift
+│   ├── kyojin/                 # ExLlamaV3-based inference engine for Strix Halo (gfx1151)
 │   ├── loop/                   # Corporate messenger for your team
 │   ├── mcp-gateway/            # MCP protocol gateway
 │   ├── omniroute/              # Unified AI router with 160+ providers

packages/kyojin/default.nix

@@ -0,0 +1,12 @@
+{
+  pkgs,
+  inputs,
+  ...
+}:
+# The HIP extension has to be compiled against a ROCm build of PyTorch, and the
+# only ROCm torch in this flake with a cached binary is the one from the pinned
+# nixpkgs-torch211 input (2.11 + ROCm 7.2, python 3.13). Current nixos-unstable
+# ships torch 2.13, whose ROCm build fails to configure in nixpkgs and is not
+# on cache.nixos.org.
+inputs.nixpkgs-torch211.legacyPackages.${pkgs.stdenv.hostPlatform.system}.callPackage ./package.nix
+  { }

packages/kyojin/package.nix

@@ -0,0 +1,204 @@
+{
+  lib,
+  stdenv,
+  pkgs,
+  python3,
+  fetchFromGitHub,
+  makeWrapper,
+  rocmPackages,
+}:
+
+let
+  py = python3.pkgs;
+
+  # Kyojin is the Yamz fork of ExLlamaV3: a Python package plus a C++ extension
+  # that torch's cpp_extension cross-compiles to HIP code objects. Upstream
+  # only supports AMD Strix Halo, so the kernels are built for gfx1151 alone.
+  rocmArch = "gfx1151";
+
+  version = "1.3";
+
+  src = fetchFromGitHub {
+    owner = "Yamz-Labs";
+    repo = "kyojin";
+    rev = "v${version}";
+    hash = "sha256-/DPoqW/iaLXRAH8XF5RdE+K5YS4jZT6Sic6QeCDQMlU=";
+  };
+
+  # setup.py passes EXL3_ROCM_DEV_INCLUDE as -I to every translation unit. The
+  # AMD wheels ship one devel tree holding hipsparse/ and thrust/; nixpkgs
+  # splits them per package, so join the include dirs. torch's public HIP
+  # headers also pull hipsolver/, and it no longer vendors pybind11.
+  rocmDevInclude = pkgs.symlinkJoin {
+    name = "kyojin-rocm-dev-include";
+    paths = map lib.getDev (
+      [ py.pybind11 ]
+      ++ [
+        rocmPackages.clr
+        rocmPackages.hipblas
+        rocmPackages.hipblaslt
+        rocmPackages.hipcub
+        rocmPackages.hipfft
+        rocmPackages.hiprand
+        rocmPackages.hipsparse
+        rocmPackages.hipsolver
+        rocmPackages.rocblas
+        rocmPackages.rocprim
+        rocmPackages.rocrand
+        rocmPackages.rocsolver
+        rocmPackages.rocsparse
+        rocmPackages.rocthrust
+      ]
+    );
+  };
+
+  # torch's HIP path in cpp_extension expects one ROCM_HOME holding bin/hipcc,
+  # the headers and the amdgcn bitcode; nixpkgs ships them separately.
+  rocmToolkit = pkgs.symlinkJoin {
+    name = "kyojin-rocm-toolkit";
+    paths = [
+      rocmPackages."rocm-core"
+      rocmPackages."rocm-runtime"
+      rocmPackages."rocm-device-libs"
+      rocmPackages."rocm-comgr"
+      rocmPackages.clr
+      (lib.getBin rocmPackages.hipcc)
+      rocmPackages.rocblas
+      rocmPackages.hipblas
+      rocmPackages.hipblas-common
+      rocmPackages.hipblaslt
+      rocmPackages.hipsparse
+      rocmPackages.hipsolver
+      rocmPackages.hiprand
+      rocmPackages.rocrand
+      rocmPackages.hipfft
+      rocmPackages.rocsparse
+      rocmPackages.rocsolver
+      rocmPackages.rocthrust
+      rocmPackages.rocprim
+      rocmPackages.hipcub
+    ];
+    # The setup hooks of the joined packages would rewrite build flags; the
+    # compiler only needs the files.
+    postBuild = ''
+      rm -rf $out/nix-support
+    '';
+  };
+
+  # The library: exllamav3 plus the compiled exllamav3_ext module.
+  library = py.buildPythonPackage {
+    inherit src version;
+    pname = "kyojin";
+
+    format = "pyproject";
+    build-system = with py; [
+      setuptools
+      wheel
+    ];
+
+    nativeBuildInputs = [
+      py.ninja
+      rocmPackages."rocm-runtime"
+    ];
+
+    buildInputs = [ (lib.getBin rocmPackages.hipcc) ];
+
+    propagatedBuildInputs = with py; [
+      aiohttp
+      huggingface-hub
+      jinja2
+      llguidance
+      marisa-trie
+      numpy
+      pillow
+      pydantic
+      pyyaml
+      rich
+      safetensors
+      tokenizers
+      torchWithRocm
+      typing-extensions
+    ];
+
+    env = {
+      PYTORCH_ROCM_ARCH = rocmArch;
+      ROCM_PATH = "${rocmToolkit}";
+      ROCM_HOME = "${rocmToolkit}";
+      HIP_PATH = "${rocmToolkit}";
+      HIPCC = "${rocmToolkit}/bin/hipcc";
+      HIP_DEVICE_LIB_PATH = "${rocmPackages."rocm-device-libs"}/amdgcn/bitcode";
+      # Upstream build scripts look these up to find the ROCm devel tree.
+      EXL3_ROCM_SDK = "${rocmToolkit}";
+      EXL3_ROCM_DEV_INCLUDE = "${rocmDevInclude}/include";
+      # Default defines of upstream build.sh (tools/strix_halo/rebuild.sh).
+      EXL3_HIP_DEFINES = "EXL3_HIP_STG_PAD";
+    };
+
+    # setup.py pulls the extension source list and the arch helper with
+    # `from exllamav3...import ...`, which runs the package __init__ and
+    # therefore exllamav3/ext.py. That module JIT-compiles the extension when no
+    # precompiled exllamav3_ext is importable, i.e. during every wheel build
+    # (and it wants a writable $HOME for it). Register an empty `exllamav3`
+    # package first so only the two leaf modules are imported, and load
+    # build_config without the package wrapper.
+    postPatch = ''
+      sed -i '1i import sys as _sys, types as _types; _exl3_pkg = _types.ModuleType("exllamav3"); _exl3_pkg.__path__ = ["exllamav3"]; _sys.modules.setdefault("exllamav3", _exl3_pkg)' setup.py
+      sed -i 's|^from exllamav3\.exllamav3_ext\.build_config import get_sources as _get_sources$|import sys as _exl3_sys; _exl3_sys.path.insert(0, "exllamav3/exllamav3_ext"); from build_config import get_sources as _get_sources; _exl3_sys.path.pop(0)|' setup.py
+      grep -q 'from build_config import' setup.py
+    '';
+
+    # tests/ needs a real gfx1151 GPU and the published model packs.
+    doCheck = false;
+
+    meta = with lib; {
+      description = "Inference engine for 100 GB-class EXL3 MoE models on AMD Strix Halo (gfx1151, ROCm 7)";
+      homepage = "https://github.com/Yamz-Labs/kyojin";
+      changelog = "https://github.com/Yamz-Labs/kyojin/releases";
+      license = licenses.mit;
+      sourceProvenance = with sourceTypes; [ fromSource ];
+      # HIP code objects are gfx1151-only and ROCm is Linux-only.
+      platforms = platforms.linux;
+    };
+  };
+
+  python = py.python.withPackages (_: [ library ]);
+in
+stdenv.mkDerivation {
+  inherit src version;
+  pname = "kyojin";
+
+  # Only the wrappers are installed here: the serve scripts are plain scripts,
+  # they import their siblings by path, and the module itself is in `library`.
+  dontConfigure = true;
+  dontBuild = true;
+
+  nativeBuildInputs = [ makeWrapper ];
+
+  installPhase = ''
+    runHook preInstall
+
+    mkdir -p $out/bin
+    for model in glm mimo qwen; do
+      makeWrapper ${python}/bin/python $out/bin/kyojin-serve-$model \
+        --set PYTORCH_ROCM_ARCH ${rocmArch} \
+        --add-flags "${src}/tools/$model/serve.py"
+    done
+
+    runHook postInstall
+  '';
+
+  passthru = {
+    inherit python rocmArch;
+    pythonPackage = library;
+    category = "AI Inference";
+    updateScript = [
+      "nix-update"
+      "--flake"
+      ".#kyojin"
+    ];
+  };
+
+  meta = library.meta // {
+    mainProgram = "kyojin-serve-qwen";
+  };
+}