Контекст главы

Шестнадцатая глава книги Fregly посвящена профилированию, отладке и настройке производительности инференса LLM в масштабе. Глава охватывает полный цикл: от мониторинга и сбора метрик до конкретных техник оптимизации — динамического батчинга, планирования, квантования и оптимизаций на уровне приложений. Значительное внимание уделяется практическим аспектам эксплуатации inference-кластеров: диагностике неполадок, работе с KV-кэшем, управлению памятью и обеспечению качества обслуживания (QoS).

Итеративный цикл настройки: наблюдение — гипотеза — действие

Фундаментальная методология настройки производительности инференса — итеративный цикл:

1. Наблюдение. Сбор метрик: загрузка SM GPU, использование памяти, перцентили задержки (p50, p95, p99), доля попаданий в кэш, частота преемпций KV-кэша, загрузка copy-движков, пропускная способность NVLink. Без измерений любая оптимизация — лотерея.
2. Гипотеза. Формулирование конкретного предположения о первопричине. Не «модель медленная», а «загрузка SM ниже 50% из-за маленьких батчей в фазе декодирования» или «p99 задержка растёт из-за преемпции KV-кэша при пиковом трафике».
3. Исправление. Внесение изменения: увеличение размера батча, включение prefix caching, применение квантования FP8, перераспределение памяти GPU. Одно изменение за раз — иначе невозможно понять, что сработало.
4. Проверка. Прогон на staging-окружении с репрезентативной нагрузкой, затем выкатка в продакшен через канареечный релиз на 5–10% трафика. Слежение за Grafana: если пропускная способность выросла, а p99 задержка осталась в рамках SLO — оптимизация успешна.
5. Повтор. Устранив одно узкое место, обнаруживается следующее. Этот цикл никогда не заканчивается, потому что характеристики трафика меняются, модели обновляются, а оборудование стареет.

Профилирование: Nsight Systems и Nsight Compute

Nsight Systems — макроуровень

NVIDIA Nsight Systems — это профилировщик уровня системы, который показывает временную шкалу работы CPU и GPU с микросекундным разрешением. Видны потоки CPU, ядра GPU, события CUDA, коммуникации NCCL, операции memcpy — всё на единой шкале времени.

Ключевая техника — аннотация NVTX (NVIDIA Tools Extension). Пометка регионов кода метками вроде «Prefill», «Decode», «All-to-all» даёт цветные блоки на временной шкале, по которым сразу видно, сколько времени занимает каждая фаза и где возникают простои.

#include <nvtx3/nvToolsExt.h>
#include "my_model.hpp"
#include <vector>

#define NVTX_PUSH(name, argb) \
  do { \
    nvtxEventAttributes_t a{}; \
    a.version = NVTX_VERSION; \
    a.size = NVTX_EVENT_ATTRIB_STRUCT_SIZE; \
    a.colorType = NVTX_COLOR_ARGB; \
    a.color = (unsigned int)(argb); \
    a.messageType = NVTX_MESSAGE_TYPE_ASCII; \
    a.message.ascii = (name); \
    nvtxRangePushEx(&a); \
  } while (0)

#define NVTX_POP() do { nvtxRangePop(); } while (0)

void run_inference(
    const std::vector<Token>& prompt_tokens,
    Model& model,
    int num_generate_steps) {

  NVTX_PUSH("Prefill", 0xFF4F86F7);
  model.encode(prompt_tokens);
  NVTX_POP();

  for (int t = 0; t < num_generate_steps; ++t) {
    NVTX_PUSH("Decode", 0xFFFF8C00);
    Token next_token = model.decode_next();
    NVTX_POP();
  }
}

Для многопоточных сценариев критически важно именовать потоки CUDA и оборачивать каждый в отдельные диапазоны NVTX. Без именования потоков дорожки GPU на временной шкале выглядят как бессмысленная мешанина — с именованием сразу видно, перекрываются ли передача данных и вычисления.

Nsight Compute — микроуровень

Nsight Compute профилирует отдельные ядра CUDA. Можно сосредоточиться на конкретных частях ядра — например, на транзакциях памяти — через профилирование по секциям.

