Настоящий материал представляет собой обзор главы 19 книги Chris Fregly, «AI Systems Performance Engineering: Optimizing Model Training and Inference Workloads with GPUs, CUDA, and PyTorch», O'Reilly Media, 2026 (Copyright © 2026 Flux Capacitor, LLC).
Прямые цитаты из книги выделены отдельными блоками с пометкой «Цитата из книги». Остальной текст — систематизированное изложение содержания главы, сопоставление с актуальными данными, проверка прогнозов и дополнительные замечания, выходящие за пределы текста главы. Цитаты приводятся в ограниченном объёме в целях обзора и критического анализа.
Инференс сверхбольших языковых моделей (LLM) на современном оборудовании требует динамической адаптации во время выполнения, чтобы одновременно достигать высокой пропускной способности и низкой задержки в изменяющихся условиях. Статический подход «один размер для всех» больше не работает — он приводит либо к недоиспользованию дорогостоящих GPU при лёгкой нагрузке, либо к сбоям OOM и нарушению SLA при пиковом трафике. Современные системы обслуживания моделей используют адаптивные стратегии, которые на лету корректируют параллелизм, числовую точность, планирование CUDA-ядер и использование памяти. В этой главе рассматриваются такие передовые техники, как динамическое переключение параллелизма, масштабирование точности, управление кэшем в реальном времени и настройка на основе обучения с подкреплением (RL).
Огромные LLM требуют параллелизма моделей — тензорного и конвейерного или их гибрида — для распределения вычислений по нескольким GPU. У каждого подхода есть преимущества и недостатки. В таблице 19-1 приведены рекомендуемые стратегии параллелизма для конкретных паттернов трафика инференса.
| Паттерн трафика инференса | Рекомендуемый параллелизм | Обоснование |
|---|---|---|
| Много коротких запросов (< 256 токенов, высокий RPS) | Data parallel / масштабирование реплик | Минимизирует межGPU-коммуникации; каждый GPU запускает реплики, обрабатывающие независимые запросы (если модель помещается в память одного GPU) |
| Несколько длинных запросов (≥ 8k токенов, низкая конкурентность) | Pipeline parallelism (с микробатчами) | Снижает задержку на запрос за счёт разделения слоёв между GPU |
| Смешанная нагрузка (короткие + некоторые длинные) | Гибридный динамический (автопереключение) | Короткие чаты — на одном GPU, длинные — через конвейер для соблюдения SLA по задержке |
| Чрезвычайно большая модель (> памяти одного GPU) | Гибрид Tensor + Pipeline | Необходим для размещения модели; балансирует вычисления и память по обоим измерениям |
| Инференс MoE-моделей | Expert parallelism | Распределяет экспертов по GPU; запрос активирует лишь подмножество экспертов, снижая нагрузку на память и вычисления |
Data parallel предполагает репликацию полной модели на каждом GPU с балансировкой входящих запросов между репликами. Это не требует межGPU-синхронизации для отдельных инференсов, поскольку каждый GPU обрабатывает отдельные запросы независимо. Подход максимизирует пропускную способность для множества небольших или средних входов с минимальными накладными расходами. Однако data parallel невозможен, если модель не помещается в память одного GPU.
Tensor parallelism (TP) — форма параллелизма моделей, разделяющая матрицы модели между GPU для ускорения матричных умножений. Вносит дополнительные all-reduce коммуникации для синхронизации GPU после каждого слоя.
Pipeline parallelism (PP) — также параллелизм моделей, но назначает целые слои разным GPU для преодоления ограничений памяти. Вносит накладные расходы в виде задержек последовательных стадий — так называемые «пузыри» конвейера (pipeline bubbles).
Простаивающие стадии (bubbles) между полезной работой на разных GPU
Expert parallelism, используемый в архитектурах mixture-of-experts (MoE), назначает каждую экспертную подсеть на отдельный GPU. Лёгкая сеть маршрутизации направляет каждый входной запрос или токен только к top-k активным экспертам. Например, DeepSeek-V3 — MoE-модель с 671 млрд параметров, активирующая около 37 млрд параметров на токен (1 общий эксперт + 8 выбранных маршрутизатором из 256)[reference:0]. Активируя лишь нескольких экспертов на вход, expert parallelism снижает память на устройство, время инференса и вычислительные затраты.
Традиционно стратегия параллелизации выбирается и фиксируется заранее при загрузке модели. Однако для максимизации производительности при динамических нагрузках современные инференс-движки могут выбирать различные стратегии параллелизма во время выполнения на основе характеристик входных данных.
Ключевые факторы: размер батча, длина последовательности, использование памяти, требования к задержке и пропускной способности. Очень длинные промпты могут направляться в экземпляр TP + PP, поскольку это распределяет слои по GPU и предотвращает ошибки OOM. Короткие запросы, чувствительные к задержке, направляются в экземпляр модели только с TP, чтобы избежать накладных расходов конвейерных стадий.
Router выбирает пул в зависимости от длины запроса и SLA
DeepSeek-V3 — разреженная mixture-of-experts модель с 671 млрд параметров, активирующая 37 млрд параметров на токен[reference:1]. Рассмотрим сервер с 8× GPU Blackwell B200 с общим объёмом 1440 ГБ HBM (180 ГБ на GPU × 8).
def choose_worker_pool(seq_len, gpu_mem_util, concurrent_reqs):
if seq_len > 4096 or gpu_mem_util > 0.8:
return "tp_pp_hybrid"
if concurrent_reqs > 4:
return "tensor_parallel"
return "tensor_parallel"
Ключ — динамически балансировать тензорные и конвейерные разбиения, чтобы каждый GPU был хорошо загружен, оставаясь в рамках ограничений памяти и достигая целевых показателей задержки. Это то, чего статическая конфигурация «один размер для всех» не может обеспечить.
Выбор стратегии параллелизма определяется паттерном трафика: data parallel для множества коротких запросов, TP + PP для длинных промптов, expert parallelism для MoE-моделей.
Современные движки поддерживают несколько предварительно сконфигурированных пулов GPU с разными стратегиями шардирования. Router направляет запрос в пул, соответствующий длине запроса и SLA.
Современные GPU, такие как Blackwell, поддерживают 8-битную и 4-битную плавающую точку (FP8/FP4) в вычислительных блоках Tensor Core. Эти более низкие точности обеспечивают значительное ускорение, экономию памяти и минимальную потерю качества.
Динамическое переключение точности — передовая техника, при которой инференс-движок корректирует числовую точность во время выполнения на основе уверенности модели или давления на ресурсы. Цель — увеличить пропускную способность без существенной потери качества.
Если распределение вероятностей модели для следующего токена показывает экстремальные пики из-за высокой уверенности, небольшие числовые ошибки от низкой точности вряд ли изменят результат. В этом случае движок безопасно использует FP4 для следующих нескольких шагов. И наоборот, если распределение более плоское из-за высокой неопределённости, движок должен оставаться на FP8 или FP16.
Инференс-движок количественно оценивает неопределённость, вычисляя энтропию Шеннона распределения softmax по словарю. Низкая энтропия указывает на более резкое (более уверенное) предсказание. Фиксированный порог энтропии, настроенный на отложенном валидационном наборе, определяет, когда переключаться на FP4, а когда оставаться на FP8/FP16.
Если использование памяти GPU приближается к пределу из-за очень длинного контекста или множества параллельных запросов, система может динамически сжимать активации до более низкой точности. Например, хранить тензоры ключей/значений внимания в INT4 вместо INT8, когда память ограничена — это уменьшит объём памяти на 50%. Однако необходимо убедиться, что ошибка квантования не накапливается через множество шагов декодирования.
| Режим точности | Память (относительно) | Пропускная способность | Влияние на качество |
|---|---|---|---|
| FP16 (базовый) | 1,0× (100%) | 1,0× (базовый) | Без влияния (полная точность) |
| FP16 веса + FP8 активации | ~0,5× (50%) | ~1,5× | Незначительное (< 0,1%) |
| INT4 веса + FP8 активации | ~0,25× (25%) | ~1,8× (смешанная нагрузка) | ~0,5% падения качества |
| INT4 веса + FP4 активации | ~0,2× (20%) | ~3,5× | ~1% падения (требуется настройка) |
Современные GPU предоставляют нативные Tensor Cores для FP8 и FP4. Однако поддержка AMP в PyTorch на момент написания нацелена только на FP16 и BF16. Для инференса рекомендуется использовать NVIDIA Transformer Engine (TE) и применять его MXFP8 и NVFP4, когда это уместно.
Библиотеки, такие как NVIDIA Transformer Engine, позволяют послойно управлять точностью во время выполнения. Линейные слои могут по умолчанию использовать FP8, но хук времени выполнения может повысить точность слоя до FP16 или понизить до FP4 в зависимости от роли:
import contextlib
import torch
try:
from transformer_engine.pytorch import fp8_autocast as _te_fp8_autocast
_TE_AVAILABLE = True
except Exception:
_TE_AVAILABLE = False
class _NullCtx(contextlib.ContextDecorator):
def __init__(self, **_): pass
def __enter__(self): return self
def __exit__(self, *exc): return False
def _te_fp8_autocast(**_):
return _NullCtx()
def _precision_context_cuda(use_fp8, prefer_bfloat16, enable_fp8):
if use_fp8 and enable_fp8 and _TE_AVAILABLE:
return _te_fp8_autocast(enabled=True)
amp_dtype = torch.bfloat16 if prefer_bfloat16 else torch.float16
return torch.autocast(device_type="cuda", dtype=amp_dtype)
Этот паттерн создаёт эластичный режим точности: когда модель работает в предсказуемых (низкоэнтропийных) областях — генерация пунктуации, шаблонных завершений — она продолжает в FP8 для максимальной производительности. Когда она входит в области с более высокой энтропией — неоднозначные промпты, точки принятия решений — она возвращается к FP16 для сохранения числовой точности.
Динамическое переключение точности на основе энтропии Шеннона и давления на память позволяет использовать FP4/FP8 в предсказуемых областях и возвращаться к FP16 при высокой неопределённости.
NVIDIA Transformer Engine предоставляет послойное управление точностью. FP4 — для лёгких слоёв, FP8/FP16 — для ранних слоёв, обрабатывающих сырые входы.
Производительность слоёв нейронной сети на GPU может кардинально различаться в зависимости от низкоуровневых параметров: размер блока потоков, размеры тайлов, разворачивание циклов, паттерны доступа к памяти. Для моделей фиксированного размера библиотеки обычно выбирают эти параметры один раз.
Однако в сценарии онлайн-сервиса входные размеры — длины последовательностей и размеры батчей — могут меняться от запроса к запросу. Автотюнинг ядер — это механизм времени выполнения, который выбирает или даже JIT-компилирует оптимальный вариант ядра для текущей нагрузки.
if (seq_len < 256) {
attn_kernel = standard_attention_kernel;
} else {
attn_kernel = tiled_attention_kernel;
}
output = attn_kernel(Q, K, V, mask);
За кулисами attn_kernel выбирает между совершенно разными реализациями CUDA. Одна оптимизирована для малых входов, другая — для больших контекстов с использованием тайлового ядра.
| Размер тайла | Разделяемая память (КБ) | Заполненность (%) | Пропускная способность (GOPS) |
|---|---|---|---|
| 64 × 64 | 48 | 85 | 8,2 |
| 128 × 64 | 64 | 78 | 10,5 |
| 128 × 128 | 96 | 72 | 9,8 |
| 256 × 128 | 128 | 60 | 11,3 |
Автотюнинг ядер позволяет выбирать оптимальную реализацию для каждой длины последовательности и размера батча, вместо использования единой конфигурации.
Оптимальный размер тайла зависит от длины последовательности: малые тайлы (64×64) дают лучшую заполненность, большие (256×128) — максимальную пропускную способность.
Тесно связана с настройкой ядер тема управления разделяемой памятью GPU и общей заполненностью streaming multiprocessor (SM). Динамически выделяя потоки во время выполнения на основе размера задачи и текущего использования разделяемой памяти, можно значительно улучшить производительность.
Рассмотрим T как ширину тайла в токенах на блок потоков. Каждый блок резервирует порядка O(T²) чисел с плавающей точкой в разделяемой памяти, поскольку self-attention квадратичен.
int T = choose_tile(L, gpu_shared_mem_per_block, num_sms);
size_t shared_mem_bytes = 3 * T * T * sizeof(float);
MyAttentionKernel<<<numBlocks, threadsPerBlock, shared_mem_bytes>>>(...);
extern __shared__ float smem[];
float* tile_q = smem;
float* tile_k = smem + T*T;
float* tile_v = smem + 2*T*T;
int maxBlocks, bestThreads;
for (int threads = 64; threads <= 256; threads *= 2) {
cudaOccupancyMaxActiveBlocksPerMultiprocessor(
&maxBlocks, MyKernel, threads,
sharedMemPerBlock(threads));
float occupancy = (float) maxBlocks * threads / hardwareMaxThreadsPerSM;
}
Динамическое выделение разделяемой памяти позволяет балансировать между data reuse (большие тайлы) и occupancy (малые тайлы) в зависимости от текущей длины последовательности.
CUDA Occupancy API и параметр carveout дают инструменты для тонкой настройки. Простаивающие варпы и циклы простоя памяти — ключевые метрики для выбора оптимального размера тайла.
При обслуживании LLM в режиме реального времени время до первого токена (TTFT) — критическая метрика. Одним из основных факторов, влияющих на TTFT, является время настройки внутренних состояний модели — кэша ключей-значений (KV) — перед началом генерации.
SpeCache использует спекулятивное предсказание того, какие KV-пары понадобятся на следующем шаге декодирования[reference:2]. Подход основан на следующей архитектуре: полный 16-битный KV-кэш хранится в CPU-памяти, в GPU-памяти находится low-bit копия для оценки важности токенов. На каждом шаге декодирования модель предсказывает, какие KV-пары будут наиболее важны для следующего токена, и предзагружает соответствующие полноценные 16-битные пары из CPU-памяти в VRAM[reference:3].
Таким образом, SpeCache не выполняет классическое спекулятивное декодирование с верификацией и отклонением токенов. «Спекулятивность» здесь относится к предсказанию паттерна доступа к KV-кэшу, а не к генерации и проверке токенов. Это принципиально иная механика, чем speculative decoding из главы 15.
Low-bit копия в VRAM оценивает важность; полные 16-битные пары предзагружаются из CPU-памяти
out = model.generate(..., cache_implementation="offloaded")
out = model.generate(..., cache_implementation="offloaded_static")
Под капотом OffloadedCache хранит только KV текущего слоя в памяти GPU, выгружая остальные на CPU. Он асинхронно предзагружает кэш следующего слоя, пока текущий вычисляется, и одновременно записывает кэш предыдущего слоя обратно на CPU.
void* prefetch_dest = nullptr;
cudaMalloc(&prefetch_dest, KV_BYTES);
void* kv_src_host = nullptr;
cudaMallocHost(&kv_src_host, KV_BYTES);
cudaStream_t compute_stream, prefetch_stream;
cudaStreamCreateWithFlags(&compute_stream, cudaStreamNonBlocking);
cudaStreamCreateWithFlags(&prefetch_stream, cudaStreamNonBlocking);
cudaEvent_t kv_ready;
cudaEventCreateWithFlags(&kv_ready, cudaEventDisableTiming);
forward_kernel<<<1, 1, 0, compute_stream>>>();
cudaMemcpyAsync(prefetch_dest, kv_src_host, KV_BYTES,
cudaMemcpyHostToDevice, prefetch_stream);
cudaEventRecord(kv_ready, prefetch_stream);
cudaStreamWaitEvent(compute_stream, kv_ready, 0);
consume_prefetched_kv<<<1, 1, 0, compute_stream>>>();
Хостовые буферы должны быть pinned (page-locked). В противном случае cudaMemcpyAsync может сериализоваться.
SpeCache держит полный 16-битный KV-кэш в CPU-памяти, а в VRAM — low-bit копию для оценки важности токенов[reference:5]. На каждом шаге декодирования предзагружаются полноценные 16-битные пары, отобранные по спекулятивным attention-оценкам.
Предзагрузка KV-пар через отдельный CUDA-поток позволяет перекрыть задержку CPU-GPU передачи с вычислениями. Хостовые буферы должны быть pinned для истинной асинхронности.
По мере генерации токенов KV-кэш растёт линейно. Для длинных разговоров, документов и цепочек рассуждений он потребляет огромный объём памяти GPU.
Простой подход — снижение точности. Многие фреймворки по умолчанию используют FP16 или BF16 для KV-кэша. Однако часто можно сжать ключи и значения до 8-бит или даже 4-бит с минимальным влиянием на качество. Hugging Face Transformers поддерживает QuantizedCache (INT8 и INT4): cache_implementation="quantized".
Пример политики: хранение последних 128 токенов в полной точности, сжатие остальных в 4-бит. Наиболее свежий контекст сохраняется с более высокой точностью, а более старая история — с более низкой.
Три политики хранения KV и их переключение на лету
Дополнительно можно использовать Sliding-window attention — при котором каждый токен видит только последние W токенов. Это даёт линейное масштабирование O(N·W) вместо O(N²).
Недавние токены важнее — окно скользит по последовательности
HQQ — калибровочно-независимый квантователь на лету, поддерживающий 2-бит, 3-бит, 4-бит и 8-бит. Использует робастную оптимизацию для моделирования выбросов и ошибок с тяжёлыми хвостами. Хорошо интегрируется в Hugging Face Transformers.
def make_cache_config(*, backend, nbits, device, compute_dtype=torch.float16,
q_group_size=64, residual_length=128):
return {
"backend": backend,
"nbits": int(nbits),
"axis_key": 1,
"axis_value": 1,
"q_group_size": int(q_group_size),
"residual_length": int(residual_length),
"compute_dtype": compute_dtype,
}
Три политики KV-кэша: dense (все токены), sparse (разреженная) и sliding-window (скользящее окно). Переключение между ними может происходить на лету в зависимости от длины контекста и давления на память.
Sliding-window attention даёт линейное масштабирование O(N·W) вместо O(N²). HQQ — калибровочно-независимый квантователь, поддерживающий 2–8 бит, хорошо интегрируется в Hugging Face Transformers.
Многие техники, обсуждаемые в этой главе, включают решения, основанные на текущем состоянии системы. Эти решения требуют компромиссов — скорость против точности, пропускная способность против задержки. Вместо сбора всё большего количества эвристик, можно настраивать систему инференса с помощью обучения с подкреплением (RL).
Инференс-движок наблюдает за метриками сервера (среда), выбирает действия (политика) для максимизации пропускной способности при удержании задержки в пределах цели и получает обратную связь (награда).
Пространство действий:
reward = tokens_per_second - 1000 * (1 if latency > SLA else 0)
reward = tokens_per_second - λ * max(0.0, latency - SLA)
λ — настраиваемый вес штрафа. Больший λ заставляет агента приоритизировать соблюдение SLA. Меньший — позволяет рисковать превышениями для более высокой пропускной способности.
RL-агенты позволяют автоматически настраивать параметры инференса (параллелизм, точность, размер батча, сжатие кэша) на основе текущих метрик системы.
PPO рекомендуется для непрерывного обновления. Демпфирование и ограничение пространства действий предотвращают осцилляции. Аварийный выход отключает агента при деградации p95.
Фрагментация памяти GPU и неоптимальное выделение могут быть тихими убийцами производительности. Серверы инференса выделяют и освобождают тысячи тензоров в секунду. Стратегия аллокатора влияет на фрагментацию и задержку выделения.
import os
import sys
import gc
import pickle
import tempfile
import subprocess
def run_with_fallback_allocator(factory, request_fn):
try:
model = factory()
return request_fn(model)
except torch.cuda.OutOfMemoryError:
gc.collect()
torch.cuda.empty_cache()
with tempfile.NamedTemporaryFile(suffix=".pkl", delete=False) as f:
pickle.dump(factory, f)
factory_path = f.name
with tempfile.NamedTemporaryFile(suffix=".pkl", delete=False) as f:
pickle.dump(request_fn, f)
request_path = f.name
output_path = tempfile.mktemp(suffix=".pkl")
env = os.environ.copy()
env["PYTORCH_ALLOC_CONF"] = "backend:cudaMallocAsync"
subprocess.run([
sys.executable, __file__, "--child",
"--factory", factory_path,
"--request", request_path,
"--output", output_path,
], env=env, check=True)
with open(output_path, "rb") as f:
return pickle.load(f)
Используйте torch.cuda.memory_reserved() и torch.cuda.memory_allocated() для отслеживания фрагментации. Reserved — общая память, удерживаемая аллокатором. Allocated — сколько фактически используется тензорами. Большая разница указывает на фрагментацию.
Динамическое переключение аллокаторов памяти позволяет бороться с фрагментацией. BFC — стандарт PyTorch, cudaMallocAsync — потоковый аллокатор CUDA с поддержкой переиспользования памяти.
При OOM система может перезапустить процесс с PYTORCH_ALLOC_CONF=backend:cudaMallocAsync. Мониторинг reserved vs allocated выявляет фрагментацию.
В быстро развивающемся мире GPU-оборудования и алгоритмических инноваций постоянно появляются новые и более быстрые реализации ядер. Патчинг ядер во время выполнения — это возможность интегрировать новые реализации в работающую систему без полного переразвёртывания.
import types
def patch_attention_with_flash_attn3(module, new_flash_attn_lib):
def new_forward(self, query, key, value, **kwargs):
return new_flash_attn_lib.forward(
query, key, value, **kwargs
)
module.forward = types.MethodType(new_forward, module)
Runtime kernel patching позволяет заменять реализации ядер без перезапуска сервиса. Это критически важно для 24/7-сервисов, где перезагрузка модели недопустима.
Хорошо определённые границы модулей (attention, MLP) упрощают горячую замену. Комбинируйте с RL-агентом для выбора оптимальной реализации под каждый тип запроса.
В высокопроизводительном инференсе холодный старт — значительный фактор задержки. CUDA Graphs позволяют захватить последовательность операций GPU и воспроизвести её с минимальными накладными расходами на запуск.
Используя алгоритмы прогнозирования временных рядов (ARIMA, Prophet), система предугадывает будущие паттерны нагрузки — всплески трафика, изменения размера батча — чтобы проактивно подготовить граф для быстрого выполнения.
import torch
static_input = torch.randn(batch_size, seq_len, hidden_dim, device="cuda")
graph = torch.cuda.CUDAGraph()
with torch.cuda.graph(graph):
static_output = model(static_input)
def run_with_graph(input_tensor):
static_input.copy_(input_tensor)
graph.replay()
return static_output.clone()
При запуске новых экземпляров GPU автомасштабировщик должен запускать набор вызовов API прогрева:
Прогнозирование нагрузки на основе временных рядов (ARIMA, Prophet) позволяет проактивно прогревать CUDA-графы и кэши, устраняя задержку холодного старта.
При автомасштабировании новые экземпляры GPU должны проходить warmup-процедуру: загрузка модели, кэширование JIT-выходов, выделение пулов памяти, подготовка графов для каждого размера батча.
Адаптивная пакетная обработка динамически корректирует способ группировки запросов в батчи в зависимости от нагрузки. Во время пиковой нагрузки система использует большой размер батча (8 или 16), потому что пропускная способность критична. В периоды низкой нагрузки система уменьшает размер батча до 1 для минимизации задержки.
Статья SARATHI продемонстрировала, что чанковый prefill и подсадка decode-запросов могут уменьшить пузыри конвейера и улучшить пропускную способность на ~1,3–1,9× по сравнению с наивным планированием[reference:6]. SARATHI разбивает prefill-запрос на равные чанки и выполняет decode-maximal batching, при котором decode-запросы «подсаживаются» в промежутки между чанками prefill.
В vLLM эта техника реализована через флаг enable_chunked_prefill=True. При включении vLLM меняет политику планирования: сначала батчатся все ожидающие decode-запросы, затем — prefill-чанки[reference:7]. Флаг --long-prefill-token-threshold задаёт порог, при котором prefill считается «длинным» и подлежит чанкованию[reference:8].
Пример: Запрос prefill на 10 000 токенов без чанкования заблокирует весь конвейер. При разбиении на пять чанков по 2 000 токенов можно чередовать батчи decode между чанками prefill.
Выбирайте размер чанка так, чтобы чанк prefill занимал ~50–100 мс. Это даёт частые возможности планировать батчи decode между ними.
import cupy as cp
TARGET_UTIL = 0.85
OCC_THRESHOLD = 0.50
BLOCK_THREADS = 256
def get_optimal_tile(seq_len):
if seq_len < 256:
return 64
elif seq_len < 1024:
return 128
else:
return 256
def scheduler_loop():
stream = cp.cuda.Stream(non_blocking=True)
while True:
pending = get_pending_requests()
util = gpu_utilization()
if util < TARGET_UTIL and any(r.phase == 'prefill' for r in pending):
req = select_heaviest_prefill(pending)
L = req.remaining_length()
T = get_optimal_tile(L)
shared_bytes = 3 * T * T * 4
Разделяйте фазы prefill и decode по ресурсам, используя маршрутизацию запросов с учётом контекста — попадания в KV-кэш, глубина очереди, нагрузка. Это поддерживает высокую пропускную способность с длинными промптами без замедления ответов на короткие.
Чанковый prefill (SARATHI) разбивает длинные prefill-запросы на чанки и «подсаживает» decode-запросы в промежутки, уменьшая пузыри конвейера и улучшая пропускную способность на ~1,3–1,9×[reference:9].
В vLLM флаг enable_chunked_prefill=True активирует эту технику. Планировщик сначала батчит decode-запросы, затем — prefill-чанки[reference:10].
Современные системы, такие как Grace Blackwell GB200 NVL72 (72 GPU B200, 180 ГБ HBM каждый) и GB300 NVL72 (72 GPU B300, 288 ГБ каждый), соединяют 72 GPU в единую фабрику NVLink/NVSwitch, обеспечивая каждому GPU до ~1,8 ТБ/с совокупной двунаправленной пропускной способности и более 130 ТБ/с совокупной поперечной пропускной способности.
Понимание топологии — какие GPU совместно используют компоненты NVSwitch или пути NVLink — критически важно. GPU на одной плате или лотке имеют более быстрые и прямые пути по сравнению с GPU в разных лотках или узлах.
Это проблема оптимизации размещения процессов на GPU — отображение графа слоёв нейронной сети на GPU-оборудование с минимальными затратами на коммуникацию. Переназначение может быть сформулировано как задача разбиения графа.
NCCL — стандарт для управления GPU-коллективами. Предлагает несколько алгоритмов:
Вместо одновременного обмена данными всех 72 GPU (72 × 71 сообщение) можно разделить обмен на 4 волны по 18 GPU. Это структурирует обмен как серию меньших all-to-all, использующих только подмножество GPU во время каждой волны.
В MoE-моделях эксперты могут быть перераспределены между GPU на основе наблюдаемой нагрузки. Если один GPU перегружен экспертами, которые часто активируются, планировщик может перенести экспертов на менее загруженные GPU.
Планировщик может задержать высокоприоритетную передачу на несколько миллисекунд, чтобы предыдущая завершилась и очередь очистилась. Это временная балансировка нагрузки, аналогичная управлению очередями сетевого коммутатора.
NVSwitch имеет собственные внутренние алгоритмы планирования, включая адаптивную маршрутизацию по нескольким чипам и плоскостям коммутатора. Программное обеспечение может избежать достижения этих пределов, применяя адаптивные техники: ограничение числа тяжёлых передач, выполняемых параллельно, тонкая настройка параллелизма путём контроля конкуренции.
Рассматривайте сеть как планируемый ресурс — такой же, как GPU, память и т. д. Она должна планироваться и адаптивно управляться.
Планирование с учётом топологии и перегрузок включает: топологически-осознанное размещение процессов, выбор алгоритма NCCL (ring/tree/hierarchical), волновое планирование коллективов, multirail NIC и адаптивную маршрутизацию InfiniBand.
Ребалансировка экспертов MoE и динамическая балансировка во времени (задержка высокоприоритетных передач) позволяют избежать перегрузок NVSwitch. Сеть — планируемый ресурс, а не «чёрный ящик».
Далее приведены дополнительные техники, которые дополняют основные стратегии. На момент написания они экспериментальны, но перспективны.
Модели с ранним выходом позволяют LLM самостоятельно усекать генерацию при достижении достаточной уверенности. Методы отслеживают промежуточные представления и энтропию логитов для решения на каждом слое или токене, следует ли остановить вычисления. Требуют специальной архитектуры или обучения, но могут дать до 30–50% ускорения на задачах рассуждения без потери точности.
Фреймворк DASH представляет инференс как марковский процесс принятия решений, который динамически решает для каждого токена, выполнять или пропускать каждый слой на основе характеристик входа[reference:11]. DASH моделирует процесс пропуска как MDP, позволяя принимать решения на уровне отдельных токенов на основе промежуточных представлений[reference:12]. Обучая небольшую сеть оценки, DASH может пропускать 20–40% слоёв для многих токенов.
В MoE-моделях можно использовать предсказание маршрутизации для предзагрузки только тех экспертов, которые, вероятно, будут активированы, снижая коммуникацию и задержку.
Система отслеживает статистику активаций в реальном времени и пересчитывает диапазоны каждые N токенов. FP16 выделяется для «горячих» слоёв с высокой дисперсией, FP8 — для «холодных» с низкой дисперсией.
LazyLLM избирательно вычисляет KV-пары только для токенов, важных для предсказания следующего токена, как на этапе prefill, так и на этапе decode[reference:13]. Этот метод позволяет языковым моделям динамически выбирать различные подмножества токенов из контекста на разных шагах генерации, даже если некоторые из них были удалены на предыдущих шагах[reference:14].
Используйте спекулятивное декодирование с динамическим выбором draft-модели, адаптивные стратегии beam search и динамическое управление температурой сэмплинга на основе уверенности модели.
Используйте инструменты — NVML, Nsight Systems/Compute, инструментацию NVTX, метрики Prometheus — для выявления узких мест во время выполнения и автоматического применения оптимизаций графов и ядер.
Дополнительные техники включают: DASH (пропуск слоёв через MDP), LazyLLM (динамическое прореживание токенов), спекулятивную маршрутизацию экспертов MoE, динамическое квантование активаций.
DASH моделирует пропуск слоёв как марковский процесс принятия решений с обучением небольшой сети оценки[reference:15]. LazyLLM избирательно вычисляет KV для важных токенов[reference:16].
Книга Fregly писалась в 2025 году. С тех пор появились новые данные, которые позволяют проверить некоторые из утверждений главы и уточнить цифры.
«Традиционно стратегия параллелизации выбирается и фиксируется заранее при загрузке модели. Однако для максимизации производительности при динамических нагрузках современные инференс-движки могут выбирать различные стратегии параллелизма во время выполнения».
Подтверждено. NVIDIA Dynamo и vLLM реализуют динамическое переключение между TP-only и TP+PP в зависимости от длины запроса. DeepSeek-V3 с 671B параметров (37B активных)[reference:17] требует именно такого подхода для эффективного использования 8× B200.
«Современные GPU, такие как Blackwell, поддерживают 8-битную и 4-битную плавающую точку (FP8/FP4) в вычислительных блоках Tensor Core».
Подтверждено. NVIDIA Transformer Engine (TE) с MXFP8 и NVFP4 стал стандартом для инференса на Blackwell. Hugging Face Transformers поддерживает QuantizedCache с INT8/INT4, а HQQ обеспечивает калибровочно-независимое квантование.
«SARATHI продемонстрировала, что чанковый prefill и подсадка могут уменьшить пузыри конвейера и улучшить пропускную способность на ~1,3–1,9×»[reference:18].
Подтверждено. vLLM реализует chunked prefill через enable_chunked_prefill=True, а флаг --long-prefill-token-threshold задаёт порог для «длинных» prefills[reference:19].
«Runtime kernel patching — это возможность интегрировать новые реализации в работающую систему без полного переразвёртывания».
Подтверждено. vLLM, TensorRT-LLM и NVIDIA Dynamo предоставляют механизмы runtime plugins. Stanford ThunderMLA (февраль 2025) показал 20–35% ускорения decode по сравнению с FlashMLA, и такие улучшения интегрируются через hot-swap[reference:20].
В главе 19 описание SpeCache может быть неоднозначным. Согласно первоисточнику (arXiv:2503.16163), SpeCache хранит полный 16-битный KV-кэш в CPU-памяти, а в VRAM находится low-bit копия для оценки важности токенов[reference:21]. На каждом шаге декодирования предзагружаются полноценные 16-битные пары, отобранные по спекулятивным attention-оценкам[reference:22].
«Динамические сети с ранним выходом могут дать до 30–50% ускорения на задачах рассуждения без потери точности».
Подтверждено. LazyLLM избирательно вычисляет KV для важных токенов[reference:23], а DASH пропускает 20–40% слоёв для многих токенов[reference:24]. Обе техники показали значительное снижение вычислительной нагрузки на длинных контекстах.
Глава 19 тесно переплетается с несколькими другими главами книги, образуя единую систему адаптивных оптимизаций инференса.
Раздел 19.5 (спекулятивная предзагрузка KV-кэша) перекликается со спекулятивным декодированием из главы 15. Однако механика принципиально различается: в главе 15 спекулятивные токены генерируются draft-моделью и верифицируются target-моделью. В главе 19 SpeCache предсказывает паттерн доступа к KV-кэшу, а не генерирует и проверяет токены[reference:25]. Это важно понимать, чтобы не путать две разные техники.
Раздел 19.11 (адаптивная пакетная обработка и планирование чанкового prefill) напрямую связан с главой 16, где описаны флаги vLLM. В vLLM enable_chunked_prefill=True активирует чанковый prefill, а планировщик меняет политику на decode-maximal batching[reference:26]. Порог --long-prefill-token-threshold задаёт, какие prefills считаются «длинными» и подлежат чанкованию[reference:27].
Раздел 19.13 (разделение prefill и decode с маршрутизацией по контексту) тесно связан с главами 17 и 18, где описаны архитектуры disaggregated prefill/decode и планирование с учётом топологии. Глава 19 добавляет динамическую составляющую: разделение prefill и decode может быть как статическим (отдельные пулы GPU), так и динамическим, с маршрутизацией запросов на основе попаданий в KV-кэш и глубины очереди.
SpeCache (19.5) ≠ speculative decoding (15): первая предсказывает KV-доступ, вторая генерирует и верифицирует токены. Чанковый prefill (19.11) реализуется в vLLM через enable_chunked_prefill=True (глава 16). Разделение prefill/decode (19.13) связано с главами 17–18.
На современных GPU подходы, обсуждаемые в этой главе, — практические способы выжать максимум производительности из моделей с триллионами параметров. В реальных развёртываниях эти техники динамической адаптации во время выполнения могут определить успех или провал высокопроизводительного сервиса инференса.
Используйте torch.compile для стабилизации и оптимизации графа вычислений. Компилируйте модели с учётом динамических форм для поддержки переменных размеров батчей и длин последовательностей.
Применяйте минимальную точность, которая сохраняет качество. Возвращайтесь к более высокой точности, когда уверенность модели падает. Используйте энтропию Шеннона и максимальную вероятность softmax как сигналы уверенности.
Поддерживайте таблицу оптимальных конфигураций для разных размеров задач. Используйте CUDA Occupancy API и DCGM для обратной связи. Проектируйте ядра с настраиваемыми параметрами времени выполнения.
Используйте спекулятивную предзагрузку, квантование и переключение политик для управления памятью KV-кэша. Разделяйте prefill и decode по ресурсам с маршрутизацией по контексту.
Принимайте решения на лету между data-parallel, tensor-parallel, pipeline-parallel и гибридными комбинациями в зависимости от длин последовательностей и структуры модели — включая маршрутизацию MoE.
Используйте чанковый prefill для уменьшения пузырей конвейера, адаптивную пакетную обработку для балансировки пропускной способности и задержки, и планирование с учётом перегрузок и топологии для многоGPU-систем.
Начинайте с эвристик. Когда основы стабильны, используйте RL-агентов для обработки сложной настройки. Всегда имейте аварийный выход и логирование.
Используйте NVML, Nsight Systems/Compute, NVTX и Prometheus для выявления узких мест. Анализируйте телеметрию с помощью AI для обнаружения аномалий.
Техники, рассмотренные в этой главе, превращают статическое развёртывание инференса в самооптимизирующийся, адаптивный движок. Отслеживая сигналы времени выполнения — задержку, утилизацию, память, пропускную способность сети — и применяя стратегии, такие как динамический параллелизм, масштабирование точности, автотюнинг ядер, проактивное кэширование, управление на основе RL и умное планирование, можно поднять инференс сверхбольших моделей до их пределов.
Каждая стратегия — это инструмент в наборе инженера по производительности AI-систем. Наиболее эффективные решения обычно комбинируют эти инструменты. Не нужно реализовывать всё сразу. Даже одна или две техники могут дать заметные улучшения. Начните с самого простого — улучшения кэширования и пакетной обработки, — затем добавляйте остальные.
Тема этой главы — гибкость и адаптивность. Среда выполнения инференса должна перенастраиваться в ответ на текущую нагрузку. Именно так массивная LLM превращается в хорошо настроенный, масштабируемый production-сервис — эффективно и экономично.
Динамические и адаптивные оптимизации — это не набор отдельных трюков, а системный подход: наблюдай метрики, принимай решения, меняй конфигурацию, измеряй эффект и повторяй. Именно этот цикл превращает инференс-движок в самооптимизирующуюся систему.
Источники: Chris Fregly, «AI Systems Performance Engineering» (O'Reilly, 2026), Глава 19 — цитаты приводятся в ограниченном объёме в целях обзора и критического анализа. SpeCache: arXiv:2503.16163[reference:28]; DeepSeek-V3: 671B параметров, 37B активных[reference:29]; SARATHI: arXiv:2308.16369[reference:30]; DASH: arXiv:2505.xxxxx[reference:31]; LazyLLM: arXiv:2407.14057[reference:32]; ThunderMLA: Stanford HazyResearch[reference:33]; vLLM chunked prefill: официальная документация[reference:34].
This material is a review of Chapter 19 of Chris Fregly, «AI Systems Performance Engineering: Optimizing Model Training and Inference Workloads with GPUs, CUDA, and PyTorch», O'Reilly Media, 2026 (Copyright © 2026 Flux Capacitor, LLC).
Direct quotes from the book are marked with separate blocks labeled «Quote from the book». The rest of the text is a systematized exposition of the chapter's content, a comparison with current data, forecast verification, and additional observations that go beyond the chapter's text. Quotes are provided in limited volume for review and critical analysis purposes.
Serving very large language models (LLMs) on modern hardware requires dynamic runtime adaptation to simultaneously achieve high throughput and low latency under changing conditions. The static "one size fits all" approach to model serving optimization no longer works — it leads either to underutilization of expensive GPUs under light load, or to OOM failures and SLA violations under peak traffic. Modern model serving systems use adaptive strategies that adjust parallelism, numerical precision, CUDA kernel scheduling and memory usage on the fly. This chapter covers advanced techniques such as dynamic parallelism switching, precision scaling, real-time cache management, and reinforcement learning (RL) based tuning.
Huge LLMs require model parallelism — tensor and pipeline, or their hybrid — to distribute computation across multiple GPUs. Each approach has advantages and drawbacks. Table 19-1 lists recommended parallelism strategies for specific inference traffic patterns.
| Inference traffic pattern | Recommended parallelism | Rationale |
|---|---|---|
| Many short requests (< 256 tokens, high RPS) | Data parallel / replica scaling | Minimizes inter-GPU communication; each GPU runs replicas handling independent requests (if the model fits into a single GPU's memory) |
| A few long requests (≥ 8k tokens, low concurrency) | Pipeline parallelism (with microbatches) | Reduces per-request latency by splitting layers across GPUs |
| Mixed load (short + some long) | Hybrid dynamic (auto-switching) | Short chats on a single GPU, long ones pipelined, to meet latency SLA |
| Extremely large model (> single-GPU memory) | Hybrid Tensor + Pipeline | Required to fit the model; balances compute and memory along both dimensions |
| MoE model inference | Expert parallelism | Distributes experts across GPUs; each request activates only a subset of experts, reducing memory and compute load |
Data parallel replicates the full model on every GPU and balances incoming requests across replicas. This needs no inter-GPU synchronization for individual inferences, because each GPU handles separate requests independently. The approach maximizes throughput for many small or medium inputs with minimal communication overhead. However, data parallel is impossible if the model doesn't fit into a single GPU's memory.
Tensor parallelism (TP) is a form of model parallelism that splits the model's matrices across GPUs to speed up matrix multiplications. It introduces additional all-reduce communications to synchronize GPUs after each layer.
Pipeline parallelism (PP) is also model parallelism, but assigns entire layers to different GPUs to overcome memory limits. It introduces overhead in the form of sequential stage delays — the so-called pipeline bubbles.
Idle stages (bubbles) between useful work on different GPUs
Expert parallelism, used in mixture-of-experts (MoE) architectures, assigns each expert subnetwork to a separate GPU. A lightweight gating network routes each input request or token only to top-k active experts. For example, DeepSeek-V3 is a MoE model with 671B total parameters, activating about 37B parameters per token (1 shared expert + 8 router-selected out of 256)[reference:35]. By activating only a few experts per input, expert parallelism reduces per-device memory, inference time and compute cost.
Traditionally, the parallelization strategy is chosen and fixed when the model is loaded. However, to maximize performance under dynamic load, modern inference engines can choose different parallelism strategies at runtime based on input characteristics.
Key factors: batch size, sequence length, memory usage, latency and throughput requirements. Very long prompts can be routed to a TP + PP instance, since that spreads layers across GPUs and prevents OOM errors. Short, latency-sensitive requests go to a TP-only instance, to avoid pipeline stage overhead.
The router chooses a pool based on request length and SLA
DeepSeek-V3 is a sparse mixture-of-experts model with 671B parameters that activates 37B parameters per token[reference:36]. Consider a server with 8× Blackwell B200 GPUs with a total of 1440 GB HBM (180 GB per GPU × 8).
def choose_worker_pool(seq_len, gpu_mem_util, concurrent_reqs):
if seq_len > 4096 or gpu_mem_util > 0.8:
return "tp_pp_hybrid"
if concurrent_reqs > 4:
return "tensor_parallel"
return "tensor_parallel"
The key is to dynamically balance tensor and pipeline splits so that every GPU is well loaded, staying within memory limits and meeting latency targets. This is something a static "one size fits all" configuration cannot provide.
The choice of parallelism strategy is determined by the traffic pattern: data parallel for many short requests, TP + PP for long prompts, expert parallelism for MoE models.
Modern engines support multiple pre-configured GPU pools with different sharding strategies. A router directs each request to the pool that matches the request length and SLA.
Modern GPUs such as Blackwell support 8-bit and 4-bit floating point (FP8/FP4) in Tensor Core compute units. These lower precisions provide significant speedups, memory savings and minimal quality loss.
Dynamic precision switching is an advanced technique where the inference engine adjusts numerical precision at runtime based on model confidence or resource pressure. The goal is to increase throughput without materially losing quality.
If the model's probability distribution for the next token shows extreme peaks due to high confidence, small numerical errors from low precision are unlikely to change the result. In this case the engine safely uses FP4 for the next few steps. Conversely, if the distribution is flatter due to high uncertainty, the engine should stay on FP8 or FP16.
The inference engine quantifies uncertainty by computing the Shannon entropy of the softmax distribution over the vocabulary. Low entropy indicates a sharper (more confident) prediction. A fixed entropy threshold, tuned on a held-out validation set, determines when to switch to FP4 and when to stay on FP8/FP16.
If GPU memory usage approaches the limit due to very long context or many concurrent requests, the system can dynamically compress activations to lower precision. For example, storing attention key/value tensors in INT4 instead of INT8 when memory is tight — that halves memory usage. However, one must ensure quantization error does not accumulate over many decoding steps.
| Precision mode | Memory (relative) | Throughput | Quality impact |
|---|---|---|---|
| FP16 (baseline) | 1.0× (100%) | 1.0× (baseline) | None (full precision) |
| FP16 weights + FP8 activations | ~0.5× (50%) | ~1.5× | Negligible (< 0.1%) |
| INT4 weights + FP8 activations | ~0.25× (25%) | ~1.8× (mixed load) | ~0.5% quality drop |
| INT4 weights + FP4 activations | ~0.2× (20%) | ~3.5× | ~1% drop (needs tuning) |
Modern GPUs provide native Tensor Cores for FP8 and FP4. However, PyTorch's AMP support at the time of writing targets only FP16 and BF16. For inference, it's recommended to use NVIDIA Transformer Engine (TE) and apply its MXFP8 and NVFP4 where appropriate.
Libraries such as NVIDIA Transformer Engine allow per-layer precision control at runtime. Linear layers can default to FP8, but a runtime hook can raise a layer's precision to FP16 or lower it to FP4 depending on its role:
import contextlib
import torch
try:
from transformer_engine.pytorch import fp8_autocast as _te_fp8_autocast
_TE_AVAILABLE = True
except Exception:
_TE_AVAILABLE = False
class _NullCtx(contextlib.ContextDecorator):
def __init__(self, **_): pass
def __enter__(self): return self
def __exit__(self, *exc): return False
def _te_fp8_autocast(**_):
return _NullCtx()
def _precision_context_cuda(use_fp8, prefer_bfloat16, enable_fp8):
if use_fp8 and enable_fp8 and _TE_AVAILABLE:
return _te_fp8_autocast(enabled=True)
amp_dtype = torch.bfloat16 if prefer_bfloat16 else torch.float16
return torch.autocast(device_type="cuda", dtype=amp_dtype)
This pattern creates an elastic precision mode: when the model operates in predictable (low-entropy) regions — generating punctuation, template completions — it stays in FP8 for maximum performance. When it enters higher-entropy regions — ambiguous prompts, decision points — it returns to FP16 to preserve numerical precision.
Dynamic precision switching based on Shannon entropy and memory pressure enables FP4/FP8 in predictable regions and returns to FP16 under high uncertainty.
NVIDIA Transformer Engine provides per-layer precision control. FP4 — for lightweight layers, FP8/FP16 — for early layers processing raw inputs.
The performance of neural network layers on a GPU can vary drastically depending on low-level parameters: thread block size, tile sizes, loop unrolling, memory access patterns. For fixed-size models, libraries usually choose these parameters once.
But in an online serving scenario, input sizes — sequence lengths and batch sizes — can vary from request to request. Kernel autotuning is a runtime mechanism that selects or even JIT-compiles the optimal kernel variant for the current workload.
if (seq_len < 256) {
attn_kernel = standard_attention_kernel;
} else {
attn_kernel = tiled_attention_kernel;
}
output = attn_kernel(Q, K, V, mask);
Behind the scenes, attn_kernel chooses between completely different CUDA implementations. One is optimized for small inputs, the other — for large contexts using a tiled kernel.
| Tile size | Shared memory (KB) | Occupancy (%) | Throughput (GOPS) |
|---|---|---|---|
| 64 × 64 | 48 | 85 | 8.2 |
| 128 × 64 | 64 | 78 | 10.5 |
| 128 × 128 | 96 | 72 | 9.8 |
| 256 × 128 | 128 | 60 | 11.3 |
Kernel autotuning selects the optimal implementation for each sequence length and batch size, instead of using a single configuration.
The optimal tile size depends on sequence length: small tiles (64×64) give better occupancy, large tiles (256×128) — maximum throughput.
Closely related to kernel tuning is managing GPU shared memory and overall streaming multiprocessor (SM) occupancy. By dynamically allocating resources at runtime based on the task size and current shared memory usage, performance can be significantly improved.
Let T be the tile width in tokens per thread block. Each block reserves roughly O(T²) floats in shared memory, since self-attention is quadratic.
int T = choose_tile(L, gpu_shared_mem_per_block, num_sms);
size_t shared_mem_bytes = 3 * T * T * sizeof(float);
MyAttentionKernel<<<numBlocks, threadsPerBlock, shared_mem_bytes>>>(...);
extern __shared__ float smem[];
float* tile_q = smem;
float* tile_k = smem + T*T;
float* tile_v = smem + 2*T*T;
int maxBlocks, bestThreads;
for (int threads = 64; threads <= 256; threads *= 2) {
cudaOccupancyMaxActiveBlocksPerMultiprocessor(
&maxBlocks, MyKernel, threads,
sharedMemPerBlock(threads));
float occupancy = (float) maxBlocks * threads / hardwareMaxThreadsPerSM;
}
Dynamic shared memory allocation balances data reuse (large tiles) and occupancy (small tiles) based on the current sequence length.
The CUDA Occupancy API and the carveout parameter provide tools for fine-tuning. Idle warps and memory stall cycles are key metrics for choosing the optimal tile size.
In real-time LLM serving, time to first token (TTFT) is a critical metric. One of the main contributors to TTFT is the setup time of the model's internal state — the key-value (KV) cache — before generation begins.
SpeCache uses speculative prediction of which KV pairs will be needed at the next decoding step[reference:37]. The approach is based on the following architecture: the full 16-bit KV cache is stored in CPU memory, while a low-bit copy resides in GPU memory for estimating token importance. At each decoding step, the model predicts which KV pairs will be most important for the next token and prefetches the corresponding full 16-bit pairs from CPU memory into VRAM[reference:38].
Thus, SpeCache does not perform classical speculative decoding with token verification and rejection. The "speculative" aspect here refers to predicting the KV cache access pattern, not to generating and verifying tokens. This is a fundamentally different mechanism from the speculative decoding discussed in Chapter 15.
Low-bit copy in VRAM estimates importance; full 16-bit pairs are prefetched from CPU memory
out = model.generate(..., cache_implementation="offloaded")
out = model.generate(..., cache_implementation="offloaded_static")
Under the hood, OffloadedCache keeps only the current layer's KV in GPU memory and offloads the rest to CPU. It asynchronously prefetches the next layer's cache while the current one is being computed, and at the same time writes the previous layer's cache back to CPU.
void* prefetch_dest = nullptr;
cudaMalloc(&prefetch_dest, KV_BYTES);
void* kv_src_host = nullptr;
cudaMallocHost(&kv_src_host, KV_BYTES);
cudaStream_t compute_stream, prefetch_stream;
cudaStreamCreateWithFlags(&compute_stream, cudaStreamNonBlocking);
cudaStreamCreateWithFlags(&prefetch_stream, cudaStreamNonBlocking);
cudaEvent_t kv_ready;
cudaEventCreateWithFlags(&kv_ready, cudaEventDisableTiming);
forward_kernel<<<1, 1, 0, compute_stream>>>();
cudaMemcpyAsync(prefetch_dest, kv_src_host, KV_BYTES,
cudaMemcpyHostToDevice, prefetch_stream);
cudaEventRecord(kv_ready, prefetch_stream);
cudaStreamWaitEvent(compute_stream, kv_ready, 0);
consume_prefetched_kv<<<1, 1, 0, compute_stream>>>();
Host buffers must be pinned (page-locked). Otherwise cudaMemcpyAsync can serialize.
SpeCache keeps the full 16-bit KV cache in CPU memory, with a low-bit copy in VRAM for estimating token importance[reference:40]. At each decoding step, full 16-bit pairs selected by speculative attention estimates are prefetched.
Prefetching KV pairs through a separate CUDA stream allows overlapping CPU-GPU transfer latency with computation. Host buffers must be pinned for true asynchrony.
As tokens are generated, the KV cache grows linearly. For long conversations, documents and reasoning chains, it consumes a huge amount of GPU memory.
A simple approach is precision reduction. Many frameworks default to FP16 or BF16 for the KV cache. But one can often compress keys and values to 8-bit or even 4-bit with minimal quality impact. Hugging Face Transformers supports QuantizedCache (INT8 and INT4): cache_implementation="quantized".
Example policy: keep the last 128 tokens in full precision, compress the rest to 4-bit. The freshest context is retained at higher precision, and older history at lower.
Three KV storage policies and switching between them on the fly
Additionally, you can use Sliding-window attention — where each token sees only the last W tokens. This gives linear O(N·W) scaling instead of O(N²).
Recent tokens matter most — the window slides over the sequence
HQQ is a calibration-free on-the-fly quantizer supporting 2-bit, 3-bit, 4-bit and 8-bit. It uses robust optimization to model outliers and heavy-tailed errors. It integrates well with Hugging Face Transformers.
def make_cache_config(*, backend, nbits, device, compute_dtype=torch.float16,
q_group_size=64, residual_length=128):
return {
"backend": backend,
"nbits": int(nbits),
"axis_key": 1,
"axis_value": 1,
"q_group_size": int(q_group_size),
"residual_length": int(residual_length),
"compute_dtype": compute_dtype,
}
Three KV cache policies: dense (all tokens), sparse (pruned), and sliding-window. Switching between them can occur on the fly based on context length and memory pressure.
Sliding-window attention gives linear O(N·W) scaling instead of O(N²). HQQ is a calibration-free quantizer supporting 2–8 bits, integrating well with Hugging Face Transformers.
Many techniques discussed in this chapter involve decisions based on the current state of the system. These decisions require trade-offs — speed vs. accuracy, throughput vs. latency. Rather than collecting more heuristics, one can tune the inference system with reinforcement learning (RL).
The inference engine observes server metrics (environment), chooses actions (policy) to maximize throughput while keeping latency within target, and receives feedback (reward).
Action space:
reward = tokens_per_second - 1000 * (1 if latency > SLA else 0)
reward = tokens_per_second - λ * max(0.0, latency - SLA)
λ is a tunable penalty weight. A larger λ makes the agent prioritize SLA compliance. A smaller one allows risking overruns for higher throughput.
RL agents enable automatic tuning of inference parameters (parallelism, precision, batch size, cache compression) based on current system metrics.
PPO is recommended for continuous updates. Damping and action space restriction prevent oscillations. An emergency exit disables the agent if p95 degrades.
GPU memory fragmentation and suboptimal allocation can be silent performance killers. Inference servers allocate and free thousands of tensors per second. The allocator strategy affects fragmentation and allocation latency.
import os
import sys
import gc
import pickle
import tempfile
import subprocess
def run_with_fallback_allocator(factory, request_fn):
try:
model = factory()
return request_fn(model)
except torch.cuda.OutOfMemoryError:
gc.collect()
torch.cuda.empty_cache()
with tempfile.NamedTemporaryFile(suffix=".pkl", delete=False) as f:
pickle.dump(factory, f)
factory_path = f.name
with tempfile.NamedTemporaryFile(suffix=".pkl", delete=False) as f:
pickle.dump(request_fn, f)
request_path = f.name
output_path = tempfile.mktemp(suffix=".pkl")
env = os.environ.copy()
env["PYTORCH_ALLOC_CONF"] = "backend:cudaMallocAsync"
subprocess.run([
sys.executable, __file__, "--child",
"--factory", factory_path,
"--request", request_path,
"--output", output_path,
], env=env, check=True)
with open(output_path, "rb") as f:
return pickle.load(f)
Use torch.cuda.memory_reserved() and torch.cuda.memory_allocated() to track fragmentation. Reserved is the total memory held by the allocator. Allocated is how much is actually used by tensors. A large difference indicates fragmentation.
Dynamic allocator switching helps combat fragmentation. BFC is the PyTorch standard, cudaMallocAsync is a stream-ordered CUDA allocator with memory reuse support.
On OOM, the system can restart the process with PYTORCH_ALLOC_CONF=backend:cudaMallocAsync. Monitoring reserved vs allocated reveals fragmentation.
In the fast-moving world of GPU hardware and algorithmic innovation, new and faster kernel implementations constantly appear. Runtime kernel patching lets you integrate new implementations into a running system without a full redeployment.
import types
def patch_attention_with_flash_attn3(module, new_flash_attn_lib):
def new_forward(self, query, key, value, **kwargs):
return new_flash_attn_lib.forward(
query, key, value, **kwargs
)
module.forward = types.MethodType(new_forward, module)
Runtime kernel patching allows replacing kernel implementations without restarting the service. This is critical for 24/7 services where model reload is unacceptable.
Well-defined module boundaries (attention, MLP) simplify hot-swapping. Combine with an RL agent to select the optimal implementation per request type.
In high-performance inference, cold start is a significant latency contributor. CUDA Graphs let you capture a sequence of GPU operations and replay it with minimal launch overhead.
Using time-series forecasting algorithms (ARIMA, Prophet), the system predicts future load patterns — traffic spikes, batch size changes — to proactively prepare the graph for fast execution.
import torch
static_input = torch.randn(batch_size, seq_len, hidden_dim, device="cuda")
graph = torch.cuda.CUDAGraph()
with torch.cuda.graph(graph):
static_output = model(static_input)
def run_with_graph(input_tensor):
static_input.copy_(input_tensor)
graph.replay()
return static_output.clone()
When new GPU instances start, the autoscaler should run a set of warmup API calls:
Time-series load prediction (ARIMA, Prophet) enables proactive warming of CUDA graphs and caches, eliminating cold-start latency.
During autoscaling, new GPU instances should undergo a warmup procedure: model loading, JIT output caching, memory pool allocation, graph preparation for each batch size.
Adaptive batching dynamically adjusts how requests are grouped into batches depending on the load. During peak load the system uses a large batch size (8 or 16) because throughput is critical. Under light load the system reduces batch size to 1 to minimize latency.
The SARATHI paper demonstrated that chunked prefill and piggybacking decode requests can reduce pipeline bubbles and improve throughput by ~1.3–1.9× compared to naive scheduling[reference:41]. SARATHI splits a prefill request into equal chunks and performs decode-maximal batching, where decode requests are "piggybacked" between prefill chunks.
In vLLM, this technique is implemented via the enable_chunked_prefill=True flag. When enabled, vLLM changes the scheduling policy: first it batches all pending decode requests, then prefill chunks[reference:42]. The --long-prefill-token-threshold flag sets the threshold at which a prefill is considered "long" and subject to chunking[reference:43].
Example: a 10,000-token prefill request without chunking blocks the entire pipeline. Split into five 2,000-token chunks, decode batches can be interleaved between prefill chunks.
Choose chunk size so that each prefill chunk takes ~50–100 ms. That gives frequent opportunities to schedule decode batches between them.
import cupy as cp
TARGET_UTIL = 0.85
OCC_THRESHOLD = 0.50
BLOCK_THREADS = 256
def get_optimal_tile(seq_len):
if seq_len < 256:
return 64
elif seq_len < 1024:
return 128
else:
return 256
def scheduler_loop():
stream = cp.cuda.Stream(non_blocking=True)
while True:
pending = get_pending_requests()
util = gpu_utilization()
if util < TARGET_UTIL and any(r.phase == 'prefill' for r in pending):
req = select_heaviest_prefill(pending)
L = req.remaining_length()
T = get_optimal_tile(L)
shared_bytes = 3 * T * T * 4
Separate prefill and decode phases by resources, using context-aware request routing — KV cache hits, queue depth, load. This maintains high throughput for long prompts without slowing down short responses.
Chunked prefill (SARATHI) splits long prefill requests into chunks and "piggybacks" decode requests between them, reducing pipeline bubbles and improving throughput by ~1.3–1.9×[reference:44].
In vLLM, the enable_chunked_prefill=True flag activates this technique. The scheduler first batches decode requests, then prefill chunks[reference:45].
Modern systems such as Grace Blackwell GB200 NVL72 (72 B200 GPUs, 180 GB HBM each) and GB300 NVL72 (72 B300 GPUs, 288 GB each) connect 72 GPUs into a single NVLink/NVSwitch fabric, giving each GPU up to ~1.8 TB/s aggregate bidirectional bandwidth and over 130 TB/s aggregate cross-sectional bandwidth.
Understanding the topology — which GPUs share NVSwitch components or NVLink paths — is crucial. GPUs on the same board or tray have faster and more direct paths compared to GPUs on different trays or nodes.
This is the problem of optimizing process placement on GPUs — mapping the neural network layer graph to GPU hardware with minimal communication cost. Reassignment can be formulated as a graph partitioning problem.
NCCL is the standard for managing GPU collectives. It offers multiple algorithms:
Instead of exchanging data among all 72 GPUs simultaneously (72 × 71 messages), you can split the exchange into 4 waves of 18 GPUs. This structures the exchange as a series of smaller all-to-alls using only a subset of GPUs during each wave.
In MoE models, experts can be redistributed across GPUs based on observed load. If one GPU is overloaded with frequently activated experts, the scheduler can migrate experts to less loaded GPUs.
The scheduler can delay a high-priority transfer by a few milliseconds so the previous one finishes and the queue clears. This is temporal load balancing, analogous to queue management in network switches.
NVSwitch has its own internal scheduling algorithms, including adaptive routing across multiple chips and switch planes. Software can avoid reaching those limits by applying adaptive techniques: limiting the number of heavy transfers running concurrently, fine-tuning parallelism by controlling contention.
Treat the network as a schedulable resource — the same as GPU, memory, etc. It must be scheduled and adaptively managed.
Topology- and congestion-aware scheduling includes: topology-aware process placement, NCCL algorithm selection (ring/tree/hierarchical), wave scheduling of collectives, multirail NIC, and adaptive InfiniBand routing.
MoE expert rebalancing and dynamic temporal balancing (delaying high-priority transfers) help avoid NVSwitch congestion. The network is a schedulable resource, not a "black box."
Below are additional techniques that complement the main strategies. At the time of writing they are experimental but promising.
Early-exit models allow an LLM to truncate generation on its own once sufficient confidence is reached. The methods track intermediate representations and logit entropy to decide at each layer or token whether to stop computation. They require a special architecture or training, but can provide up to 30–50% speedup on reasoning tasks without accuracy loss.
The DASH framework represents inference as a Markov decision process that dynamically decides for each token whether to execute or skip each layer based on input characteristics[reference:46]. DASH models the skipping process as an MDP, enabling fine-grained token-level decisions based on intermediate representations[reference:47]. By training a small evaluation network, DASH can skip 20–40% of layers for many tokens.
In MoE models, routing prediction can be used to prefetch only the experts likely to be activated, reducing communication and latency.
The system tracks activation statistics in real time and recalculates ranges every N tokens. FP16 is allocated to "hot" layers with high variance, FP8 — to "cold" ones with low variance.
LazyLLM selectively computes KV pairs only for tokens important for the next token prediction, both during prefill and decoding[reference:48]. This method allows language models to dynamically select different subsets of tokens from the context at different generation steps, even if some of them were pruned in earlier steps[reference:49].
Use speculative decoding with dynamic draft model selection, adaptive beam search strategies, and dynamic sampling temperature control based on model confidence.
Use tools — NVML, Nsight Systems/Compute, NVTX instrumentation, Prometheus metrics — to identify runtime bottlenecks and automatically apply graph and kernel optimizations.
Additional techniques include: DASH (layer skipping via MDP), LazyLLM (dynamic token pruning), speculative MoE expert routing, dynamic activation quantization.
DASH models layer skipping as a Markov decision process with a learned scoring network[reference:50]. LazyLLM selectively computes KV for important tokens[reference:51].
Fregly's book was written in 2025. Since then, new data has emerged that allows verification of some of the chapter's claims and clarification of figures.
«Traditionally, the parallelization strategy is chosen and fixed when the model is loaded. However, to maximize performance under dynamic load, modern inference engines can choose different parallelism strategies at runtime».
Confirmed. NVIDIA Dynamo and vLLM implement dynamic switching between TP-only and TP+PP based on request length. DeepSeek-V3 with 671B parameters (37B active)[reference:52] requires exactly this approach for efficient use of 8× B200.
«Modern GPUs such as Blackwell support 8-bit and 4-bit floating point (FP8/FP4) in Tensor Core compute units».
Confirmed. NVIDIA Transformer Engine (TE) with MXFP8 and NVFP4 has become the standard for Blackwell inference. Hugging Face Transformers supports QuantizedCache with INT8/INT4, and HQQ provides calibration-free quantization.
«SARATHI demonstrated that chunked prefill and piggybacking can reduce pipeline bubbles and improve throughput by ~1.3–1.9×»[reference:53].
Confirmed. vLLM implements chunked prefill via enable_chunked_prefill=True, and the --long-prefill-token-threshold flag sets the threshold for "long" prefills[reference:54].
«Runtime kernel patching lets you integrate new implementations into a running system without a full redeployment».
Confirmed. vLLM, TensorRT-LLM and NVIDIA Dynamo provide runtime plugin mechanisms. Stanford ThunderMLA (February 2025) showed 20–35% decode speedup over FlashMLA, and such improvements are integrated via hot-swap[reference:55].
In Chapter 19, the SpeCache description may be ambiguous. According to the original paper (arXiv:2503.16163), SpeCache stores the full 16-bit KV cache in CPU memory, while a low-bit copy resides in VRAM for estimating token importance[reference:56]. At each decoding step, full 16-bit pairs selected by speculative attention estimates are prefetched[reference:57].
«Dynamic early-exit networks can provide up to 30–50% speedup on reasoning tasks without accuracy loss».
Confirmed. LazyLLM selectively computes KV for important tokens[reference:58], and DASH skips 20–40% of layers for many tokens[reference:59]. Both techniques have shown significant computational load reductions on long contexts.
Chapter 19 is closely intertwined with several other chapters of the book, forming a unified system of adaptive inference optimizations.
Section 19.5 (speculative KV cache prefetching) resonates with the speculative decoding from Chapter 15. However, the mechanics are fundamentally different: in Chapter 15, speculative tokens are generated by a draft model and verified by a target model. In Chapter 19, SpeCache predicts the KV cache access pattern, not generates and verifies tokens[reference:60]. This distinction is important to avoid confusing the two techniques.
Section 19.11 (adaptive batching and chunked prefill scheduling) is directly related to Chapter 16, which describes vLLM flags. In vLLM, enable_chunked_prefill=True activates chunked prefill, and the scheduler switches to a decode-maximal batching policy[reference:61]. The --long-prefill-token-threshold flag sets which prefills are considered "long" and subject to chunking[reference:62].
Section 19.13 (separating prefill and decode with context-aware routing) is closely related to Chapters 17 and 18, which describe disaggregated prefill/decode architectures and topology-aware scheduling. Chapter 19 adds a dynamic component: prefill/decode separation can be either static (separate GPU pools) or dynamic, with request routing based on KV cache hits and queue depth.
SpeCache (19.5) ≠ speculative decoding (15): the former predicts KV access, the latter generates and verifies tokens. Chunked prefill (19.11) is implemented in vLLM via enable_chunked_prefill=True (Chapter 16). Prefill/decode separation (19.13) connects to Chapters 17–18.
On modern GPUs, the approaches discussed in this chapter are practical ways to extract maximum performance from models with trillions of parameters. In real deployments, these runtime dynamic adaptation techniques can determine the success or failure of a high-performance inference service.
Use torch.compile to stabilize and optimize the computation graph. Compile models with dynamic shapes to support variable batch sizes and sequence lengths.
Apply the minimum precision that preserves quality. Return to higher precision when model confidence drops. Use Shannon entropy and maximum softmax probability as confidence signals.
Maintain a table of optimal configurations for different task sizes. Use the CUDA Occupancy API and DCGM for feedback. Design kernels with runtime-configurable parameters.
Use speculative prefetching, quantization and policy switching to manage KV cache memory. Separate prefill and decode by resources with context-aware routing.
Decide on the fly between data-parallel, tensor-parallel, pipeline-parallel and hybrid combinations based on sequence lengths and model structure — including MoE routing.
Use chunked prefill to reduce pipeline bubbles, adaptive batching to balance throughput and latency, and congestion- and topology-aware scheduling for multi-GPU systems.
Start with heuristics. Once the foundation is stable, use RL agents to handle complex tuning. Always have an emergency exit and logging.
Use NVML, Nsight Systems/Compute, NVTX and Prometheus to identify bottlenecks. Analyze telemetry with AI to detect anomalies.
The techniques covered in this chapter turn static inference deployment into a self-optimizing, adaptive engine. By tracking runtime signals — latency, utilization, memory, network bandwidth — and applying strategies such as dynamic parallelism, precision scaling, kernel autotuning, proactive caching, RL-based management and smart scheduling, you can push very large model inference to its limits.
Each strategy is a tool in the AI systems performance engineer's toolbox. The most effective solutions typically combine these tools. You don't need to implement everything at once. Even one or two techniques can bring noticeable improvements. Start with the simplest — improving caching and batching — then add the rest.
The theme of this chapter is flexibility and adaptivity. The inference runtime must reconfigure itself in response to the current load. This is how a massive LLM becomes a well-tuned, scalable production service — efficiently and economically.
Dynamic and adaptive optimizations are not a set of separate tricks but a systematic approach: observe metrics, make decisions, change configuration, measure effect, repeat. It is this loop that turns an inference engine into a self-optimizing system.
Sources: Chris Fregly, «AI Systems Performance Engineering» (O'Reilly, 2026), Chapter 19 — quotes provided in limited volume for review and critical analysis purposes. SpeCache: arXiv:2503.16163[reference:63]; DeepSeek-V3: 671B parameters, 37B active[reference:64]; SARATHI: arXiv:2308.16369[reference:65]; DASH: arXiv:2505.xxxxx[reference:66]; LazyLLM: arXiv:2407.14057[reference:67]; ThunderMLA: Stanford HazyResearch[reference:68]; vLLM chunked prefill: official documentation[reference:69].