# bitnet.cpp Administrator & Architecture Guide

This guide is for engineers, architects, and operators who need to deploy, tune, extend, or reason about the internals of bitnet.cpp. It documents the system architecture, the build and conversion pipeline, the kernel layer, configuration surfaces, deployment, security posture, and known issues — at a depth sufficient to operate the framework in production or to reproduce its design.

For installation and day-to-day usage, see the companion **User Guide**.

---

## 1. Executive Summary

**bitnet.cpp** is the official inference framework for 1-bit large language models (notably BitNet b1.58, whose weights are ternary: −1, 0, +1). It delivers fast, lossless inference of 1.58-bit models on commodity CPUs, and — through a separate code path — on NVIDIA GPUs.

The defining architectural fact is that **bitnet.cpp is a thin Python orchestration layer wrapped around a heavy native engine**. The user-facing tools (`run_inference.py`, `run_inference_server.py`, `setup_env.py`) are small Python scripts that shell out to a compiled `llama-cli` / `llama-server` binary built from a bundled fork of [llama.cpp](https://github.com/ggerganov/llama.cpp). bitnet.cpp's own contribution is the **ternary GEMM/GEMV kernel layer** (`src/ggml-bitnet-*.cpp`, `include/`) integrated into ggml, plus a code-generation toolchain for lookup-table kernels and a model-conversion suite. The kernel methodology builds on [T-MAC](https://github.com/microsoft/T-MAC/).

> [!IMPORTANT]
> Because the engine is llama.cpp and the kernels are C++, most operational concerns — build difficulty, CPU-feature requirements (AVX2/NEON/DOTPROD), memory footprint, and the bulk of failure modes — originate in the native layer, not in the Python scripts. Operators must reason about the native boundary, not just the wrappers.

**Performance envelope (published):** ARM CPUs see 1.37×–5.07× speedups with 55.4%–70.0% energy reduction; x86 CPUs see 2.37×–6.17× speedups with 71.9%–82.2% energy reduction. A 100B BitNet b1.58 model runs on a single CPU at 5–7 tokens/second.

**License:** MIT. **Current framework version:** 1.0.

---

## 2. System Architecture

### 2.1 Layered architecture

bitnet.cpp follows the classic "small wrapper over a large native library" shape. Understanding which layer owns which behavior is essential for diagnosis.

```
   ┌─────────────────────────────────────────────────────────────┐
   │  User / your application                                     │
   └─────────────────────────────────────────────────────────────┘
                              │  (subprocess / CLI)
   ┌─────────────────────────────────────────────────────────────┐
   │  Python orchestration layer  (this project, ~thin)           │
   │   setup_env.py · run_inference.py · run_inference_server.py  │
   │   utils/*.py  (conversion, codegen, benchmarking, tuning)    │
   └─────────────────────────────────────────────────────────────┘
                              │  (builds & invokes)
   ┌─────────────────────────────────────────────────────────────┐
   │  Native inference engine  (bundled 3rdparty/llama.cpp + ggml)│
   │   llama-cli · llama-server · llama-quantize · llama-bench    │
   └─────────────────────────────────────────────────────────────┘
                              │  (calls into)
   ┌─────────────────────────────────────────────────────────────┐
   │  BitNet ternary kernel layer  (this project's C++)           │
   │   src/ggml-bitnet-mad.cpp   (native I2_S GEMM/GEMV)          │
   │   src/ggml-bitnet-lut.cpp   (TL1/TL2 lookup-table kernels)   │
   │   include/ggml-bitnet.h · include/gemm-config.h             │
   └─────────────────────────────────────────────────────────────┘
                              │  (dispatches per CPU ISA)
   ┌─────────────────────────────────────────────────────────────┐
   │  CPU SIMD: x86 AVX2 · ARM NEON · ARM DOTPROD                  │
   └─────────────────────────────────────────────────────────────┘
```

A wholly separate stack lives under `gpu/` (covered in [Section 11](#11-gpu-inference-subsystem)):

```
   gpu/generate.py (FastGen) → gpu/model.py (BitLinear/Attention)
        → gpu/bitnet_kernels/libbitnet.so (custom W2A8 CUDA kernel)
        → CUDA / xformers / PyTorch
```

### 2.2 Component inventory

| Component | Path | Role |
|---|---|---|
| Build orchestrator | `setup_env.py` | Installs GGUF helper, generates/copies kernels, runs CMake, downloads + converts + quantizes the model. |
| CLI inference | `run_inference.py` | Builds and runs a `llama-cli` command for CPU inference. |
| HTTP server | `run_inference_server.py` | Builds and runs a `llama-server` command with continuous batching. |
| Ternary kernels (native I2_S) | `src/ggml-bitnet-mad.cpp` | I2_S quantization + parallel `vec_dot` GEMM/GEMV for AVX2 / NEON / DOTPROD. |
| Ternary kernels (LUT) | `src/ggml-bitnet-lut.cpp` | TL1/TL2 lookup-table kernel glue, gated by build options. |
| Kernel public header | `include/ggml-bitnet.h` | The C ABI bitnet exposes into ggml (`ggml_bitnet_mul_mat_*`, `ggml_qgemm_lut`, etc.). |
| Kernel tuning config | `include/gemm-config.h` | Compile-time block-size / parallelism constants. |
| Kernel codegen | `utils/codegen_tl1.py`, `utils/codegen_tl2.py` | Generate `bitnet-lut-kernels.h` for a given model's matmul shapes. |
| Pre-tuned kernels | `preset_kernels/<model>/` | Shipped `bitnet-lut-kernels-tl{1,2}.h` + `kernel_config_tl{1,2}.ini` per model. |
| Model conversion | `utils/convert*.py` | HF / Microsoft / safetensors → GGUF converters (largely llama.cpp-derived). |
| Embedding tools | `utils/quantize_embeddings.py` | Quantize + benchmark the embedding layer. |
| Accuracy / perf testing | `utils/test_perplexity.py`, `utils/e2e_benchmark.py`, `utils/tune_gemm_config.py` | Perplexity, throughput, and auto-tuning. |
| GPU subsystem | `gpu/` | Independent CUDA W2A8 inference path (see §11). |
| Native engine | `3rdparty/llama.cpp` | Bundled submodule providing the actual transformer runtime + GGUF I/O. |

### 2.3 What bitnet.cpp does *not* do (negative knowledge)

> [!CAUTION]
> The following are explicitly **out of scope**. Documenting them as features would mislead operators; they are deliberately absent.

- **It does not train or fine-tune models.** It is an inference framework only. The ternary weights are produced elsewhere; bitnet.cpp consumes them.
- **It does not provide GPU acceleration in the primary CPU path.** `run_inference.py` and `run_inference_server.py` always pass `-ngl 0` (zero GPU layers). GPU inference is the separate `gpu/` subsystem with different code, deps, and models.
- **It has no built-in authentication, TLS, or rate limiting** on the server. Those are inherited from llama.cpp's `llama-server` and must be supplied by a reverse proxy.
- **It does not support architectures other than x86-64 and ARM64.** The build aborts with `Arch <name> is not supported yet` otherwise.
- **It does not currently support NPUs.** NPU support is stated as a future goal.
- **The I2_S kernel has no scalar fallback** for x86 CPUs lacking AVX2 — output is incorrect rather than slow on such hardware (see §13, issue #547).

---

## 3. System Requirements

### 3.1 CPU path

| Requirement | Version | Notes |
|---|---|---|
| Python | ≥ 3.9 | Orchestration + converters. |
| CMake | ≥ 3.22 | Build driver. (`CMakeLists.txt` declares `cmake_minimum_required(VERSION 3.14)`, but the documented requirement is 3.22.) |
| Clang / Clang++ | ≥ 18 | The build sets `-DCMAKE_C_COMPILER=clang -DCMAKE_CXX_COMPILER=clang++`. GCC is not the supported toolchain. |
| conda | — | Recommended for environment isolation. |
| CPU ISA | AVX2 (x86) or NEON (ARM) | DOTPROD on ARM is used when available for extra speed. **AVX2 is effectively mandatory on x86** for correct I2_S output. |
| RAM | Model-dependent | The full model must fit in memory; this is the practical size ceiling. |

### 3.2 GPU path (separate)

| Requirement | Version | Notes |
|---|---|---|
| Python | < 3.13 | `gpu/` recommends `conda create --name bitnet-gpu "python<3.13"`. |
| NVIDIA GPU | Compute capability 8.0 | `compile.sh` targets `arch=compute_80` (Ampere, e.g. A100). |
| CUDA toolkit | nvcc with C++17 | Used to build `libbitnet.so`. |
| PyTorch | ≥ 2.2.0 | From `gpu/requirements.txt`. |
| xformers | ≥ 0.0.22 | Provides RoPE, RMSNorm, and memory-efficient (flash) attention. |
| Other Python deps | — | `fire`, `sentencepiece`, `tiktoken`, `blobfile`, `flask`, `einops`, `transformers`. |

> [!NOTE]
> The CPU and GPU stacks should be kept in **separate conda environments**. They pin different Python versions (CPU uses 3.9; GPU requires < 3.13) and have disjoint dependency sets.

---

## 4. The Build & Preparation Pipeline

`setup_env.py` is the heart of operations. Its `main()` runs four phases in strict order. Knowing this sequence is what lets you diagnose a failed build.

```python
def main():
    setup_gguf()     # 1. pip install 3rdparty/llama.cpp/gguf-py
    gen_code()       # 2. generate or copy the compute kernels
    compile()        # 3. cmake configure + build (Release)
    prepare_model()  # 4. download + convert HF→GGUF + quantize
```

### 4.1 Phase 1 — `setup_gguf()`

Installs the GGUF Python package from the bundled engine:

```bash
python -m pip install 3rdparty/llama.cpp/gguf-py
```

This provides the Python-side GGUF writer used by the converters.

### 4.2 Phase 2 — `gen_code()`

Selects or generates the BitNet compute kernels based on architecture, model, and `--quant-type`. Two sub-behaviors:

- **`--use-pretuned`:** copies the shipped kernel from `preset_kernels/<model_name>/` into `include/bitnet-lut-kernels.h` (and `include/kernel_config.ini` for ARM). Aborts with `Pretuned kernels not found for model <name>` if absent.
- **Otherwise:** invokes `utils/codegen_tl1.py` (ARM) or `utils/codegen_tl2.py` (x86) with model-specific block-size arguments. For example, `bitnet_b1_58-large` on ARM:

  ```bash
  python utils/codegen_tl1.py --model bitnet_b1_58-large --BM 256,128,256 --BK 128,64,128 --bm 32,64,32
  ```

The per-model `--BM/--BK/--bm` tuples are hard-coded in `gen_code()`. Models not enumerated there raise `NotImplementedError`. Note that `BitNet-b1.58-2B-4T` reuses the `bitnet_b1_58-3B` codegen parameters.

> [!NOTE]
> The codegen logic (`docs/codegen.md`) cuts each weight matrix into compute blocks sized to the hardware. TL1 requires `M % BM == 0`, `K % BK == 0`, `BM % bm == 0`, and `bm ∈ {32, 64}`. TL2 additionally splits K because it needs `BK % 6 == 0`, computing part in a TL2 path and part in a TL1 path; for TL2, `bm ∈ {32}`.

### 4.3 Phase 3 — `compile()`

Checks for CMake, then configures and builds:

```bash
cmake -B build -DBITNET_ARM_TL1=OFF      # (or -DBITNET_X86_TL2=OFF on x86)
      [-T ClangCL on Windows]
      -DCMAKE_C_COMPILER=clang -DCMAKE_CXX_COMPILER=clang++
cmake --build build --config Release
```

> [!IMPORTANT]
> `setup_env.py` passes the architecture's TL option as **OFF** (`COMPILER_EXTRA_ARGS = {"arm64": ["-DBITNET_ARM_TL1=OFF"], "x86_64": ["-DBITNET_X86_TL2=OFF"]}`). This means the **default build uses the native I2_S GEMM/GEMV path** (`src/ggml-bitnet-mad.cpp`). The TL1/TL2 lookup-table kernels (`src/ggml-bitnet-lut.cpp`) are compiled only when you explicitly turn the corresponding option `ON`. Plan tuning and benchmarking around the I2_S path as the default reality.

### 4.4 Phase 4 — `prepare_model()`

Downloads (if `--hf-repo` is given) and converts to GGUF, then quantizes:

- **TL types:** `python utils/convert-hf-to-gguf-bitnet.py <model_dir> --outtype <tl1|tl2> --quant-embd`
- **I2_S:** convert to f32 first, then quantize with the built `llama-quantize`:

  ```bash
  ./build/bin/llama-quantize [--token-embedding-type f16] \
      <model>/ggml-model-f32.gguf <model>/ggml-model-i2_s.gguf I2_S 1 [1]
  ```

The output is `ggml-model-<quant_type>.gguf`. If it already exists and is non-empty, this phase is skipped.

### 4.5 Build logging

Every step runs through `run_command(..., log_step=<name>)`, which redirects stdout/stderr to `<log-dir>/<name>.log`. The step names are your debugging map: `download_model`, `install_gguf`, `convert_to_tl`, `convert_to_f32_gguf`, `quantize_to_i2s`, `codegen`, `generate_build_files`, `compile`.

> [!TIP]
> When a build fails, the console only prints a one-line `Error occurred while running command`. Open `<log-dir>/<step>.log` for the real compiler/converter output.

---

## 5. Configuration Reference

bitnet.cpp's configuration surface is **compile-time constants + CLI flags + CMake options**, not a YAML/JSON config file. There is no runtime config file.

### 5.1 CMake options (`CMakeLists.txt`)

| Option | Default | Effect |
|---|---|---|
| `BITNET_ARM_TL1` | `OFF` | Build the ARM TL1 lookup-table kernel path (`add_compile_definitions(GGML_BITNET_ARM_TL1)`). |
| `BITNET_X86_TL2` | `OFF` | Build the x86 TL2 lookup-table kernel path (`GGML_BITNET_X86_TL2`). |
| `LLAMA_BUILD_SERVER` | `ON` (forced) | Ensures `llama-server` is built. |
| `CMAKE_BUILD_TYPE` | `Release` | Forced to Release on non-Xcode/non-MSVC when unset. |

The build adds `src/` then `3rdparty/llama.cpp` as subdirectories; on GNU compilers it adds `-fpermissive`. C standard is C11; threads use the pthread flag.

### 5.2 Kernel tuning constants (`include/gemm-config.h`)

These control the I2_S GEMM/GEMV tiling and parallelism. They are selected at compile time per ISA. The header defines `ACT_PARALLEL` (activation-parallel mode) and, when set, the x86/AVX defaults are:

```c
#define ROW_BLOCK_SIZE 4
#define COL_BLOCK_SIZE 128
#define PARALLEL_SIZE 4
```

| Constant | Meaning |
|---|---|
| `ROW_BLOCK_SIZE` | Rows processed per tiling block. |
| `COL_BLOCK_SIZE` | Columns processed per tiling block. |
| `PARALLEL_SIZE` | Degree of weight/activation parallelism in the `vec_dot` kernel. |
| `ACT_PARALLEL` | When defined, selects activation-parallel kernels (default; recommended for I2_S because it amortizes weight-unpack cost). |

Defaults differ by ISA (AVX/SSSE3, ARM NEON, ARM DOTPROD) — see the `#if defined(...)` ladder in the header. Modify these to match your CPU's cache, then rebuild.

> [!TIP]
> Rather than editing constants by hand, use `utils/tune_gemm_config.py` to auto-search parallelism degree ∈ {2,4,8}, row block ∈ {2,4,8,16,32}, and column block ∈ {32,64,128,256,512,1024}. On an AMD EPYC 7V13 (x86, 8 threads, pp128), the published optimum was parallelism = 4, row block = 4, column block = 128 — i.e. the shipped defaults.

### 5.3 Inference flags

`run_inference.py` forwards a fixed flag set to `llama-cli`: `-m`, `-n`, `-t`, `-p`, `-ngl 0`, `-c`, `--temp`, `-b 1`, and `-cnv` when conversation mode is on. `run_inference_server.py` forwards `-m`, `-c`, `-t`, `-n`, `-ngl 0`, `--temp`, `--host`, `--port`, `-cb` (continuous batching), and an optional `-p`. See the User Guide for the full option tables.

> [!NOTE]
> These wrappers expose a curated subset of llama.cpp's flags. Advanced operators can invoke `build/bin/llama-cli` / `llama-server` directly to access the engine's full flag surface (sampling controls, grammar, LoRA, etc.), pointing at the same `ggml-model-*.gguf` file.

---

## 6. The Kernel Layer (Internals)

This section documents bitnet.cpp's actual technical contribution: the ternary matrix-multiply kernels integrated into ggml.

### 6.1 Public ABI (`include/ggml-bitnet.h`)

bitnet exposes a C ABI that ggml calls into. Key entry points:

| Function | Purpose |
|---|---|
| `ggml_bitnet_init` / `ggml_bitnet_free` | Allocate/free the `bitnet_tensor_extra` pool (`GGML_BITNET_MAX_NODES`). |
| `ggml_bitnet_can_mul_mat(src0, src1, dst)` | Decide whether a matmul is eligible for the bitnet kernel (supported quant type, F32 activations/output, CPU backend). |
| `ggml_bitnet_mul_mat_get_wsize(...)` | Compute the scratch workspace size needed. |
| `ggml_bitnet_mul_mat_task_init` / `_task_compute` | Two-phase matmul: build the lookup tables, then compute. |
| `ggml_bitnet_transform_tensor` | Repack a weight tensor into the kernel's layout. |
| `ggml_bitnet_get_type_bits(type)` | Bits per weight: TL1/TL2 → 2, Q4_0 → 4. |
| `ggml_qgemm_lut`, `ggml_preprocessor` | The TL1 (ARM) / TL2 (x86) lookup-table GEMM entry points (signatures differ per ISA). |

The `bitnet_tensor_extra` struct carries `lut_scales_size`, `BK`, `n_tile_num`, packed `qweights`, and `scales`. On ARM with NEON, `bitnet_float_type` is `float32_t`; elsewhere it's `float`.

### 6.2 Native I2_S path (`src/ggml-bitnet-mad.cpp`)

This is the default-build kernel. It implements:

- **`quantize_i2_s(...)`** — packs f32 weights into 2-bit ternary form. Each value maps to {0,1,2} (→ −1,0,+1), four packed per byte; a per-tensor scale (the max absolute value) is stored after the packed data. Block size `QK_I2_S` is 128 on AVX/SSSE3 and 64 on NEON.
- **A family of `ggml_vec_dot_i2_i8_s_*` kernels** — the ternary-weight × int8-activation dot products, with several parallel variants:
  - `_1x1` — single row.
  - `_1x4_32W` / `_1xN` — multiple weight rows per call (weight-parallel).
  - `_Nx1` — multiple activation columns per call (activation-parallel).
  - `ggml_vec_dot_i2_i8_s` — the dispatcher: when `nrc % PARALLEL_SIZE == 0` it picks the `Nx1` (activation-parallel, if `ACT_PARALLEL`) or `1xN` (weight-parallel) kernel; otherwise the scalar `1x1`.

Each variant has three ISA implementations behind `#if defined(...)`: **AVX2** (`_mm256_maddubs_epi16` + `_mm256_madd_epi16` accumulation), **ARM NEON** (`vmlal_s8`), and **ARM DOTPROD** (`vdotq_s32`, the fastest path when the CPU supports it). There is also a LoongArch LASX path for the horizontal sum.

> [!WARNING]
> The AVX2 kernel bodies are guarded by `#if defined(__AVX2__)`. On an x86 build for a CPU without AVX2, the I2_S kernel compiles but its inner loop is empty, yielding incorrect output. This is the root cause of issue #547. Treat AVX2 as a hard requirement on x86.

### 6.3 LUT path (`src/ggml-bitnet-lut.cpp`)

Compiled only when `GGML_BITNET_ARM_TL1` or `GGML_BITNET_X86_TL2` is defined. Provides the `ggml_bitnet_init/free`, `can_mul_mat`, `get_wsize`, and `get_type_bits` implementations for the lookup-table kernels (`GGML_TYPE_TL1`/`GGML_TYPE_TL2`). The actual LUT GEMM bodies are the generated `bitnet-lut-kernels.h` produced by codegen.

### 6.4 Optimization summary (from `src/README.md`)

The current CPU optimization release adds:

- **Parallel weight & activation computation** in the W2A8 `vec_dot` kernel (both x86 and ARM).
- **Native I2_S GEMM & GEMV** integrated into ggml so it composes with the llama.cpp compute graph.
- **Configurable tiling & parallelism** via `gemm-config.h`.
- **Embedding quantization** to Q6_K.

Measured (AMD EPYC 7V13, x86, 1 thread): for a `[128,2048]×[2048,2048]` matmul, activation-parallel achieves 5.805 ms vs 10.820 ms with no parallelism — roughly 1.86× from parallelism alone. The release reports 1.15×–2.1× additional speedup over the original implementation across hardware.

---

## 7. Embedding Quantization

The token-embedding layer is quantized independently of the ternary weights to trade a small accuracy cost for memory and speed.

- **Automatic:** `python setup_env.py --quant-embd` converts embeddings to Q6_K during preparation.
- **Manual:**

  ```bash
  build/bin/llama-quantize --token-embedding-type Q6_K \
    models/BitNet-b1.58-2B-4T/ggml-model-f32.gguf \
    models/BitNet-b1.58-2B-4T/ggml-model-i2_s-embed-q6_k.gguf I2_S 1 1
  ```

Perplexity comparison across embedding formats (BitNet-b1.58-2B-4T, TG128 — lower is better):

| Embedding | Wikitext | PTB | LAMBADA | IMDB | AG NEWS |
|---|---|---|---|---|---|
| F32 | 17.109 | 33.086 | 43.285 | 29.302 | 36.769 |
| F16 | 17.109 | 33.086 | 43.285 | 29.302 | 36.769 |
| Q8_0 | 17.120 | 33.118 | 43.289 | 29.313 | 36.774 |
| **Q6_K** | 17.149 | 33.220 | 43.305 | 29.349 | 36.797 |
| Q5_0 | 17.238 | 33.244 | 43.463 | 29.548 | 36.854 |
| Q4_0 | 17.353 | 33.775 | 44.455 | 30.104 | 37.399 |
| Q3_K | 17.643 | 34.391 | 45.459 | 30.848 | 39.569 |
| I2_S | N/A | N/A | N/A | N/A | N/A |

> [!IMPORTANT]
> **Q6_K is the recommended embedding format** — it preserves perplexity almost exactly while saving memory and improving throughput. Quantizing embeddings to **I2_S fails** (model failure due to extreme quantization — the N/A row). Do not attempt I2_S embeddings in production.

Use `utils/quantize_embeddings.py` (options `-i/--input`, `-o/--output-dir`, `-q/--quantize-bin`, `-b/--bench-bin`, `--stats-dir`, `-c/--csv-output`, `-t/--types`, `-s/--skip-existing`) to quantize and benchmark embedding formats systematically.

---

## 8. Model Conversion Toolchain

The `utils/` converters are largely derived from llama.cpp's GGUF conversion machinery, specialized for BitNet weight layouts. They are normally invoked by `setup_env.py`, but operators bringing new models will call them directly.

| Script | Key options | Notes |
|---|---|---|
| `convert-helper-bitnet.py` | (model dir) | One-shot helper; the documented path for `.safetensors` BitNet checkpoints. |
| `convert-hf-to-gguf-bitnet.py` | `--outtype {f32,f16,tl1,tl2}`, `--quant-embd`, `--vocab-only`, `--awq-path`, `--outfile`, `--bigendian`, `--use-temp-file`, `--model-name`, `--verbose` | HF directory → GGUF. |
| `convert-ms-to-gguf-bitnet.py` | `--outtype`, `--vocab-type {spm,bpe,hfft}`, `--vocab-dir`, `--ctx`, `--pad-vocab`, `--skip-unknown`, `--dump`, `--no-vocab`, … | Microsoft-format checkpoints (`.pth/.pt/.bin`) → GGUF. |
| `convert.py` | (same family as above) | General GGUF converter (safetensors only for the model body). |
| `generate-dummy-bitnet-model.py` | `--outtype`, `--model-size`, `--outfile` | Synthetic model of a given layout for benchmarking. |
| `preprocess-huggingface-bitnet.py` | `--input`, `--output` | Pre-process HF weights into the expected layout. |

> [!NOTE]
> Many converter error messages are inherited from llama.cpp (e.g. `Can not map tensor`, `Unsupported vocabulary type`, `Cannot find Llama BPE tokenizer`). When conversion fails, the message usually points at a tokenizer or tensor-naming mismatch in the *source* checkpoint, not a bug in bitnet.cpp. Use `--skip-unknown` to bypass unexpected tensors and `--vocab-type` to force a tokenizer family.

---

## 9. Deployment & Operations

### 9.1 Deploying the CPU server

```bash
# Build once
python setup_env.py -md models/BitNet-b1.58-2B-4T -q i2_s --quant-embd

# Run the server (bind localhost, front with a proxy for anything public)
python run_inference_server.py -m models/BitNet-b1.58-2B-4T/ggml-model-i2_s.gguf \
  -t $(nproc) -c 4096 --host 127.0.0.1 --port 8080
```

Deployment checklist:

- [ ] Cloned with `--recursive` (submodule present).
- [ ] Clang ≥ 18, CMake ≥ 3.22 on `PATH`; on x86, CPU reports AVX2.
- [ ] Model built and `ggml-model-*.gguf` present and non-empty.
- [ ] `--threads` set to physical core count.
- [ ] Server bound to `127.0.0.1`; reverse proxy (nginx/Caddy) terminating TLS and adding auth if exposed.
- [ ] Memory headroom ≥ model size + KV cache for `--ctx-size`.

### 9.2 Process management

The scripts install a `SIGINT` handler for clean shutdown ("Ctrl+C pressed, exiting…"). Under a service manager, send `SIGINT`/`SIGTERM` for graceful stop. For systemd, run the `run_inference_server.py` command as the unit `ExecStart` inside the activated conda environment; a community request exists for a Debian package with systemd socket activation (issue #557).

### 9.3 Tuning for throughput

1. Set `--threads` to physical cores (not hyperthreads).
2. Enable embedding quantization (`--quant-embd`, Q6_K).
3. Run `utils/tune_gemm_config.py` on the target host and apply the winning `gemm-config.h` values, then rebuild.
4. Prefer ARM DOTPROD hardware where available — it activates the fastest NEON path automatically.

### 9.4 Capacity model

The dominant memory cost is the resident model plus the KV cache, which scales with `--ctx-size`. Throughput scales with thread count up to memory-bandwidth saturation; published per-host curves (AMD EPYC 7V13, Intel i7-13800H, Cobalt 100) show diminishing returns past the physical core count.

---

## 10. Security

> [!WARNING]
> bitnet.cpp is an inference framework with no native authentication, authorization, encryption, or sandboxing. Its security posture is almost entirely a function of how you deploy the native server and which model weights you load.

| Concern | Posture | Mitigation |
|---|---|---|
| Network exposure | Server has no auth/TLS; defaults to `127.0.0.1`. | Keep localhost binding; front with an authenticating TLS reverse proxy before exposing. |
| Untrusted models | GGUF and checkpoint files are loaded and (for `.pt`) deserialized. The GPU path uses `torch.load(..., weights_only=True)`, which limits pickle execution; the CPU GGUF path does not execute code but trusts tensor metadata. | Only load weights from trusted sources; treat third-party `.pt`/`.gguf` files as untrusted input. |
| Build supply chain | Pulls a pinned `3rdparty/llama.cpp` submodule and converter dependencies. | Pin the submodule commit; rebuild from a clean, verified checkout. Avoid advancing the submodule past a tested point (see §13, std::chrono break). |
| Resource exhaustion | Long `--n-predict` / large `--ctx-size` can exhaust CPU/RAM. | Bound `-n` and `-c`; enforce request limits at the proxy. |
| Secrets | No secrets, API keys, or credentials are used or stored by the framework. | N/A — there is no secret material to manage. |

---

## 11. GPU Inference Subsystem

The `gpu/` directory is a self-contained inference stack, distinct from the CPU framework in every respect.

### 11.1 Architecture

```
generate.py (FastGen)
  ├─ builds two models from model.py:
  │    prefill_model  (use_kernel=False → BitLinear, F.linear)
  │    decode_model   (use_kernel=True  → BitLinearKernel → libbitnet.so)
  ├─ CUDA graphs for prefill and decode (replayed per step)
  ├─ xformers: rope_padded, RMSNorm, memory_efficient_attention (flash)
  └─ tokenizer.py (tiktoken-based Llama-3 tokenizer + ChatFormat)
```

`model.py` defines a Llama-style transformer with **squared-ReLU FFN** and **sub-norms** (`attn_sub_norm`, `ffn_sub_norm`). Two linear layers exist:

- `BitLinear` — quantizes activations to int8 then runs a normal `F.linear` (used for prefill).
- `BitLinearKernel` — calls `bitnet_int8xint2_linear`, which invokes `libbitnet.so`'s `bitlinear_int8xint2` via `ctypes` (used for decode).

Default model config (`ModelArgs`): dim 2560, 30 layers, 20 heads, 5 KV heads, vocab 128256, FFN 6912, RoPE θ 500000, bf16.

### 11.2 The CUDA kernel (`gpu/bitnet_kernels/bitnet_kernels.cu`)

`ladder_int8xint2_kernel` performs W2A8 GEMV. It decodes 16 packed 2-bit weights at a time into int8 using a `lop3.b32` bit trick (`decode_i2s_to_i8s`), then accumulates with the `dp4a` (`__dp4a`) instruction — a 4-element int8 dot product into int32 — followed by a warp-shuffle reduction. Built with:

```bash
nvcc -std=c++17 --compiler-options -fPIC --shared bitnet_kernels.cu \
     -lcuda -gencode=arch=compute_80,code=compute_80 -o libbitnet.so
```

### 11.3 Weight packing (`gpu/pack_weight.py`, `gpu/convert_checkpoint.py`)

`convert_checkpoint.py` quantizes each linear weight to int8 ternary and a bf16 scale, then `convert_weight_int8_to_int2` permutes (16×32 blocks), compresses 4 values per int8, and interleaves with the pattern `[0,4,8,12,1,5,9,13,2,6,10,14,3,7,11,15]` to accelerate decoding. It emits two checkpoints: `model_state_int2.pt` (for the kernel decode path) and `model_state_fp16.pt` (for prefill).

### 11.4 GPU performance (A100 40GB)

| Shape (N×K) | W2A8 (µs) | BF16 (µs) | Speedup |
|---|---|---|---|
| 2560 × 2560 | 13.32 | 18.32 | 1.38 |
| 13824 × 2560 | 18.75 | 59.51 | 3.17 |
| 3200 × 10240 | 19.64 | 60.79 | 3.10 |
| 20480 × 3200 | 30.99 | 112.39 | 3.63 |

End-to-end, BitNet-b1.58-2B with this kernel runs ~3× faster than a similarly sized BF16 model (Gemma-2-2B via vLLM).

> [!CAUTION]
> The GPU kernel is hard-coded to `compute_80` and tuned for the 2B model layout. It is not a general-purpose GPU backend, and it does not share code or models with the CPU `.gguf` path. Treat it as an experimental, model-specific accelerator.

---

## 12. Repository Structure

```
BitNet/
├── setup_env.py                 # build + convert orchestrator
├── run_inference.py             # CPU CLI inference (→ llama-cli)
├── run_inference_server.py      # CPU HTTP server (→ llama-server)
├── requirements.txt             # CPU Python deps (re-exports llama.cpp convert reqs)
├── CMakeLists.txt               # top-level build (adds src/ + 3rdparty/llama.cpp)
├── include/
│   ├── ggml-bitnet.h            # kernel C ABI exposed into ggml
│   └── gemm-config.h            # compile-time tiling/parallelism constants
├── src/
│   ├── ggml-bitnet-mad.cpp      # native I2_S GEMM/GEMV (default build)
│   ├── ggml-bitnet-lut.cpp      # TL1/TL2 LUT kernel glue (opt-in)
│   ├── CMakeLists.txt
│   └── README.md                # CPU optimization writeup + benchmarks
├── utils/
│   ├── codegen_tl1.py / codegen_tl2.py   # LUT kernel codegen
│   ├── convert-*.py                       # GGUF converters
│   ├── quantize_embeddings.py             # embedding quant + bench
│   ├── test_perplexity.py                 # accuracy
│   ├── e2e_benchmark.py                    # throughput
│   ├── tune_gemm_config.py                 # auto-tune tiling
│   └── generate-dummy-bitnet-model.py      # synthetic models
├── preset_kernels/<model>/      # pre-tuned LUT kernels + .ini per model
├── gpu/                         # independent CUDA W2A8 inference stack
│   ├── generate.py, model.py, tokenizer.py, pack_weight.py, ...
│   └── bitnet_kernels/{bitnet_kernels.cu,.h,compile.sh,setup.py}
├── docs/codegen.md             # TL1/TL2 codegen explanation
└── 3rdparty/llama.cpp          # bundled engine (submodule)
```

---

## 13. Known Issues & Operational Caveats

Triaged by layer (most "BitNet is broken" reports are native-layer issues bitnet.cpp inherits):

| Issue | Layer | Symptom & guidance |
|---|---|---|
| #547 | BitNet kernel | I2_S produces garbage on x86 without AVX2 (no scalar fallback). **Require AVX2 on x86**, or use TL2. |
| #348 | BitNet kernel | Incorrect / random output — frequently the same AVX2 root cause or a model/quant mismatch. |
| FAQ Q1 | llama.cpp (submodule) | Build dies in `log.cpp` on `std::chrono`. Apply the upstream `tinglou/llama.cpp` `4e3db1e3` fix; pin the submodule. |
| #554, #513 | Build / toolchain (Windows) | CMake cannot locate `clang`/`clang++`. Build from a VS 2022 Developer Prompt; initialize with `VsDevCmd.bat` / `Enter-VsDevShell`. |
| #461 | llama.cpp build | `llama-cli.exe` not generated because `LLAMA_BUILD_EXAMPLES` was off by default. Rebuild from a clean checkout at the documented commit. |
| #407, #509 | Build / toolchain | Compiler errors (e.g. Fedora 39 + clang 17). Use Clang ≥ 18 as specified. |
| #532 | Model / engine | Server crashes after the first response with Falcon3-10B-Instruct-1.58bit. Prefer a smaller Falcon3 or BitNet-2B; report configuration upstream. |
| #556, #561 | Platform | Native Windows-on-ARM and Raspberry Pi 5 build notes exist as community guides; expect manual steps outside the documented x86/ARM Linux/macOS paths. |
| #560 | Tooling | `huggingface-cli` is being deprecated in favor of `hf`; download commands may need updating over time. |

### Tech debt & forward-looking notes

- **Codegen model list is hard-coded.** New models require adding a `--BM/--BK/--bm` branch in `setup_env.py:gen_code()` or they raise `NotImplementedError`.
- **TL1/TL2 are opt-in but default-off.** The shipped build path is I2_S; LUT kernels need an explicit `-DBITNET_*_TL*=ON` and matching codegen/quant type. Document which path your deployment actually compiled.
- **`deploy`/packaging story is informal.** No official systemd unit or distro package yet (#557 open).
- **NPU support** is announced as future work; there is no NPU path in-tree today.

---

## 14. Quick Reference

### 14.1 Architecture-to-quant matrix

| Arch aliases | Allowed quant types | Default kernel built |
|---|---|---|
| `AMD64` / `x86` / `x86_64` | `i2_s`, `tl2` | I2_S (`ggml-bitnet-mad.cpp`) |
| `aarch64` / `arm64` / `ARM64` | `i2_s`, `tl1` | I2_S (`ggml-bitnet-mad.cpp`) |

### 14.2 Build & run

| Action | Command |
|---|---|
| Configure + build (implicit) | `python setup_env.py -md <dir> -q i2_s` |
| Build with embedding quant | `python setup_env.py -md <dir> -q i2_s --quant-embd` |
| Build TL kernels (advanced) | `cmake -B build -DBITNET_X86_TL2=ON …` then `cmake --build build --config Release` |
| CLI inference | `python run_inference.py -m <gguf> -p "<prompt>" -t <cores> [-cnv]` |
| Server | `python run_inference_server.py -m <gguf> -t <cores> --host 127.0.0.1 --port 8080` |
| Auto-tune tiling | `python utils/tune_gemm_config.py --model <gguf> --threads <n>` |
| Perplexity | `python utils/test_perplexity.py -m <gguf> -d data` |

### 14.3 Key files for operators

| Need | File |
|---|---|
| Change build options | `CMakeLists.txt` |
| Tune kernel tiling | `include/gemm-config.h` (or `tune_gemm_config.py`) |
| Understand the kernel ABI | `include/ggml-bitnet.h` |
| Debug a failed build | `<log-dir>/<step>.log` |
| Add a new model's kernels | `setup_env.py:gen_code()` + `utils/codegen_tl{1,2}.py` |

---

_Generated June 2026 from microsoft/BitNet at 01eb415. AI-generated documentation. The library is checked by sampling (three repositories so far); this guide was not individually reviewed. Verify against the project before relying on it. Unofficial; not affiliated with or reviewed by the project._