Менее известная, но крайне полезная функция — CUDA Program Counter (PC) Sampling. Она сэмплирует счётчики программы и выявляет горячие точки внутри ядер с минимальными накладными расходами. В отличие от полной трассировки, PC Sampling можно запускать на живых продакшен-серверах, не деградируя обслуживание.

PyTorch Profiler

PyTorch Profiler, основанный на библиотеке Kineto, предоставляет разбивку на уровне операторов для CPU и GPU. Он записывает формы тензоров, снимки памяти и поддерживает экспорт трассировок в Chrome Tracing / Perfetto для совместного анализа.

with torch.profiler.profile(
    activities=[ProfilerActivity.CPU, ProfilerActivity.CUDA],
    record_shapes=True,
    profile_memory=True,
    with_stack=True
) as prof:
    with profiler.record_function("train_step"):
        torch.cuda.nvtx.range_push("forward")
        outputs = model(input_ids, attention_mask=attention_mask, labels=labels)
        loss = outputs.loss
        torch.cuda.nvtx.range_pop()

        torch.cuda.nvtx.range_push("backward")
        loss.backward()
        torch.cuda.nvtx.range_push("optimizer_step")
        optimizer.step()
        torch.cuda.nvtx.range_pop()
        optimizer.zero_grad()
        torch.cuda.nvtx.range_pop()

print(prof.key_averages().table(
    sort_by="self_cuda_time_total",
    row_limit=10,
    fields=["self_cuda_time_total", "calls"]
))

Мониторинг: Prometheus, Grafana и DCGM

В продакшене постоянный мониторинг важнее эпизодического профилирования. Стек Prometheus + Grafana + DCGM — стандарт де-факто для GPU-кластеров.

Ключевые метрики GPU

Метрики уровня приложения

• Запросы в секунду (RPS) — основной индикатор нагрузки.
• Средняя задержка, p95, p99 — для соответствия SLO.
• Число активных контекстов — сколько одновременных сессий обслуживается.
• Пропускная способность в токенах/с — реальная скорость генерации.
• Использование и размер KV-кэша (общий и по узлам).
• Доля попаданий в префиксный кэш — vllm:gpu_prefix_cache_queries и vllm:gpu_prefix_cache_hits в vLLM.
• Частота преемпций KV-кэша — если растёт, памяти GPU не хватает.

Алерты в Alertmanager

Числовые значения — ориентировочные. Фактические пороги зависят от архитектуры GPU, условий охлаждения и SLO. Настраивайте их на основе baseline-измерений.

Корреляция метрик

• RPS + p99 задержка: всплеск задержки совпадает со всплеском трафика? Возможно, динамический батчинг увеличивает размер батча и задержку.
• Загрузка SM + загрузка copy-движка: SM простаивает, copy-движок загружен? Скорее всего, движок выгружает и загружает KV-кэш — памяти GPU не хватает.
• Счётчик ошибок NCCL + загрузка GPU по узлам: один узел на 0%, остальные на 90%? Ищите NCCL_COMM_FAILURE в логах.
• Доля попаданий в префиксный кэш + TTFT: падение hit-rate совпадает с ростом time-to-first-token? Возможно, кэш слишком мал.

Логи: что мониторить

Пример лога vLLM, указывающего на преемпцию KV-кэша:

WARNING 2025-05-03 14:22:07 scheduler.py:1057 Sequence group 0 is preempted by
PreemptionMode.RECOMPUTE because not enough KV cache space.
total_cumulative_preemption_cnt=1

Это означает, что движку не хватило памяти для KV-кэша, и он вытеснил активную последовательность. Результат — пересчёт KV при повторном планировании, дополнительные вычисления и рост задержки. Если такие сообщения появляются регулярно: увеличьте порог использования памяти GPU, уменьшите максимальное число батчированных токенов, включите PagedAttention или рассмотрите квантование KV-кэша.

Рецепты устранения неполадок

Отладка корректности

