# Apsi on NVIDIA GPUs

Everything Apsi does on a graphics card, in one place: running language models, the weight formats, training,
serving, temperature control, and what to do when something goes wrong. Apsi needs **only the NVIDIA display
driver** — no CUDA toolkit, no PyTorch, nothing to download. Its kernels are generated by Apsi itself (LLVM → PTX)
and the driver compiles them for whatever card is present.

| Card | What works |
|---|---|
| GTX 16xx, RTX 20xx (Turing, compute 7.5) | everything; matrix products on the float32 kernels (no bfloat16 tensor cores on Turing) |
| RTX 30xx, A100 (Ampere, 8.0 / 8.6) | everything; tensor cores for prompts and training |
| RTX 40xx, L4, H100, RTX 50xx, B200 (Ada / Hopper / Blackwell, 8.9+) | everything; also fp8 products on the tensor cores with `backend="torch"` |
| Older than 7.5, AMD, Intel, Apple | the processor path (or `backend="torch"` with a PyTorch build for that hardware) |

`apsi doctor` shows the card, the driver and what Apsi will use. `hub.devices()` from a program.

### Quick start: train on your card in three steps

1. **Once**, give Apsi the right to set the card's power limit (it holds the temperature this way, at full and
   steady speed). The Windows and Linux installers do it for you: keep **Full GPU speed without administrator
   rights** ticked and answer *Yes* once. Otherwise: **Apsi GPU Tuner → Apply** from the download page, or
   `apsi gpu setup` from a terminal opened as administrator.
2. Train with `backend="gpu"` (the example in section 2). Nothing else to install.
3. Read the report: the speed of every epoch, the temperature, and how it was held. `apsi gpu` shows the card at any
   time, `apsi gpu watch` every second.

