Tutorial: como construir seu próprio runtime de LLM do zero com CUDA e H100

Como construir seu próprio runtime de LLM do zero: um tutorial completo com CUDA e H100

Se você já quis entender exatamente o que acontece dentro de um motor de inferência de LLM — empacotar seus próprios pesos, gerenciar cada barreira de sincronização e capturar seus próprios grafos CUDA — este tutorial é o roteiro definitivo. O engenheiro Anubhab Banerjee documentou uma jornada completa de construção de um runtime de inferência do zero para o Qwen2.5-Coder-7B rodando em uma NVIDIA H100.

O resultado é o annotated-llm-runtime, um runtime com cerca de duas dúzias de arquivos CUDA/C++ onde cada caminho quente está comentado para explicar o porquê, não apenas o o quê. O projeto está disponível no GitHub e serve como referência de aprendizado — não como substituto do llama.cpp, que continua sendo a escolha certa para produção.

Números que impressionam

• Decodificação em estado estacionário: ~16,7 ms/token (~60 tok/s) em contexto de até 4096 tokens
• Latência do primeiro token: ~128 ms para um prompt de 512 tokens
• Ganho com CUDA graphs: 7× de melhoria — a decodificação caiu de ~119 ms/token para ~17 ms/token usando grafos CUDA capturados, sem mudar um único kernel

As três lições que só o hardware ensina

O tutorial é pontuado por três bugs que ensinaram mais do que qualquer documentação:

1. __syncthreads() dentro de branch especializado por warp: comportamento indefinido que corrompe o softmax nas bordas das páginas KV. O tipo de bug que não aparece em teste unitário, só em carga real com contexto longo.
2. Grafos CUDA não são opcionais: a diferença entre 119 ms/token e 17 ms/token está inteiramente no overhead do driver — cada lançamento de kernel individual tem um custo que os grafos CUDA eliminam ao submeter todo o grafo de uma vez.
3. prmt.b32 venceu __dp4a no Hopper: para as formas GEMV de gate/up/down, a instrução de permutação de bytes foi mais rápida que o produto escalar de inteiros. O caminho com ativações INT8 foi implementado, testado e removido.

Para quem é este tutorial

Não é um tutorial para iniciantes. Você precisa de familiaridade com C++, CUDA e conceitos de inferência de LLMs (atenção, projeções lineares, softmax, sampling). Mas se você está nesse ponto da curva de aprendizado, o material é ouro: cada kernel é explicado com o contexto de hardware que motivou as decisões de design.

O autor é claro: o llama.cpp continua sendo a referência. Na mesma configuração (H100, Q4_K_M, pp512/tg128), ele entrega ~43 ms de TTFT e ~4,95 ms/token. O runtime didático não compete — ele explica.