• Утечки памяти CUDA: использование памяти растёт монотонно. Используйте compute-sanitizer --tool memcheck your_binary для обнаружения ошибок памяти, гонок данных и выходов за границы.
• Сбои NCCL: один GPU показывает 0% загрузки, остальные 90%. Включите NCCL_DEBUG=WARN для подробных логов. Используйте ncclCommGetAsyncError с ncclCommAbort для обнаружения и обработки асинхронных ошибок.
• Аппаратные ошибки: ошибки ECC, повторы NVLink, повторы PCIe. Любое появление неисправимых ошибок ECC — повод немедленно оповестить и запланировать замену узла.

Полностековые оптимизации

Высокопроизводительный инференс требует скоординированных оптимизаций на каждом уровне стека. Проблема только на одном уровне — это удача. В реальности узкие места мигрируют между уровнями по мере их устранения.

Ключевая идея — синергии между уровнями. Квантование (модель) снижает потребление памяти, позволяя увеличить размеры батчей (среда выполнения) без OOM, что, в свою очередь, позволяет оркестрации объединять больше запросов на цикл GPU. Одна оптимизация на нижнем уровне может разблокировать несколько оптимизаций на верхних.

Динамический батчинг и планирование

Статический vs динамический vs непрерывный

Статический батчинг фиксирует размер батча, дополняет все последовательности до самой длинной и ждёт завершения каждого запроса перед возвратом. Результат: неограниченная задержка для ранних прибывших запросов и впустую потраченные циклы GPU на паддинг.

Динамический батчинг собирает батчи на лету. Система накапливает входящие запросы в течение короткого таймаута (обычно 2–10 мс) или до достижения целевого размера батча, затем отправляет батч на GPU.

Непрерывный батчинг (batching in-flight, iteration-level scheduling) — следующий шаг. Вместо ожидания завершения полных последовательностей, система пополняет батч на каждой итерации генерации токенов. Завершённые запросы вытесняются, новые немедленно занимают их слоты.

Непрерывное планирование

Непрерывное планирование относится к GPU как планировщик ОС к CPU. Если модель маленькая или размер батча ограничен SLO, одна копия модели может не насыщать GPU. Тогда движок запускает несколько потоков инференса одновременно на одном GPU.

PagedAttention в vLLM — конкретный пример. KV-кэш разбивается на страницы, и вычисления динамически группируются по страницам между активными последовательностями. Под тяжёлой нагрузкой это даёт почти 100% загрузку GPU.

Чанковый префилл

Когда промпты очень длинные (например, 20 000 токенов), можно разбить их на чанки и чередовать работу префилла и декодирования. Это называется чанковым префиллом (chunked prefill) или планированием без простоев.

Ключевое понимание: чанкирование не уменьшает общее число операций внимания. Self-attention выполняет N(N+1) ÷ 2 скалярных произведений Q × K на слой на голову — стоимость квадратична по N, O(N²). Чанкирование меняет только то, когда работа становится доступной декодеру. Выгода — сглаживание задержки и улучшение перекрытия, а не сокращение вычислений.

Планирование с учётом задержки

Наивный планировщик FIFO может перегрузить один GPU, если порядок прибытия не оптимален. Рассмотрим шесть промптов с длинами 6K, 2K, 6K, 2K, 2K, 2K и два GPU.

• FIFO: Батч 1 = [6K, 2K, 6K] → GPU 1 (≈38M операций Q×K), Батч 2 = [2K, 2K, 2K] → GPU 2 (≈6M). Критический путь — GPU 1.
• С учётом задержки: Батч 1 = [2K, 2K, 6K] → GPU 1 (≈22M), Батч 2 = [2K, 2K, 6K] → GPU 2 (≈22M). Критический путь снижается на 42%.

Системные оптимизации

Перекрытие коммуникаций и вычислений

В распределённом инференсе перекрытие коммуникаций с вычислениями — критический фактор производительности. Системы вроде GB200/GB300 NVL72 обеспечивают до ~130 ТБ/с агрегированной пропускной способности NVLink внутри стойки, но даже с такими скоростями синхронные коммуникации создают пузыри.