| Section | What it covers |
|---|---|
| [1. Language models](#1-language-models-use-hub) | running Hugging Face models on the card, weight formats, tensor cores, threads |
| [2. Training](#2-training-on-the-gpu-with-apsi-alone-backendgpu) | layers, precisions, measured speed against PyTorch, several cards, model size |
| [3. Serving](#3-serving-the-openai-api) | the OpenAI API in front of a model, queues, metrics |
| [4. Temperature control](#4-temperature-control) | the 77 °C control, the power limit, the GPU Tuner, `apsi gpu hold` |
| [5. Images, video, your own kernels](#5-images-video-and-your-own-kernels-gpu) | `@gpu` functions |
| [6. Troubleshooting](#6-troubleshooting) | slow training, out of memory, the card not found, and more |

> **Important — for the full, steady speed of long training, give Apsi the right to set the card's power limit.**
> Without it, the only way Apsi can keep a hot card at 77 °C is to *pause* the work, and a card that needs cooling
> then loses a large share of its speed (measured on an RTX 3060: a 127 M-parameter model trained at 10,332 tokens/s
> with pauses, **18,074 tokens/s** with the power limit, at the same temperature). Do one of these:
>
> 1. **Run Apsi as administrator** (Windows: right-click the terminal → *Run as administrator*; Linux: `sudo`), or
> 2. **once**, run `apsi gpu setup` as administrator (or the *Apsi GPU Tuner — Apply* file from the download page):
>    a small helper then lets every Apsi program set the power limit without administrator rights, at every logon.
>
> The training report says which one is in use: `temperature held by the power limit, directly (no pauses)` or
> `… through the GPU helper`. If it reports cooling pauses, neither is set up.

---

## 1. Language models (`use hub`)

```apsi
let lm = hub.model("Qwen/Qwen2.5-3B-Instruct")        # device="auto": the GPU when there is one
print(lm.device, lm.plan)                             # where every layer runs, in which format, how many bytes
print(lm.chat("Explain how a rainbow forms."))
```

**Where the model goes.** With `device="auto"` the whole model goes to the card when it fits; when it does not, the
weights are stored on the card in a smaller format (below) so that every layer still fits; only when even 4-bit
weights do not fit are the remaining layers run on the processor (from RAM, or memory-mapped from the SSD).
`gpu_layers=N` chooses the split yourself, `device="cpu"` keeps it off the card, `lm.to("cuda")` moves it later.

**Measured on a desktop** (Ryzen 5 5500, RTX 3060 12 GB, Windows 11, display driver only):

| Model | Placement | Answer speed | Prompt of 512 tokens |
|---|---|---:|---:|
| Qwen2.5-0.5B | 24/24 layers, bfloat16, 0.9 GB | 200–222 tokens/s | 107 ms |
| Qwen2.5-3B | 36/36 layers, bfloat16, 5.9 GB | 45–48 tokens/s | 556 ms |
| Qwen2.5-14B (27.5 GB in bfloat16) | 48/48 layers, int8 + int4, 9.0 GB | 19.6–21.0 tokens/s | 2.6 s |

### Weight formats on the card

One scale per 32 values, quantized on the card from the bfloat16 checkpoint and read directly by the matrix-vector
kernels (no unpacking to bfloat16 first). `APSI_GPU_WEIGHTS` or `hub.model(…, gpu_weights=…)`:

| Format | Bytes / weight | 3B model on the RTX 3060 | Notes |
|---|---:|---|---|
| `bf16` | 2 | 44 tokens/s, 5.9 GB | the checkpoint as stored |
| `int8` | 1.125 | 62 tokens/s, 3.7 GB | the same 64 greedy tokens as bfloat16 |
| `fp8` (E4M3) | 1.125 | 58 tokens/s, 3.7 GB | the same 64 greedy tokens as bfloat16 |
| `fp6_e2m3` | 0.875 | 57 tokens/s, 3.0 GB | OCP MX fp6; about as close as fp8 at three quarters of its size |
| `fp6` (E3M2) | 0.875 | 55 tokens/s, 3.0 GB | wider range, coarser steps |
| `int4` | 0.625 | 67 tokens/s, 2.3 GB | ggml Q4_0-like; small quality loss |
| `fp4` (E2M1) | 0.625 | 57 tokens/s, 2.3 GB | OCP MX fp4 |
| `auto` (default) | — | — | bf16 → int8 → int4, the first in which every layer fits; int8 again for the first layers while room is left |

Every format answers sensibly ("The sky appears blue because of a phenomenon called Rayleigh scattering…" from all
seven); smaller formats drift from the bfloat16 wording sooner. The float formats round exactly as ml_dtypes does
(tested bit for bit).

### Tensor cores, cache, long prompts

* Prompts run on the tensor cores (bfloat16 inputs with float32 sums; `APSI_TC_PRECISE=1`, the default, splits each
  input into two bfloat16 parts for about float32 accuracy; quantized weights use one pass). A 512-token prompt on
  the 3B model: 556 ms, 754 ms without tensor cores. `APSI_TENSOR_CORES=0` turns them off.
* The key/value cache is float16 by default (twice the context in the same memory; `APSI_KV_CACHE=fp32`).
* Long prompts through layers that run on the processor stream those weights to the card (`APSI_PROMPT_ON_GPU`).
* Sampling (temperature, top-k, top-p, repetition penalty) runs on the card; only the few candidates come back.
* A conversation reuses the cache of its earlier turns: only the new part is computed.

### Many threads

A model can be used from any thread (a server's request threads, a worker pool); Apsi makes the card's context
current on the calling thread by itself, and threads take turns on one model.

---

## 2. Training on the GPU with Apsi alone (`backend="gpu"`)

```apsi
use layers
let model = transformer_lm(257, d=384, heads=6, layers=6, context=128)
let r = fit(model, model.init(seed=0), X, Y, loss="cross_entropy", opt=adamw(lr=1e-3), epochs=3, batch_size=32,
            backend="gpu", precision="auto")
print(r.device, r.history, r.gpu_memory_mb, r.gpu_temperature)
let logits = predict(model, r.params, X_test, backend="gpu")
```

`backend="auto"` picks it whenever there is an NVIDIA card and the model is made of the layers it has.

| | |
|---|---|
| Layers | linear, activation (relu, gelu, silu/swish, tanh, sigmoid, leaky_relu, identity), dropout, layer_norm, rms_norm, embedding, positional_encoding (learned or fixed), residual (with pre_norm), multi_head_attention (self-attention, causal or not), transformer_block — so `mlp`, `resnet_mlp`, `transformer_lm`, `transformer_encoder` |
| Losses | `cross_entropy` (negative labels left out), `mse` |
| Optimizers | Adam, AdamW, SGD (momentum); `clip=` (global norm, computed on the card), `schedule=` |
| Precision | `"auto"`/`"bf16"`: bfloat16 tensor cores, float32 sums; `"fp32"`: three tensor-core passes (about float32); `"fp8"`, `"fp6"`, `"fp6_e2m3"`, `"fp4"`, `"int8"`, `"int4"`: weights rounded to that format every forward pass, float32 master copy |
| Not yet | convolutions, LSTM/GRU, graph layers, mixture of experts, custom layers, losses other than the two — `backend="torch"` or `"numpy"` train those, and `fit` says so |

How it works: parameters, gradients and the AdamW state are three flat arrays updated by one kernel; the data set is
copied to the card once and every batch is gathered there, so inside an epoch the processor never makes the card
wait; the loss is read back once per epoch. Gradients equal Apsi's autodiff to 1e-5 and the loss curve equals the
NumPy backend's (tests/test_gpu_train.py), on the Emulator and on the real card.

| Extras | `ema=0.999` keeps a moving average of the weights (`r.ema_params`); `ewc=(anchor_params, fisher, strength)` pulls each weight toward an earlier task's value in proportion to its importance (continual learning without forgetting) |

**Measured** (the GPT above, 10.9 M parameters; the same data.bin for both: 4,441 windows of 128 bytes of Apsi's
documentation to train on, 234 held out; batch 32, AdamW 1e-3, 3 epochs; RTX 3060 12 GB; Apsi held the card near
77 °C by its power limit, PyTorch 2.14 ran without temperature control and reached 79 °C at 121–132 W):

| Run | Throughput | Every epoch | Held-out loss | GPU memory |
|---|---:|---|---:|---:|
| **Apsi `backend="gpu"`, bf16** | **151,744 tokens/s** | 153,552 · 154,009 · 151,667 | 6.25 → 2.371 | 873 MB |
| Apsi, fp8 weights | 144,176 tokens/s | 147,038 · 145,831 · 145,160 | 2.365 | 974 MB |
| Apsi, fp4 weights | 142,284 tokens/s | 145,197 · 143,729 · 141,405 | 2.456 | 974 MB |
| Apsi, fp32 (3 bfloat16 passes) | 60,000 tokens/s | 60,647 · 60,294 · 59,811 | 2.362 | 1,124 MB |
| PyTorch bf16 autocast (cuBLAS, flash attention) | 126,993 tokens/s | | 6.89 → 2.320 | 610 MB |
| PyTorch tf32 | 94,031 tokens/s | | 2.309 | 797 MB |
| PyTorch fp32 | 73,929 tokens/s | | 2.310 | 796 MB |

Without administrator rights, through the GPU helper (set up once by the installer or `apsi gpu setup`), the same
training ran at 146,051 tokens/s (148,783 · 148,898 · 143,629 by epoch) and the 127 M model at 17,968 tokens/s, with
no pauses: an ordinary run gets the same steady speed.

In bfloat16 Apsi trains this model **1.19× as fast as PyTorch**, every epoch at the same speed. One step takes
26.8 ms of kernels on the card; the matrix products are 75 % of it.

**A larger model** (10 blocks, d=1024, 16 heads: 126.6 M parameters; batch 32, 2 epochs, same card):

| Run | Throughput | Held-out loss | GPU memory |
|---|---:|---:|---:|
| **Apsi bf16, held at 77 °C** | **18,074 tokens/s** (18,418 · 18,284 by epoch) | 6.65 → 2.810 | 4,564 MB |
| Apsi, the kernels of one step alone (204 ms) | 20,043 tokens/s | | |
| PyTorch 2.13 bf16 autocast, no temperature control | 20,111 tokens/s | 6.58 → 3.020 | 3,854 MB |
| PyTorch tf32 / fp32 | 11,236 / 8,011 tokens/s | 3.025 / 3.027 | 4,565 MB |

Held at 77 °C, Apsi trains the 127 M model at 2.3× PyTorch's fp32 and 1.6× its tf32 speed, with a lower held-out
loss after the same two epochs; its kernels run as fast as PyTorch's bfloat16 does with the card left to heat up.
Without administrator rights (cooling by pauses) the same model ran at 10,332 tokens/s: see the box at the top.

Why the numbers differ: Apsi's bf16 kernels (tiled tensor-core products with the bias, activation and residual
fused in, fused attention, a warp per row for the norms, one AdamW kernel over all parameters, the data resident on
the card) are now faster than PyTorch's general kernels for this model. The RTX 3060 has no fp8 or fp4 hardware
(that starts with the RTX 40 / Ada generation), so fp8 and fp4 there are *simulated* (rounded weights): they cost a
little speed and serve to train models meant to run in those formats. True float32 is bounded by the card's float32
rate; PyTorch's cuBLAS float32 is still faster than Apsi's three-pass float32. PyTorch uses less memory (its
activations are bfloat16; Apsi keeps float32 activations). The products come in four builds and the card times them
for each shape the first time it sees it, keeping the fastest (`APSI_GEMM_VARIANT=a|x|w|h` fixes one).

**Several cards** — `fit(…, backend="gpu", device="cuda:all")` (or `device="cuda:0,1,2,3"`): data-parallel
training for any model the trainer supports. Every card holds the model and takes its share of each batch; then each
card sums only *its slice* of the gradient, pulled from all the others at the same time (reduce-scatter; NVLink /
PCIe peer copies where the driver allows, else through the processor), updates only that slice with its slice of the
optimizer's state, and the cards exchange the updated slices (all-gather). The gradient-norm clip adds the slices'
squared sums across the cards. The result equals one card with the whole batch (tested: Adam, AdamW, SGD, clip, EMA,
EWC); per card the memory is 8 + 8 / N bytes per parameter plus that card's activations.

**Any program at your temperature** — `apsi gpu hold 77` keeps the card near 77 °C by its power limit for whatever
runs on it (a PyTorch job, a renderer) until Ctrl+C, with the same control Apsi's training uses; the benchmark uses
it so PyTorch and Apsi are measured under the same temperature.

How big a model trains depends on the architecture: per card the weights and gradients (8 bytes per parameter) and
its share of AdamW's state (8 / N bytes), plus activations that depend on the layers, the context and the batch.
Nothing is refused in advance: when the parameters alone look larger than the card, a warning gives the numbers,
and if the card really runs out, the error says what to change (batch size, layers, context, more cards). The
architecture is yours to choose at any size: `transformer_lm(vocab, d=, heads=, layers=, context=)`, or your own
`Sequential` of the layers in the table above.

Large products in four builds: the card times each one (8 warps of 64 × 32; the same loading first; 4 warps of
64 × 64 over slices of 16; the operands rounded to bfloat16 once and copied into shared memory with cp.async four
slices ahead) for every shape the first time, the rounding pass included, and keeps the fastest; the tiles are
visited in groups of 8 rows so a large weight matrix is read from memory far less often.

Older and newer cards: on Turing (RTX 20) and GTX 16 the trainer runs on its float32 kernels (no bfloat16 tensor
cores there); on Hopper and Blackwell (H100, H200, B200, RTX 50) the Ampere kernels run as they are (the driver
compiles them for the card), without the newer cards' own instructions (wgmma, TMA, fp8 tensor cores) yet.

---

## 3. Serving: the OpenAI API

```
apsi serve-model Qwen/Qwen2.5-3B-Instruct --port 8000 --api-key KEY
```
```apsi
let ai = hub.serve(lm, port=8000, background=true)       # or without background: serves until Ctrl+C
```

`/v1/chat/completions` (streaming), `/v1/completions`, `/v1/models`, `/health`, `/metrics` (requests, tokens,
latency percentiles, the device, the GPU temperature). Details and options: [HUB.md](/docs/2.4.4/hub#serving-a-model-the-openai-api);
your own routes next to it: `examples/31_ai_server.apsi`.

**Measured under load** (RTX 3060, each model in turn): one client, 4 clients at once (12 requests), a burst of 16,
3 clients that hang up half way, the same question again, and 60 s of non-stop traffic from 3 clients.

| Model | One client | 4 clients | Burst of 16 | Hang-ups | Same answer | Soak |
|---|---|---|---|---|---|---|
| 0.5B | 165 tokens/s, first piece 0.06 s | 12/12 ok | 16/16 in 2.2 s | next ok in 0.07 s | identical | 167 answers, 0 errors |
| 3B | 42 tokens/s, first piece 0.11 s | 12/12 ok | 16/16 in 9.1 s | next ok in 0.38 s | identical | 42 answers, 0 errors |
| 14B | 17 tokens/s, first piece 0.48 s | 12/12 ok | 16/16 in 23.9 s | next ok in 1.1 s | identical | 25 answers, 0 errors |

GPU memory stayed flat during the soak (0.5B: 10,225 → 10,221 MB free) and the model never left the card.

---

## 4. Temperature control

Long work — training, long answers, a busy server — keeps the card near **77 °C (75–78 °C)** instead of letting it
run up to its own limit (often 83–90 °C on desktop cards). Apsi reads the temperature from the driver's management
library (NVML, installed with the driver) twice a second; above the band it paces the work (after each training
step or generated token it waits a fraction of the time that step took, so the card does less per second and
cools), below the band it gives the speed back. The control follows the temperature smoothed over about two seconds
(the die jumps several degrees within a fraction of a second whenever the load changes, and the sensor reads whole
degrees, so reacting to the raw reading made the pauses swing between none and most of the time); the pauses are
gathered into pieces of at least 20 ms, so the card is never left waiting between short steps. The temperature and
the speed stay level, and a card that stays cool is never slowed.

| Setting | |
|---|---|
| `APSI_GPU_TEMP=77` | the target in °C (default 77 → band 75–78); any value from 50 to 95 |
| `APSI_GPU_TEMP=off` | no control (the card's own limits only) |

**Steady speed: the power limit instead of pauses (administrator).** When Apsi runs as an administrator (Windows)
or root (Linux), it holds the temperature by moving the card's power limit (between 55 % and 100 % of the default)
rather than pausing: the card then runs at slightly lower clocks and voltage, where it does more work per watt, so the
speed stays level (typically 95 % or more of full speed) and the temperature stays at the target. Without those
rights Apsi cannot change the card's limits and paces the work instead: the temperature is just as steady, but the
speed falls as far as the cooler needs (on the RTX 3060 in a small case: from 141,000 to about 100,000 tokens/s once
the card is warm). `APSI_GPU_POWER_CONTROL=0` keeps pauses only. Better airflow or a higher fan curve gives the speed
back in both cases.

**Never as administrator again: `apsi gpu setup` (once).** It installs a small helper that Windows' Task Scheduler
(or systemd on Linux) starts at every logon with administrator rights; its only job is to set the power limit when
an Apsi program asks (localhost only, with a key only your account can read, never above the card's default). From
then on every Apsi program holds the temperature by the power limit without administrator rights. `apsi gpu setup 80%`
also applies 80 % at every logon; `apsi gpu remove` takes it all away and restores the default.

**The Apsi GPU Tuner (download page).** Two plain-text files per system, no installation: *Apply* sets every
NVIDIA card in the computer to the power limit measured as the best for sustained training and keeps it after
every restart (a Task Scheduler task on Windows, a systemd service on Linux); *Restore* removes all of it and puts
every card back to its factory default. Windows asks once for administrator rights (the file asks by itself).

| Card | Power limit | Effect |
|---|---|---|
| GeForce RTX 50 / RTX 40 | 75 % | about -25 % power for about -5 % training speed |
| GeForce RTX 30 | 80 % | about -5 % speed, 6–10 °C cooler |
| GeForce RTX 20, GTX 16, GTX 10 | 85 % | about -4 % speed, cooler |
| Workstation (RTX PRO, RTX A-series, Ada, Quadro) | 90 % | steadier clocks for about -2 % |
| Data center (H100, H200, B200, A100, L40S, L4, T4…) | unchanged | built and cooled for full sustained power |
| Laptop GPUs | unchanged | the laptop manages the limit |

With Apsi installed, *Apply* also runs `apsi gpu setup`, so Apsi fine-tunes the limit around 77 °C while it trains.

**Once per start of the computer: `apsi gpu power 80%`.** Run it once from a terminal opened as administrator (or
with sudo): the card keeps that limit for every program until the computer or the driver restarts, so Apsi (and
everything else) runs cooler at almost full speed without administrator rights afterwards. `apsi gpu power default`
puts it back; `apsi gpu` (or `apsi gpu watch`) shows the temperature, power, limits, clocks, fan and load.

**Cooler without pauses: a lower power limit.** A card at its full power limit runs at its highest clocks and
voltage, where each extra percent of speed costs several percent of power — that is the heat. `APSI_GPU_POWER=80`
sets the limit to 80 % of the default while Apsi runs (and puts it back at exit): typically about 5 % slower and
several degrees cooler, with no pauses at all. Windows and Linux allow it only to an administrator / root; without
that, Apsi says so and paces as above. The pacing itself is a PI controller (it follows the temperature smoothly and
settles at the target rather than switching between hot and paused).

**The processor too.** `APSI_CPU_TEMP=77` (default; `off` for none) paces training and answering on the processor
the same way. Linux reads the processor's own sensor (k10temp / zenpower on AMD, coretemp on Intel). Windows only
shows a temperature to programs run as administrator (the ACPI thermal zone, which on some boards is the
motherboard's sensor rather than the processor's); without it there is no processor control on Windows — the
processor's own boost limits still apply.

`r.gpu_temperature` after `fit`, `/metrics` of the server and `hub.gpu_status()` report the temperature, the highest
seen, and how much time was spent pacing. Paced work is slower in proportion — that is the trade for a cooler card.
Without NVML (very old drivers), there is no control and nothing else changes.

---

## 5. Images, video and your own kernels (`@gpu`)

A function written once compiles for the processor (`@native`) and the card (`@gpu`), one thread per element:
luminance + Sobel edges on RGB frames ran at 69 fps in 4K and 34 fps in 8K on the RTX 3060 (9 and 2 fps on all 12
processor threads), with exactly the processor's results. Frames kept on the card between calls (`to_gpu`) avoid
the copies. Several GPUs: `APSI_GPUS=0,2`. Small launches go to the processor by themselves (`APSI_GPU_AUTO=0`,
`APSI_GPU_MIN`). See [NATIVE.md](/docs/2.4.4/native) and [REALTIME.md](/docs/2.4.4/realtime).

---

## 6. Troubleshooting

**"no NVIDIA GPU" / `lm.device` is `cpu`.** `lm.device_note` says why. Install or update the NVIDIA display driver
(any driver from 2020 on); on Linux, `nvidia-smi` must work. `APSI_NO_GPU=1` is set? Remote desktops and some
virtual machines hide the card.

**The answers suddenly got slow.** The model left the card after a driver error and carries on on the processor
(`apsi hub: the graphics card stopped working …`; the server prints a WARNING and `/health` shows `cpu (the GPU failed:
…)`). Restart the program; if it repeats, update the driver and check the card's temperature and power. `APSI_QUIET=1`
hides these messages — do not set it while investigating.

**Out of GPU memory.** Apsi retries in smaller pieces by itself; for a model, use a smaller format
(`APSI_GPU_WEIGHTS=int4`), a shorter context, or `gpu_layers=`; for training, a smaller `batch_size`. Other programs
on the card (games, browsers with hardware acceleration, another Apsi) use memory too: `nvidia-smi` lists them.

**Training is slower than it should be, or its speed swings.** Look at the training report. *Cooling pauses* mean
Apsi cannot set the power limit: run it as administrator, or set up the helper once (`apsi gpu setup` as
administrator, or the GPU Tuner's Apply file) — see the box at the top of this page. *fused attention unavailable*
next to the GPU line means the attention kernel could not be built (a bug: report it with that line); training
still works, more slowly, on separate kernels. A card held at a lower power limit (the GPU Tuner's 80 %) trades about
5 % of speed for 6–10 °C; `Apsi-GPU-Tuner-Restore` puts the default back.

**The card is hot or loud.** Temperature control is on by default (77 °C). Lower it with `APSI_GPU_TEMP=72`; check the
case airflow and fan curve. If the card still reaches 85 °C+, the control may be off (`APSI_GPU_TEMP=off`) or NVML
missing (`r.gpu_temperature` is `none`).

**Windows: the screen flickers and the program stops with "launch timed out".** A single kernel ran longer than
Windows allows (TDR, 2 s) — very large prompts or batches on a slow card. Use a smaller `batch_size` or prompt pieces;
as a last resort raise `TdrDelay` in the registry (Microsoft's documentation).

**"the GPU driver rejected the kernel".** The driver is too old for the PTX version; update it. Delete
`~/.apsi/cache/gpu` after a driver update if the message stays (`apsi cache clean gpu`).

**Tensor cores are not used (`lm.plan` says `tensor_cores: false`).** Turing (7.5) has no bfloat16 tensor cores; on
Ampere and newer `lm.plan` / `r.device` give the reason (`APSI_TENSOR_CORES=0` set, a failed self-check).

**`backend="gpu"`: "layer … is not on the GPU trainer yet".** The model uses a layer the trainer does not have yet;
`backend="torch"` (with PyTorch) or `backend="numpy"` trains it. `backend="auto"` chooses by itself.

**Training loss is NaN.** Lower the learning rate, add `clip=1.0`; with `"fp16"`-style instability use `"fp32"`
(three passes) to check whether precision is the cause.

**The server answers 429.** More than `--max-queue` requests are waiting: the model computes one request at a time.
Raise `--max-queue`, use a smaller model or format, or run another server on another card.

**Two copies of a model (int8 for the processor) fill the disk.** `apsi cache show`, `apsi cache clean int8`.

## Settings for the GPU

| Variable | Default | Effect |
|---|---|---|
| `APSI_NO_GPU=1` | | everything on the processor |
| `APSI_GPU_WEIGHTS` | `auto` | `bf16`, `int8`, `fp8`, `fp6`, `fp6_e2m3`, `int4`, `fp4` |
| `APSI_GPU_TEMP` | `77` | temperature target in °C; `off` for none |
| `APSI_CPU_TEMP` | `77` | the processor's target in °C; `off` for none |
| `APSI_GPU_POWER` | | `80` (% of the default) or `130W`: a lower power limit while Apsi runs — cooler at almost the same speed; needs an administrator / root |
| `APSI_TENSOR_CORES` | `1` | `0`: float32 kernels only |
| `APSI_TC_PRECISE` | `1` | `0`: one tensor-core pass for bfloat16 weights too (faster, about 1e-2 instead of 1e-5) |
| `APSI_KV_CACHE` | `fp16` | `fp32` for the old cache |
| `APSI_PROMPT_ON_GPU` | `64` | prompts of this many tokens or more through processor layers stream to the card |
| `APSI_CPU_WEIGHTS` | `int8` | processor layers in `bf16` instead |
| `APSI_PRECISION` | `auto` | training precision for every program |
| `APSI_FP8_MATMUL` | `1` | `0`: no fp8 tensor-core products with `backend="torch"` |
| `APSI_API_KEY` | | the key `apsi serve-model` requires |
| `APSI_GPUS` | all | which cards `@gpu` code uses, e.g. `0,2` |
| `APSI_QUIET` | | `1`: no messages about fall-backs and placement |
| `APSI_PROGRESS` | `1` | `0`: no progress lines while preparing big models |
| `APSI_NO_CACHE` | | `1`: rebuild the GPU kernels instead of using `~/.apsi/cache/gpu` |