Inference Engineer Interview Questions
196 inference engineer interview questions with worked answers, organized by the twelve topics an inference loop tests. Every answer carries its arithmetic as a chain you can recompute, names the hardware and precision behind each number, and ends with how a panel scores it. The topics follow the path of a request, from the router and the engine loop down to the attention kernel and the GEMM.
Worked answers on inference mechanisms, with calculations, implementation details, and follow-up questions.
What inference engineer interviews test
Inference loops are built from five kinds of round. The proportions vary by team, but each one expects numbers, not adjectives.
- Implement-from-scratch coding
Attention with a KV cache, a paged block allocator, a continuous-batching scheduler or a speculative verifier, graded on invariants and edge cases.
- Inference math
KV bytes per token, decode latency floors, crossover batch sizes and fleet sizing, produced quickly, with units and a sanity check.
- Serving system design
An LLM API or multi-tenant platform sized from the load: tokens per second per phase, replicas, routing, autoscaling and cost per token.
- Kernels and GPU performance
FlashAttention's online softmax, GEMV bandwidth, occupancy and bytes in flight, tensor-core generations and profiler evidence.
- Debugging from metrics
A TTFT regression, an OOM at peak or nondeterministic outputs, decomposed into queue, prefill and decode time before any guess.
Inference Physics & Roofline Interview Questions
Bytes and FLOPs per token, why batch-1 decode reads every weight once per step, the ridge point and the batch size that crosses it, MBU versus MFU, and the millisecond latency floors you must produce before anyone draws a box.
- Why is batch-1 decode on a 70B model bandwidth-bound, and what is its latency floor on two H100s?OpenAI · Anthropic · NVIDIA
- Derive the FLOPs per token for a dense transformer from its config. Where does the 2N rule stop being accurate?OpenAI · Anthropic · Google DeepMind
- Why does prefill turn compute-bound after a few hundred tokens on an H100, and how long does a 4K-token prompt take on a 70B model?OpenAI · Anthropic · NVIDIA
- Decode attention intensity is fixed at the query-to-KV head ratio. Derive that, then tell me what MLA does to it.OpenAI · Anthropic · Google DeepMind
- A vendor claims 1,000 tokens per second per user for a 70B model on one H200. Sanity-check that from the roofline.Fireworks AI · Together AI · Baseten
- We are moving a 70B model from H100s to B200s. How do decode latency and throughput change, regime by regime?NVIDIA · OpenAI · Together AI
- DeepSeek V3 has 671B parameters but only 37B active. Why does a decode step at batch 64 read far more than 37B?OpenAI · Google DeepMind · xAI
- How would you measure MBU on a running server? What do you time, what do you count, and which numbers make the result wrong?NVIDIA · Fireworks AI · Together AI
- Your decode step runs at 60% of the bandwidth floor. Where does the rest of the time go? Build me the budget.OpenAI · Anthropic · NVIDIA
- Long context: at what sequence length do KV reads overtake weight reads for a 70B GQA model at batch 16?Anthropic · Google DeepMind · OpenAI
- Derive the curve of decode tokens per second against batch size, and show me where the knee is.OpenAI · Fireworks AI · Together AI
- Estimate the energy per output token for a 70B model on H100s from power and throughput. What moves the number most?Google DeepMind · Meta · NVIDIA
- An 8B model at batch 1 on an H100 should take 4.8 ms per step but takes 12 ms. Name the three likeliest causes, with numbers.OpenAI · Anthropic · NVIDIA
- Rank the decode-time arithmetic intensity of the QKV projection, attention, the MLP and the LM head. Which ones can fusion help?NVIDIA · OpenAI · Anthropic
- Design a roofline-based performance model that predicts TTFT and TPOT for any model, GPU and batch. Where will it be wrong?OpenAI · Anthropic · NVIDIA
KV Cache Interview Questions
The KV algebra across MHA, GQA, MQA and MLA, block tables and fragmentation, hash and radix prefix reuse, eviction, FP8 KV, CPU and NVMe offload, and the gpu_memory_utilization budget that decides how many sequences actually fit.
- Derive the KV cache footprint for MHA, GQA and MLA, then tell me how many 8K-token conversations fit on four H100s.OpenAI · Anthropic · Google DeepMind
- Why does a paged KV cache waste at most one block per sequence, and how would you pick the block size?OpenAI · Anthropic · Together AI
- Account for every gigabyte on a GPU running vLLM: weights, KV pool, activations, CUDA graphs and NCCL buffers.NVIDIA · Baseten · Modal
- FP8 KV cache: what scales does it need, where is the error largest, and how would you validate it on long-context tasks?NVIDIA · Anthropic · Google DeepMind
- Tensor parallelism beyond the KV head count replicates the cache. Derive the per-GPU KV at TP=16 for an 8-KV-head model.Meta · NVIDIA · OpenAI
- A model mixes sliding-window and global attention layers. Compute its KV footprint at 128K and tell me what the window actually saves.Google DeepMind · OpenAI · Mistral AI
- Design KV offload to CPU memory. When does fetching blocks back over PCIe beat recomputing them?Anthropic · Perplexity · Baseten
- Prefix-cache eviction: LRU by block or by request? And what does a radix tree change about it?Anthropic · Perplexity · Fireworks AI
- Why can MLA decode attend over the 512-dimensional latent without decompressing keys, and what does absorption cost in FLOPs?Google DeepMind · NVIDIA · Fireworks AI
- With parallel sampling at n=8, a 2K-token prompt and 500 output tokens each, how many KV blocks does copy-on-write save?OpenAI · Anthropic · Fireworks AI
- The KV pool is full and requests keep arriving. Compare admission control, preemption and spilling, and what each does to p99.OpenAI · Anthropic · Fireworks AI
- In prefill-decode disaggregation, how does KV transfer size scale, and what link bandwidth keeps the transfer off the critical path?NVIDIA · Fireworks AI · Together AI
- A model's config says max context 1M. Size the KV cache for a single 1M-token request on a GQA model and on an MLA model.Google DeepMind · Anthropic · Meta
- When you quantize the KV cache, why are keys more sensitive than values, and what per-channel or per-token scheme follows from that?Google DeepMind · Meta · NVIDIA
- Design a multi-tier KV cache across HBM, host DRAM and SSD for a fleet with heavy prefix reuse, and show the hit-rate math.Anthropic · OpenAI · NVIDIA
- A hybrid model mixes attention and recurrent layers. Budget its state and design safe prefix reuse and rollback.NVIDIA · Together AI · Fireworks AI
Continuous Batching & Scheduling Interview Questions
Static, dynamic and iteration-level batching, chunked prefill and the token budget, preemption by swap or recompute, priority and fairness, queueing arithmetic, and the goodput a scheduler is really optimising.
- Quantify what continuous batching buys over static batching, then pick the chunked-prefill token budget and defend it.Anthropic · OpenAI · Google DeepMind
- What is goodput for an LLM server, and how can pushing throughput higher make it worse?Anthropic · OpenAI · Fireworks AI
- Use Little's law on an LLM server: how many requests are in flight at 10 per second, and does the KV cache hold them?OpenAI · Anthropic · Modal
- Preempt by recompute or by swapping KV to host memory? Derive the break-even from PCIe bandwidth and prefill speed.Anyscale · Red Hat · Databricks
- Where does head-of-line blocking come from under FCFS, when does shortest-job-first help, and how does it work without job lengths?Anthropic · Perplexity · Together AI
- Interactive and batch traffic share one engine. How do you stop batch jobs from wrecking interactive TTFT and TPOT?OpenAI · Anthropic · Google DeepMind
- Why does p99 TTFT shoot up once utilization passes about 70%? Model it as a queue, then say where that model is wrong for a batching engine.OpenAI · Anthropic · Perplexity
- Two tenants share a replica. Should fairness count requests or tokens, and how does the scheduler enforce it step by step?Together AI · Fireworks AI · Baseten
- The scheduler admits a 100K-token prompt while 64 users are decoding. Walk through what happens to their inter-token latency, step by step.Anthropic · Google DeepMind · Perplexity
- max_num_seqs or max_num_batched_tokens: which one binds, when, and how would you set max_num_seqs from first principles?Red Hat · Anyscale · Baseten
- Design a scheduler that admits a request only if it predicts the TTFT and TPOT targets will both hold. What does it predict, and how?OpenAI · Anthropic · Google DeepMind
- vLLM turns on async scheduling by default. What does it overlap, what does it need from the model runner, and when must it be off?Red Hat · NVIDIA · Anyscale
- Design a batch inference service for 10 billion tokens a day on a 70B model, optimizing for cost rather than latency.Anyscale · Databricks · Snowflake
- Speculative decoding inside a continuous batch: how does the scheduler budget draft tokens, and what happens to the KV on rejection?Google DeepMind · Meta · NVIDIA
- Traffic triples for a minute on your LLM API. What do you reject, at what point, and what do you tell the client?OpenAI · Anthropic · Perplexity
FlashAttention & Attention Kernel Interview Questions
Online softmax and its proof, FlashAttention 1 to 3 and what each generation changed, split-KV decoding, paged attention kernels, MLA absorption, sliding windows, and why IO, not FLOPs, is the cost model.
- Derive online softmax, prove it is exact, and explain why it makes FlashAttention fast without reducing FLOPs.NVIDIA · OpenAI · Anthropic
- Write the causal mask for a decode step with a KV cache, where the query is shorter than the keys. Which corner does it align to?NVIDIA · Meta · Together AI
- Implement online softmax for a whole tile of queries, not one query, and write the rescaling step in matrix form.NVIDIA · OpenAI · Anthropic
- Count the HBM bytes FlashAttention reads and writes for one head at N = 8K and d = 128, and compare it with naive attention.NVIDIA · Anthropic · Meta
- Why does FlashAttention-2's backward pass recompute attention, and what exactly does the forward pass save for it?OpenAI · Anthropic · Google DeepMind
- Split-KV decoding: how many splits would you use for batch 1 at 128K context on 132 SMs, and how are the partial results merged?NVIDIA · Fireworks AI · Together AI
- How does a paged attention kernel gather keys through a block table, and what does that cost compared with contiguous KV?Anyscale · Red Hat · Fireworks AI
- GQA in the attention kernel: how do you load each KV head once for its whole group of query heads?Meta · Mistral AI · NVIDIA
- FlashAttention-3 on Hopper: explain warp specialization, TMA and ping-pong scheduling, and which bottleneck each one removes.NVIDIA · OpenAI · Anthropic
- FP8 attention: where does the precision break, and how do incoherent processing and per-block scales help?NVIDIA · AMD · Together AI
- Sliding-window attention in a tiled kernel: which blocks get skipped, and what speedup follows at 32K context with a 4K window?Mistral AI · OpenAI · Google DeepMind
- Should RoPE run inside the attention kernel or as a separate pass? Count the bytes saved, then find the precision trap at long positions.Meta · NVIDIA · Fireworks AI
- Variable-length batches without padding: what is cu_seqlens, how does the kernel index it, and what is the bug that mixes sequences?Together AI · Fireworks AI · Cohere
- Your attention kernel is exact in FP32 and wrong in BF16 at long context, but fine at short context. Find the accumulation bug.NVIDIA · OpenAI · Anthropic
- Design attention for a 1M-token context on one 8-GPU node: the kernel, the parallelism across GPUs, and the memory plan.Google DeepMind · Anthropic · Meta
- A decode-attention kernel wins at 128K context. What evidence would justify enabling it for other shapes?NVIDIA · AMD · OpenAI
GPU Performance Interview Questions for Inference
The memory hierarchy in numbers, coalescing and bank conflicts, occupancy, Hopper TMA and WGMMA, Blackwell tensor memory, decode GEMV versus prefill GEMM, fused MoE and dequant kernels, CUDA graphs, and reading Nsight like an interviewer.
- Your decode GEMV kernel reaches only about half of HBM bandwidth on an H100. Find the reasons and get it above 80%.NVIDIA · OpenAI · Anthropic
- In a decode GEMV over a row-major weight matrix, which thread-to-element mapping coalesces, and why does the transposed one not?NVIDIA · OpenAI · Meta
- A kernel uses 128 registers per thread and 48 KB of shared memory per block. How many warps are active per SM on an H100?NVIDIA · AMD · OpenAI
- Your shared-memory transpose tile has bank conflicts. Do the bank arithmetic, apply the padding fix, and say when padding is the wrong fix.NVIDIA · AMD · Meta
- Why do tensor core GEMMs use tiles like 128 by 128 by 64, and what limits the tile size on Hopper?NVIDIA · AMD · OpenAI
- A GEMM launches 140 output tiles on a 132-SM H100. What does that cost, and how do you fix it?NVIDIA · OpenAI · Anthropic
- Write a fused RMSNorm kernel in Triton and count its memory traffic against the unfused PyTorch version.OpenAI · Meta · Together AI
- Fused MoE kernels sort tokens by expert before the GEMM. Why, and what does a grouped GEMM do that a loop over experts cannot?NVIDIA · Together AI · Fireworks AI
- In a W4A16 GEMM, where does the INT4 unpacking happen, and how does the kernel keep that work hidden under the memory loads?NVIDIA · Red Hat · Together AI
- Here is an Nsight Systems timeline of one decode step. Find the launch gaps, synchronizations and CPU stalls, and price each one.NVIDIA · OpenAI · Anthropic
- Hopper replaced Ampere's cp.async and mma.sync with TMA and WGMMA. What changed in each, and how does it change the way you design a kernel?NVIDIA · OpenAI · Anthropic
- Blackwell added tensor memory and 2-CTA MMA. What problem does each one solve, and what changes in a GEMM kernel written for it?NVIDIA · OpenAI · Anthropic
- Could you keep the KV cache of hot sequences resident in the H100's 50 MB L2 during decode, and would it be worth it?NVIDIA · OpenAI · Anthropic
- You are about to report one number for your new kernel in a design review. How do you measure it so the number survives scrutiny?NVIDIA · AMD · OpenAI
- A kernel runs 20% slower on an H200 than on an H100, even though the H200 has 43% more memory bandwidth. How can that happen?NVIDIA · OpenAI · Together AI
LLM Quantization Interview Questions
FP8 E4M3 and E5M2 scaling, INT4 weight-only kernels, GPTQ, AWQ, SmoothQuant, EXL2 and GGUF k-quants, MXFP4 and NVFP4 block scales, KV quantization, accuracy evaluation beyond perplexity, and batch-invariant determinism.
- AWQ INT4 beat FP8 at batch 1, lost to it at batch 256, and cost 4 points on a coding eval. Explain both and pick a format.NVIDIA · OpenAI · Meta
- FP8 comes in E4M3 and E5M2. Give the range and precision of each, and say which tensors use which in inference.NVIDIA · AMD · OpenAI
- Per-tensor, per-channel, per-group, per-block: where does each quantization scale live, and what does it cost in bytes and in the kernel?NVIDIA · Meta · Red Hat
- GPTQ and AWQ both produce INT4 weights with group scales. What does each one optimize, and when does one beat the other?NVIDIA · Red Hat · Meta
- Why do activation outliers break INT8 W8A8 on LLMs, and how does SmoothQuant fix it by moving scale into the weights?NVIDIA · Meta · Microsoft (Microsoft AI)
- NVFP4 versus MXFP4: block size, scale format and bits per element, and why does the smaller block with a finer scale help accuracy?NVIDIA · AMD · OpenAI
- You quantized a model and perplexity barely moved. How do you evaluate it before it serves traffic, and why is perplexity not enough?OpenAI · Anthropic · Meta
- Which layers should stay in higher precision after quantization, and how do you find them with a sensitivity sweep?NVIDIA · Meta · Qualcomm
- In a block-scaled FP8 GEMM like DeepSeek V3's, where are the scales applied, and why must accumulation be promoted to FP32?NVIDIA · AMD · OpenAI
- GGUF k-quants and EXL2 mix bit-widths inside one model. What do they trade, and where do they fit next to datacenter GPU formats?Meta · Qualcomm · Mistral AI
- The same prompt gives different logits at batch 1 and batch 64 on an FP8 server. Why, and how can the kernels prevent it?OpenAI · Anthropic · Google DeepMind
- Fit a 70B model on one 80 GB H100 with room left for KV cache. Choose the weight and KV formats and show the arithmetic.NVIDIA · Meta · Baseten
- Your serving team does not own pretraining. When is quantization-aware training worth it over post-training quantization?Google DeepMind · Meta · NVIDIA
- A quantized MoE holds its benchmark scores but loses accuracy on rare domains, traced to rarely routed experts. Explain why and fix it.Mistral AI · Together AI · Fireworks AI
- Design the numerics policy for a fleet that spans Hopper and Blackwell: formats per generation, fallbacks, and how you validate them.NVIDIA · OpenAI · Anthropic
Speculative Decoding Interview Questions
The draft and verify algebra with its losslessness proof, acceptance rate versus draft length, Medusa, EAGLE and multi-token prediction heads, tree verification, why speculation loses at high batch, sampling kernels, and grammar-constrained output.
- Derive the expected speedup of speculative decoding, prove it does not change the output distribution, and tell me when it stops paying.OpenAI · Anthropic · Google DeepMind
- What do temperature, top-k and top-p each do to the next-token distribution, and in what order does a serving engine apply them?OpenAI · Anthropic · Google DeepMind
- Implement speculative-decoding verification for temperature zero. How do you handle the bonus token and the rejected drafts' KV?OpenAI · Anthropic · NVIDIA
- Medusa heads or a separate draft model for a 70B target? Compare acceptance, cost per step and the training each one needs.OpenAI · Meta · Together AI
- Build the tree attention mask that verifies a draft tree in one pass. How many tokens does each step verify, and at what cost?NVIDIA · Together AI · Fireworks AI
- EAGLE drafts from the target model's hidden states. Why does that raise acceptance, and what did EAGLE-3 change?OpenAI · Anthropic · Meta
- How are DeepSeek V3's multi-token prediction modules trained, and how does the engine turn that one MTP module into a speculative drafter?NVIDIA · Together AI · Fireworks AI
- When does n-gram or prompt-lookup speculation work well with no draft model at all? Walk me through a worked acceptance example.Anthropic · Snowflake · Databricks
- You turned on speculative decoding in production. What do you measure to prove it helps, and what is the right baseline to compare against?OpenAI · Anthropic · Fireworks AI
- How does grammar-constrained decoding build the allowed-token mask at every step, and what does that cost per step at batch 128?OpenAI · Anthropic · Databricks
- Serve JSON-schema outputs at high throughput: what do you precompile, what do you cache, and what is the tokenizer boundary problem?OpenAI · Anthropic · Databricks
- Implement beam search on a paged KV cache. How do forking and pruning work on block tables, and what does beam width 4 cost in memory?Google DeepMind · Meta · NVIDIA
- Speculative decoding with a quantized target or a quantized draft: does losslessness still hold, and against which distribution?NVIDIA · AMD · Anthropic
- Choose the speculative draft length per request at runtime. What signal, what policy, and how much does it beat a fixed gamma?OpenAI · Anthropic · Google DeepMind
- Design speculative decoding for a 70B fleet whose batch size swings from 1 at night to 256 at peak. What runs at each load level?OpenAI · Anthropic · Google DeepMind
Distributed & Disaggregated Inference Interview Questions
Tensor, pipeline, expert and data-parallel attention, the all-reduce bytes per layer, MoE dispatch and combine all-to-all, prefill and decode disaggregation, KV transfer over NVLink and RDMA, and rack-scale NVLink domains.
- Going from TP=4 to TP=8 on one H100 node cut a 70B model's latency by less than half and lowered tokens per GPU. One TP=8 or two TP=4?OpenAI · Anthropic · Meta
- In Megatron tensor parallelism, where do the column and row splits go in attention and the MLP, and why two all-reduces per layer?NVIDIA · Meta · OpenAI
- Pipeline parallelism for inference: why does it raise throughput but never lower per-token latency? Work the bubble arithmetic.Meta · NVIDIA · Together AI
- Size the expert-parallel all-to-all for a 256-expert, top-8 MoE at batch 256. How many bytes per token, and how long on the network?NVIDIA · Google DeepMind · Fireworks AI
- Why do MLA models like DeepSeek V3 run data-parallel attention with expert-parallel MoE instead of plain tensor parallelism?NVIDIA · Fireworks AI · Together AI
- You split prefill and decode onto separate nodes. Compute the KV transfer per request and the network bandwidth it needs.NVIDIA · Perplexity · Fireworks AI
- Why does a one-shot custom all-reduce beat NCCL for small decode messages on an NVLink node, and when does it lose?NVIDIA · AMD · Fireworks AI
- Serve Llama 3.1 405B: TP=8 on one H100 node, or TP=4 times PP=2 across two nodes? Work the latency and the memory.Meta · NVIDIA · Together AI
- Some experts get far more tokens than others. What do hot experts do to step time, and how do redundant experts and rebalancing fix it?NVIDIA · Google DeepMind · Mistral AI
- A GB200 NVL72 puts 72 GPUs in one NVLink domain. What does that change for MoE serving and for prefill-decode disaggregation?NVIDIA · Microsoft (Microsoft AI) · OpenAI
- A 128k-token prefill takes 10 seconds on one node. How does context parallelism split it across nodes, and what does it send?Google DeepMind · Meta · NVIDIA
- A multi-node MoE deployment's TPOT doubled overnight, and one network link is degraded. How do you find it, and how do you design around it?NVIDIA · OpenAI · Meta
- Each decode layer needs the previous all-to-all to finish. How do you overlap communication with compute anyway, and at what cost?NVIDIA · Fireworks AI · Together AI
- You need to get 500 GB of weights onto 1,000 GPUs as fast as possible. What is the lower bound, and what design gets close to it?OpenAI · Anthropic · Meta
- Design the parallelism plan for a trillion-parameter MoE when your fleet is half H100 nodes and half B200 nodes.OpenAI · Anthropic · Google DeepMind
vLLM, SGLang & TensorRT-LLM Interview Questions
vLLM V1, SGLang, TensorRT-LLM and llama.cpp from the inside: the engine core loop, schedulers, KV managers, model runners, CUDA graph capture, detokenisation and streaming, and the configuration knobs whose defaults you must be able to defend.
- Walk one request through vLLM V1, from the HTTP call to the streamed token. Which process owns each step, and why is it split that way?Anyscale · Databricks · Red Hat
- What do gpu_memory_utilization, max_model_len and max_num_seqs each control in vLLM, and how do they interact at startup?Anyscale · Red Hat · Databricks
- Why do serving engines capture a CUDA graph per batch size, and what does padding to the nearest captured size actually cost?NVIDIA · Fireworks AI · Together AI
- Walk through how vLLM's prefix caching identifies a reusable KV block and hands it to a new request, step by step.Anyscale · Red Hat · Databricks
- SGLang's radix cache versus vLLM's block hashing: compare the data structures, the eviction, and the workloads where each one wins.Databricks · xAI (SpaceXSI) · Together AI
- What does SGLang's overlap scheduler hide, how does it build the next batch before the current tokens exist, and what does it cost?xAI (SpaceXSI) · Databricks · Together AI
- TensorRT-LLM in 2026: what is the PyTorch backend, how does its in-flight batching admit requests, and when do you still pick it?NVIDIA · Perplexity · Baseten
- Why is streaming detokenization harder than decoding each token, and what breaks with multi-byte characters and stop strings?OpenAI · Anthropic · Fireworks AI
- How does an engine batch requests for different LoRA adapters in one forward pass, and what do Punica-style shrink and expand kernels do?Fireworks AI · Together AI · Databricks
- With JSON-schema decoding on, where in the engine step is the grammar mask computed and applied, and what does it cost?OpenAI · Anthropic · Fireworks AI
- You upgraded the serving engine with the same flags and throughput fell. Which defaults do you diff first, and how do you prove it?Anyscale · Databricks · Red Hat
- llama.cpp versus vLLM: what is each engine built for, and when does the CPU or Apple Silicon path actually win?Meta · Mistral AI · Qualcomm
- Write the main loop of a minimal LLM serving engine: request queue, scheduler, model step, sampling and streaming.OpenAI · Anthropic · Fireworks AI
- What changed in vLLM's Model Runner V2 compared with the V1 runner, how does the engine decide which one to use, and what falls back?Red Hat · Anyscale · Databricks
- You are serving a reasoning model whose answers run 30K output tokens. Which engine do you pick, and how do you configure it?OpenAI · Anthropic · Google DeepMind
- More CUDA graph captures made decode faster but increased preemption. Diagnose the regression and choose a policy.NVIDIA · Fireworks AI · Together AI
LLM Serving System Design Interview Questions
Design an LLM API on a fixed GPU pool: prefix and KV-aware routing, autoscaling on the right signal, multi-LoRA and multi-model serving, tenant isolation, cold starts and weight distribution, and cost per million tokens under an SLO.
- Size the fleet for a 70B chat API at 200 req/s, 2K tokens in and 300 out, p99 TTFT under 1 s, TPOT under 50 ms. What is the token cost?Anthropic · OpenAI · Fireworks AI
- Design the request path for an LLM API: gateway, router, engines, streaming. Where is each SLO measured, and what does each hop cost?OpenAI · Anthropic · Together AI
- Design prefix-aware routing that keeps cache hits high without overloading the replica that owns a hot prefix. What is your fallback rule?Anthropic · OpenAI · Fireworks AI
- Design autoscaling for an LLM fleet whose model takes four minutes to load. What do you scale on, and how much headroom do you hold?Baseten · Modal · Together AI
- Design a multi-tenant LLM platform with per-tenant token quotas. How do you stop one tenant from degrading everyone else's latency?OpenAI · Anthropic · Fireworks AI
- Design serving for a chat product and a batch API on the same GPU pool. How much batch work fits, and how do you protect chat latency?OpenAI · Anthropic · Google
- Design serving for 500 customer LoRA fine-tunes of one base model. Where do the adapters live, and how does a request find one?Fireworks AI · Together AI · Modal
- Design a multi-region deployment for a 70B model that survives the loss of a whole region. How much spare capacity do you buy, and where?OpenAI · Anthropic · Google
- Design a platform serving 20 models of different sizes on one shared fleet. How do you place them, and when does a model get its own GPUs?Together AI · Fireworks AI · Baseten
- Price an LLM API: go from a GPU-hour cost to a per-million-token price for input and for output. Why do they differ?OpenAI · Anthropic · Together AI
- Design the rollout of a new model version so that neither quality nor latency regresses in production. How do you know when to promote?OpenAI · Anthropic · Google DeepMind
- Design serving for a coding assistant: 32K-token prompts, 200-token completions, p99 TTFT under one second. What do you build?Anthropic · OpenAI · Fireworks AI
- Design real-time voice inference with 300 ms from the end of the user's turn to the first audio. Where does every millisecond go?OpenAI · Google DeepMind · Meta
- Plan capacity for a 70B fleet that mixes H100 and B200 nodes under one latency SLO. How do you count capacity and route across them?OpenAI · Anthropic · Meta
- Design batch inference over 1 billion documents: embed each one and run a 70B model over each. How many GPUs, how long, what cost?OpenAI · Databricks · Snowflake
- Cut inference cost per million output tokens by 70% while preserving quality and latency. Can you defend the target?OpenAI · Anthropic · Together AI
Inference Performance Debugging Interview Questions
p99 TTFT regressions with flat utilisation, OOMs under continuous batching, numerical divergence after an engine upgrade, nondeterministic outputs, slow replicas, and benchmark methodology that does not lie about warm-up, clocks or prompt mix.
- After a prompt-template deploy, p99 TTFT went from about 100 ms to over a second while GPU utilization did not move. Debug it.Anthropic · OpenAI · Perplexity
- TTFT is fine but inter-token latency is bad. Decompose the decode step and rank the causes.Fireworks AI · Together AI · Baseten
- vLLM fails to start with an out-of-memory error on a model that should fit. Walk through the startup memory profile.Red Hat · Anyscale · Baseten
- At peak traffic the engine preempts requests constantly. Diagnose it from the metrics and fix it.Fireworks AI · Together AI · Anyscale
- Throughput fell 15% after you enabled FP8. Find out whether it is the kernel, the scales or the config.NVIDIA · Fireworks AI · Baseten
- Two identical requests at temperature zero return different outputs. Explain why, and decide what you will promise customers.OpenAI · Anthropic · Together AI
- One replica in eight is 30% slower than the rest. Find it from the metrics and explain the likely hardware causes.NVIDIA · Together AI · Baseten
- A quantized model passes its perplexity checks but fails long-context retrieval. Debug it.Anthropic · Google DeepMind · Cohere
- The benchmark showed a 2x speedup, and production saw 10%. Reconcile the two numbers before anyone rolls it back.Fireworks AI · Together AI · Baseten
- p99 TPOT spikes every few seconds in a regular pattern while p50 stays flat. Find the cause.OpenAI · Anthropic · Perplexity
- A two-node deployment loads weights, reports healthy, then hangs on the first request. Debug NCCL and the engine.NVIDIA · Meta · Together AI
- After turning on speculative decoding, p50 latency improved and p99 got worse. Explain it, and decide what to ship.OpenAI · Anthropic · Google DeepMind
- An 8B model on one H100 stops scaling at high request rates and the GPU is not saturated. Prove the CPU is the bottleneck, then fix it.Fireworks AI · Together AI · Baseten
- Memory on an inference server grows slowly all day until the engine OOMs and restarts. Find the leak.OpenAI · Anthropic · Databricks
- Build the dashboard you want open before your next inference incident, and say what decision each panel drives.OpenAI · Anthropic · Databricks
- Structured output slowed generation. Is the cause grammar compilation, mask construction, or host scheduling?NVIDIA · Fireworks AI · Together AI
- MoE decode slowed after the traffic mix changed. Are hot experts or communication causing the delay?NVIDIA · DeepSeek · Together AI
- Prefill/decode disaggregation increased latency. Is KV transfer, admission, or token delivery on the critical path?NVIDIA · Fireworks AI · Together AI
- Image requests OOM while the KV pool is mostly free. Separate encoder peaks from decoder memory and choose a fix.OpenAI · Anthropic · NVIDIA
- A router raises prefix-cache hits but worsens tail latency. Replay its decisions and find the missing state.NVIDIA · Fireworks AI · Together AI
Inference Coding Interview Questions
The coding round of an inference loop: grouped-query attention with a cache, online-softmax attention, a paged block allocator, a continuous batching scheduler, a lossless speculative verifier, top-p sampling, RoPE, INT4 packing and Triton kernels.
- Implement a paged KV-cache block manager with copy-on-write forking, then a scheduler step that preempts when blocks run out.Anthropic · OpenAI · NVIDIA
- Implement grouped-query attention over a preallocated KV cache, with no torch.cat, for both prefill chunks and decode.OpenAI · Anthropic · Meta
- Implement batched top-p sampling with temperature. It must handle ties at the cutoff, p = 1, temperature 0 and masked tokens.OpenAI · Anthropic · Cohere
- Implement a continuous batching scheduler with a token budget and chunked prefill, so decodes never wait behind a long prompt.Anthropic · OpenAI · Databricks
- Implement the lossless speculative sampling verifier for a batch with ragged draft lengths, and prove it matches the target.Google DeepMind · OpenAI · Anthropic
- Implement a radix tree prefix cache: match, insert with edge splitting, request locks and LRU eviction of unlocked leaves.Anthropic · OpenAI · Perplexity
- Implement RoPE for a batched decode step where every row sits at a different position, and verify it against the full-sequence version.Meta · Mistral AI · NVIDIA
- Implement INT4 packing with group scales and a dequantizing matmul, and test it against FP32 with odd K and a partial group.NVIDIA · AMD · Qualcomm
- Write a row-wise softmax kernel in Triton that stays correct when a row is longer than the block size, and test it.NVIDIA · OpenAI · AMD
- Implement a ring all-reduce as a simulation over p ranks, verify it against a direct sum, and count the bytes each rank sends.NVIDIA · Google DeepMind · Meta
- Implement beam search over a paged KV cache, forking beams by sharing blocks and copying only the tail on write.NVIDIA · Meta · Google DeepMind
- Implement a per-tenant rate limiter for an LLM API that limits tokens, not requests, using a token bucket.OpenAI · Anthropic · Together AI
- Implement a KV block allocator that shares prefix blocks across requests by chained hashes, with LRU eviction of cached free blocks.Anthropic · OpenAI · Perplexity
- Implement a minimal autoregressive generate loop that streams text and honors stop sequences, including ones split across tokens.OpenAI · Anthropic · Perplexity
- Implement split-KV decoding attention and the merge of the partial results, and prove it matches attention over the whole sequence.NVIDIA · OpenAI · Anthropic
- This KV allocator passes ordinary append tests but corrupts forked requests. Find two bugs and prove the repair.Anthropic · OpenAI · NVIDIA
- Implement cancellation for an inference request without reusing KV pages while GPU work still references them.OpenAI · Anthropic · NVIDIA
- A continuous batch changes order while a forward pass runs. How do you return each token to its owner?NVIDIA · OpenAI · Anthropic
- A speculative verifier returns correct tokens, but the next step diverges. Debug its cache commit boundary.OpenAI · Anthropic · NVIDIA
- Logits are identical, but a seeded request changes output after batch reordering. Repair the sampler's RNG contract.OpenAI · NVIDIA · Together AI
Inference interview questions by company
Each company weights the rounds differently. These pages say what public evidence supports about each loop, with a confidence level, and say so when the evidence is thin.
Before the interview: the role itself
Inference engineer interviews: FAQ
What is asked in an inference engineer interview?
Implement-from-scratch coding of serving components (attention with a KV cache, a paged block allocator, a scheduler, a speculative verifier), inference math such as KV bytes per token and decode latency floors, serving system design sized from a traffic profile, kernel and GPU performance deep dives, and a debugging case built from production metrics.
How is an inference interview different from an ML engineer interview?
An ML engineer loop tests modelling: data, objectives and evaluation. An inference loop starts after the model is trained and tests how it is served: memory, batching and scheduling, kernels, number formats, parallelism and cost per token, usually with arithmetic expected in every round.
What math do inference engineer interviews expect?
Back-of-envelope numbers produced quickly and with units: KV cache bytes per token for a given config, the decode latency floor from weight bytes and memory bandwidth, the batch size where decode turns compute-bound, expected tokens per step of speculative decoding, and GPU counts and cost for a traffic profile.
Do I need to know CUDA for an inference engineer interview?
For kernel and performance roles, yes: expect CUDA or Triton, the memory hierarchy, coalescing, occupancy and profiler evidence. For engine and fleet roles, expect to reason about kernels and read profiles rather than write them from scratch.
How should I prepare for an LLM inference interview?
Start with the cost model (the roofline, prefill versus decode), then KV memory and scheduling, then one serving engine read from its current source, then numerics and kernels, then parallelism and fleet design. Practise each answer out loud with the arithmetic, and check it against the scoring lines.
Every inference interview question, worked to the number
196 questions across 12 topics, ordered easy to expert, each with the derivation, the follow-ups interviewers hold back and the scoring lines. The first questions in every topic are free.
Every worked answer and derivation in all 12 topics, every concept page, and the full practice-test bank.
6 months · One payment · No auto-renewal
A serving team at a lab and a kernel team at a chip vendor run different loops, and both change between hiring cycles. Ask your recruiter which rounds your loop includes before you plan around this list.