Базовый приём — отдельные потоки CUDA для передачи и вычислений:

cudaStream_t transfer_stream, compute_stream;
cudaStreamCreate(&transfer_stream);
cudaStreamCreate(&compute_stream);

// Передача в transfer_stream
cudaMemcpyAsync(h_logits, d_logits, bytes, 
                cudaMemcpyDeviceToHost, transfer_stream);

// Вычисление в compute_stream — перекрывается с передачей
my_kernel<<<grid, block, 0, compute_stream>>>(...);

// Синхронизация только когда данные нужны
cudaEventSynchronize(transfer_done_event);

Обязательно используйте page-locked (pinned) буферы хоста для передач по PCIe или NIC — драйвер CUDA может выполнять DMA напрямую и перекрывать с вычислениями.

Для MoE-моделей all-to-all можно перекрыть, разделив батч токенов пополам: пока первая половина обрабатывается экспертами, вторая отправляется. К моменту завершения первой половины данные второй уже прибыли на GPU-получатели.

GPUDirect RDMA позволяет сетевым адаптерам (InfiniBand NIC) читать и писать в память GPU напрямую, минуя память хоста. Это снижает задержку кросс-узловых передач KV-кэша или активаций экспертов и освобождает ядра CPU.

CUDA Graphs позволяют захватить асинхронные паттерны и снизить накладные расходы на запуск ядер — особенно полезно для повторяющихся паттернов декодирования.

Энергопотребление и тепловые ограничения

Непрерывная работа GPU на 100% вызовет тепловой троттлинг, если охлаждение недостаточно. Иногда работа на 90% с эффективными ядрами превосходит 100% с троттлингом.

• Ограничение частот через nvidia-smi -lgc — незначительно снижает пропускную способность, но обеспечивает стабильность.
• Небольшая задержка между батчами для снижения утилизации.
• Жидкостное охлаждение — стандарт для современных дата-центр GPU.
• Повышение лимита мощности через nvidia-smi -pl (если охлаждение адекватно).

Управление памятью

Страничный KV-кэш (PagedAttention в vLLM) разбивает KV-кэш на страницы и динамически выделяет их между последовательностями. Это устраняет фрагментацию памяти, свойственную подходам с фиксированным выделением.

Выгрузка KV-кэша (swap) в память CPU или на NVMe — вариант для редко используемых данных. LMCache в vLLM позволяет регулировать соотношение кэша на GPU и CPU. RadixAttention в SGLang лениво вытесняет наименее недавно использованные префиксы.

Однако выгрузка — не бесплатно: при высокой частоте свопа загрузка copy-движков растёт, а загрузка SM падает. Если виден всплеск copy-движка с одновременным падением SM — система свопит. Решения: применить квантование FP8/FP4 для уменьшения размера KV, увеличить выделение памяти GPU под кэш, изменить стратегию свопа.

Квантование для инференса в реальном времени

Спектр точности

Квантование только весов: GPTQ и AWQ

Квантование только весов сжимает веса до 4-битных целых, сохраняя активации в более высокой точности (FP8, FP16). Это снижает след памяти весов примерно в 4 раза по сравнению с FP16.

GPTQ применяет приближённый алгоритм второго порядка для послойного квантования весов до 3–4 бит за несколько часов на GPU. Для многих моделей 4-битное GPTQ сохраняет 99%+ точности FP16 с ~2× ускорением инференса.

AWQ (Activation-aware Weight Quantization) идентифицирует небольшую долю «значимых» каналов весов, которые производят большие величины активаций. AWQ применяет масштабирование на конкретный канал перед приведением весов к 4-битной точности.

Квантование активаций: SmoothQuant и W4A8

Квантование активаций сложнее, потому что их распределения сильно варьируются при разных входах. SmoothQuant — метод PTQ для 8-битного квантования активаций без обучения. Он переносит часть ошибки квантования с активаций на веса через простое масштабирование по строкам/столбцам.

Гибридный подход W4A8 сочетает 4-битные веса (INT4) с 8-битными активациями (FP8 или INT8). TensorRT-LLM и другие движки предоставляют этот путь с почти без потерь точности при правильной калибровке.

Практический рабочий процесс

1. Начните с инференса в FP16/BF16 — это базовая линия.
2. Оцените 8-битное квантование (FP8 или INT8) — обычно без потерь.
3. Примените 4-битное квантование только весов (GPTQ или AWQ) — для дополнительного сжатия.
4. Если нужна максимальная оптимизация — перейдите к W4A8 (INT4 веса + FP8 активации).
5. Калибруйте на данных, отражающих реальную нагрузку.
6. Проводите A/B-тестирование ответов квантованной модели против исходной.

Оптимизации на уровне приложений

Сжатие промптов

Механизм внимания трансформера имеет временную сложность O(n²), где n — размер входа в токенах. Сокращение промпта с 10 000 до 5 000 токенов уменьшает работу self-attention в 4 раза.

• Удаление избыточности: если пользовательский промпт содержит большой фрагмент текста, не относящийся к вопросу, вышестоящий компонент может суммировать или извлечь релевантные части.
• Сжатие истории диалога: при превышении 75% максимального контекста суммировать самые старые 25% сообщений.
• Специальные токены: системный промпт из 200 токенов можно заменить 10 специальными токенами, если модель дообучена распознавать их.
• Ссылки вместо содержимого: если промпт ссылается на ранее сохранённый документ, заменить полное содержимое ссылкой и обработать через RAG.

Кэширование префиксов

Кэширование префиксов (мемоизация префиксов) — одна из самых эффективных оптимизаций. Если несколько запросов разделяют общий префикс (системный промпт, документ, история диалога), KV-кэш для префикса вычисляется один раз и переиспользуется.

Без кэширования: N запросов длины L требуют O(N × L) работы. С кэшированием: O(N + L) — префикс обрабатывается один раз, каждый запрос добавляет только свой уникальный суффикс.

Каскадирование моделей

Не все запросы требуют мощности самой большой модели. Каскадирование — подход, при котором простые запросы направляются к меньшей, более быстрой модели, а сложные — к большой.

• Эвристическая: короткие фактические запросы → малая модель. Длинные промпты или запросы со словами «объясни», «проанализируй» → большая модель.
• Спекулятивная: запустить малую модель, оценить уверенность. Высокая уверенность — вернуть ответ. Низкая — вызвать большую модель.
• По нагрузке: при насыщении кластера большой модели направлять менее критичные запросы к малой.

Маршрутизация 60% запросов к модели в 10 раз меньше может сократить вычислительные затраты в ~5 раз.

Потоковая выдача

Модель генерирует по одному токену за раз (за исключением спекулятивного декодирования EAGLE и много-головного декодирования Medusa). Система группирует токены в батчи по 2–5, сбрасывает поток и отправляет через SSE/WebSockets/HTTP Streaming.

Человек читает со скоростью 4–13 токенов в секунду. Без потоковой выдачи пользователь смотрит на пустой экран 5 секунд, затем получает весь ответ разом. С потоковой выдачей первый батч токенов приходит через 0,25 секунды — и пользователь начинает читать сразу.

Технические детали: TCP_NODELAY для минимизации задержки; потокобезопасные очереди для передачи токенов сетевому потоку; управление контролем потока; HTTP/2 и постоянные соединения.

Проверка прогнозов и актуальные данные

Книга Fregly писалась в 2025 году, когда inference-стек активно развивался. Ниже представлено сопоставление прогнозов главы 16 с открытыми данными.

Дополнительные выводы

1. Производительность инференса — это не только скорость ядер. Глава 16 показывает, что производительность inference-системы определяется совокупностью факторов: эффективностью планирования, управления памятью, обработки запросов и мониторинга. Оптимизация только одного уровня даёт ограниченный эффект.
2. Мониторинг и профилирование — основа эксплуатации. Без систематического сбора метрик (DCGM, Prometheus, Grafana) и профилирования (Nsight Systems, Nsight Compute, PyTorch Profiler) невозможно выявить узкие места. Инструменты должны быть интегрированы в CI/CD и производственные дашборды.
3. Непрерывный батчинг и чанковый префилл — ключевые техники. Они максимизируют пропускную способность при сохранении низкой задержки. Continuous batching в сочетании с PagedAttention и chunked prefill стали стандартом для production-инференса.
4. Квантование — самый эффективный способ снижения стоимости. GPTQ, AWQ, SmoothQuant и W4A8 позволяют снизить потребление памяти и увеличить пропускную способность без значительной потери точности. Начинать следует с FP8, затем переходить к 4-битным весам.
5. Оптимизации на уровне приложений недооценены. Сжатие промптов, кэширование префиксов и потоковая выдача дают значительный эффект при минимальных затратах на внедрение. Они не требуют изменения модели или железа.
6. Перекрытие коммуникаций и вычислений — критический фактор масштабирования. GPUDirect RDMA, CUDA Graphs и отдельные потоки CUDA позволяют скрыть задержки передачи данных за вычислениями.
7. QoS и раннее отклонение запросов необходимы в масштабе. При перегрузке кластера лучше быстро отклонить низкоприоритетные запросы, чем допустить деградацию latency для всех. Mooncake и аналогичные системы реализуют admission control на основе прогнозирования задержки.

Ключевые положения

1. Профилирование — двухэтапный процесс. Nsight Systems для общей картины и поиска горячих ядер, Nsight Compute для детального анализа. NVTX-аннотации делают временную шкалу читаемой.
2. Мониторинг через Prometheus + Grafana + DCGM. Ключевые метрики: загрузка SM, использование памяти, перцентили задержки, доля попаданий в кэш. Корреляция метрик позволяет быстро локализовать первопричину.
3. Диагностика неполадок — систематический процесс. Таблица рецептов покрывает типовые сценарии: низкая загрузка SM, преемпции KV-кэша, высокая хвостовая задержка, OOM.
4. Полностековые оптимизации. Узкие места мигрируют между уровнями. Оптимизации на уровне модели, ядра, среды выполнения, оркестрации и развёртывания взаимодополняют друг друга.
5. Непрерывный батчинг и чанковый префилл. Continuous batching максимизирует пропускную способность. Chunked prefill предотвращает блокировку коротких запросов длинными.
6. Планирование с учётом задержки. Снижает критический путь на 42% в примере с шестью промптами.
7. Перекрытие коммуникаций и вычислений. Отдельные потоки CUDA, GPUDirect RDMA и CUDA Graphs скрывают задержки передачи данных.
8. Управление памятью. PagedAttention, LMCache и RadixAttention устраняют фрагментацию и позволяют обслуживать больше одновременных запросов.
9. Квантование. GPTQ и AWQ для 4-битных весов, SmoothQuant для 8-битных активаций, W4A8 для максимальной оптимизации. Начинайте с FP8, калибруйте на реальных данных.
10. Сжатие промптов и кэширование префиксов. Сжатие уменьшает работу self-attention квадратично. Кэширование префиксов превращает O(N × L) в O(N + L).
11. Каскадирование моделей. Маршрутизация 60% запросов к модели в 10 раз меньше сокращает затраты в ~5 раз.
12. Потоковая выдача. TCP_NODELAY, потокобезопасные очереди, HTTP/2. Первый батч токенов через 0,25 секунды вместо 5 секунд ожидания.
13. QoS и раннее отклонение. При перегрузке — быстро отклонить низкоприоритетные запросы, чтобы сохранить latency для остальных.

Источники: Chris Fregly, «AI Systems Performance Engineering» (O'Reilly, 2026), Глава 16 — цитаты приводятся в ограниченном объёме в целях обзора и критического анализа. Документация NVIDIA Nsight Systems, Nsight Compute, DCGM, vLLM, SGLang, TensorRT-LLM; NVIDIA Dynamo Documentation; MLPerf Training/Inference v5.0; Red Hat Developer «Deploying distributed AI inference: Blueprints & troubleshooting» (2026); NVIDIA Technical Blog «Benchmarking LLM Inference at Scale with AIPerf» (2026).