Inference optimisation · serving · edge · GPU sim

The model is right. Now make it fast, small and cheap enough to ship.

You've found the LLM or VLM that answers correctly. Getting it to the tokens per second you need, into the memory you have, or onto the device in the field is a different job. That's ours: an agentic optimizer that searches the stack from the serving engine down and verifies at the SASS, a server that runs the result, and the engineers behind both.

Book an inference reviewTake the walkthrough

Open-weight LLMs and VLMs · H100, B200, Jetson, Apple, Qualcomm · accuracy floor enforced · runs in your environment

Optimize · the runs list · a run ships to production as the clip plays01 / 02
3.3×throughput on the sample run · 1,840 → 6,120 tok/s
−55%time to first token at p99 · 612 → 275 ms, under the 400 ms SLO
84.1%accuracy held above the 83.5 floor · int4 rejected by the eval
−69%cost per million tokens · $1.02 → $0.31

What one Optimize run produced for Qwen3-32B on two H100s against a support-QA eval, in 2 h 14 m of search.

Where the work happens

Summarise this contract.
Supplier may raise prices yearly, capped at 4%…

Application

Capture the real traffic shape, the eval set and the SLO we optimise to.

VitalOps works here
vLLMSGLangTRT-LLM
fp8 kv cache
spec decode
prefix cache
tok/s1,8403,1205,4606,120
bf16fp8TP 22×H100

01Serving engine

Configure the engine: quantised weights and KV cache, TP/PP, scheduler flags.

02Scheduling

Tune how requests share the GPU: batching, chunked prefill, prefix caching.

03Compile and fuse

Fuse ops, remove copies, capture CUDA graphs, compile the decode step.

04Kernels

Profile, then rewrite the hot kernels for your shapes; verify vs the stock path.

rmsnorm.ptxrmsnorm.sasssteered
1
ld.global.u16 r4,[rd2]LDG.E.U16LDG.E.12831%22%
2
bar.sync 0BAR.SYNCSHFL.BFLY11%7.5%
3
st.global.b32 [rd6],r20STG.E.32STG.E.6414%8.5%

05SASS verification

Inspect the SASS ptxas emits: registers, spills, stalls. Steer the compiler.

Silicon

Measure every candidate on the target GPU and report what it did.

Most teams stop at the first layer. We work all five and verify at the bottom, in the SASS the GPU actually executes.

Walkthrough

01 · Optimize

Point it at the model, the hardware and the number you need

Give the optimizer a model, a GPU target, an SLO and your eval set. The agent searches the serving engine's knobs, quantisation, parallelism, scheduling, speculative decoding and custom kernels, ranks candidates in the built-in simulator, verifies the best on real GPUs and ships a reproducible bundle. Every candidate is held to your accuracy floor; kernels are checked at PTX and SASS level and numerically verified.

What changes

  • Weeks of trial and error become one run with a report.
  • Every speed-up is tied to an accuracy number.
  • The result is a reproducible bundle anyone on the team can rerun.

Works with

vLLM, SGLang, TensorRT-LLM, llama.cpp, Hugging Face Transformers; lm-eval or your own evals. Targets: tokens/s, TTFT, TPOT, p99, memory, cost per million tokens.

02 · Optimize · Kernels

Rank kernels without a GPU attached

Optimize's kernel page runs your candidates on real GPUs when you have them. When you don't, it runs them on Sim: an optional GPU simulator. A replacement libcuda.so takes the same binary, loads the cubin into a SASS interpreter and executes the launches on CPU threads under a timing model, so you get a predicted time per target, a per-instruction stall breakdown and a ranking of candidates on a laptop or in CI. Three fidelities: static in milliseconds, sampled in seconds, cycle-level in hours. Rank correlation against real hardware is reported on every run, and the best candidates are verified on a real GPU before anything ships.

What it unlocks

  • Kernel work no longer waits for a GPU slot.
  • Agents rank thousands of variants for the cost of CPU time.
  • A performance regression fails CI before it reaches hardware.

Works with

CUDA C++, Triton, CUTLASS, PyTorch through a driver shim; KernelBench and TritonBench; GitHub Actions. Targets: H100, H200, B200, L40S, RTX 4090 and 5090, Jetson Orin and Thor.

03 · Serve

Ship the bundle as a canary, promote when the numbers hold

An OpenAI-compatible server built around the bundles Optimize produces: multi-model pools, continuous batching, prefix caching, chunked prefill, speculative decoding, structured output, and canary rollouts with instant rollback. Runs in your cloud account or data centre; air-gapped supported.

What you get

  • The optimised configuration is the one actually in production.
  • Rollouts you can reverse in seconds.
  • One endpoint for every model you run.

Works with

OpenAI SDKs, LangChain and LlamaIndex, Kubernetes, NVIDIA and AMD GPUs. /v1/chat/completions, /v1/embeddings, tool calls, JSON-schema and grammar-constrained decoding.

04 · Serve · Monitoring

Watch it in production, in the numbers that decide your bill

Generic APM shows CPU and memory. Serve's monitoring shows TTFT, TPOT, tokens per second, KV-cache hit rate, queue depth, cost per million tokens, and whether model quality is drifting, per model, pool, tenant and prompt template. When something moves, the fix is handed back to Optimize. It also attaches to an existing vLLM, SGLang or TensorRT-LLM deployment.

What you catch

  • You see the cost of a prompt change the hour it ships.
  • Evals catch quality regressions before they reach users.
  • Every alert arrives with a proposed fix.

Works with

Prometheus, OpenTelemetry, Grafana, PagerDuty and Slack; existing vLLM, SGLang and TensorRT-LLM deployments.

05 · Edge

Fit it on the device, and train it there

Each device gets the largest model that meets your accuracy floor inside its memory and latency budget. A Jetson or a flagship phone takes a 3B vision-language model once the weights, KV cache and encoder are quantised and a couple of layers pruned; a Raspberry Pi gets a 0.5 to 1.5B model and the same accuracy guard. Where the data can't leave the device, fine-tuning runs on it: QLoRA inside the memory budget, a live loss curve, and the adapter pushed to every device on the same base.

What becomes possible

  • A model that didn't fit now runs on a phone, a Pi or a Jetson.
  • Accuracy is measured after every step.
  • Fine-tuning runs on the device, so the data never leaves it.

Works with

NVIDIA Jetson, iPhone and Apple silicon, Pixel and Android NPUs, Qualcomm Snapdragon and Dragonwing, Raspberry Pi 5, x86 CPUs. Exports to GGUF, MLX, ExecuTorch, CoreML, QNN and TensorRT.

06 · Fleet

Run N agents on M GPUs, on your premises

Agent workloads don't look like chat: long shared prefixes, bursts of parallel sub-agents, and long idle gaps while tools run. Fleet schedules a swarm of agents across your GPUs with KV-aware routing, prefix sharing between agents, and preemption while an agent waits on a tool. Scaling from 8 agents to 96 is one command, and the capacity check tells you where the SLO breaks.

What it buys you

  • More agents per GPU, because the KV cache stops being thrown away.
  • Scaling means changing one number.
  • A straight answer to how many agents you can run.

Works with

Claude Agent SDK, OpenAI Agents, LangGraph, or your own loop; Kubernetes and Slurm; air-gapped by default.

Services

Engineers on your problem, fixed scope

The products came out of this work. We still do it, fixed-scope, for teams who want the result delivered.

Inference optimisation sprint

Model, hardware and SLO in; verified configuration and report out. Two weeks, fixed fee.

vLLM · SGLang · TensorRT-LLM · FP8 · INT4 · NVFP4 · speculative decoding · prefix caching

Custom kernels

Fused and specialised kernels where the stock path leaves throughput on the table, tuned down to the SASS.

Triton · CUDA · CUTLASS · wgmma · TMA · cp.async · Hopper · Blackwell · Ada

Speculative decoding

Draft models trained on your traffic, so the speed-up holds on your prompts.

EAGLE-3 · Medusa · n-gram and prompt-lookup drafts · acceptance tuning

Quantisation and distillation

Smaller models that keep the accuracy you signed off on, with the eval to prove it.

AWQ · GPTQ · SmoothQuant · KV-cache quantisation · distillation · pruning

Edge deployment and on-device training

From a model that doesn't fit to one that runs, updates and fine-tunes in the field.

Jetson · Apple silicon · Qualcomm · Raspberry Pi · ExecuTorch · GGUF · MLX · QLoRA on device

Serving for agent workloads

Design and tuning of the serving layer for long prefixes, sub-agent bursts and tool-call gaps.

KV-aware routing · prefix sharing · capacity planning

Benchmarking and capacity planning

Tokens per second per dollar across GPUs and clouds, measured on your workload.

H100 · H200 · B200 · MI300X · L40S · consumer cards · buy-versus-rent models

Training throughput

When the bottleneck is the training run: data loading, parallelism, kernels and memory.

FSDP · DeepSpeed · Megatron · torch.compile · fused optimisers · MFU audits
How we work together

Start small, prove it on your hardware, then run it

  1. Start here

    Inference review Free · 30 minutes

    Bring the model, the hardware you have or plan to buy, and the number you need to hit. You leave with an honest estimate of what's achievable and how.

    Book the review
  2. Prove it

    Optimisation sprint Fixed fee · 2 weeks

    We run Optimize and our engineers on your model and hardware. You get the verified configuration, the kernels, the eval numbers and a report.

    If we don't beat your baseline, you don't pay.
  3. Run it

    Managed inference Monthly

    Serve with its monitoring, and continuous re-optimisation as your models and traffic change, in your environment, with an SLO we sign.

Stack

Everything we work with

Serving and compilers

Hardware

Techniques

Models

Clients

“VitalOps helped us optimise our data pipelines for the finance databases, which directly translated into a smoother and more intuitive experience for our users.”

ChatfinCEO, Chatfin · AI for finance teams

“The team is highly responsive, pragmatic, and product-minded, and the best part is that they understand what it takes to operationalize data, not just analyze it.”

StraightopsCEO, Straightops · Source-to-Pay automation
Questions

What engineering teams ask us

Do we have to send you our weights or data?

No. Optimize, Serve, Edge and Fleet run in your cloud account or data centre. Evals run on your data where it already lives. For sprints, our engineers work inside your environment.

Which models can you speed up?

Open-weight models and your own fine-tunes: Llama, Qwen, DeepSeek, Mistral, Gemma, GLM, Whisper, VLMs, embedding and reranker models. We can't make a closed API model faster, but we can route, cache and batch around it.

How do you make sure the faster model is still correct?

You set an accuracy floor on your eval set. Every candidate configuration is evaluated against it, with several seeds, and anything below the floor is rejected no matter how fast it is. Custom kernels are checked numerically against the reference implementation and inspected at SASS level.

What does "down to the SASS" actually mean?

Most optimisation stops at the serving engine's flags. When the remaining bottleneck is a kernel, we write or tune it in Triton or CUDA and then inspect what the compiler produced: register pressure, spills, occupancy, shared-memory bank conflicts and stall reasons in the SASS the GPU executes. We don't write SASS; NVIDIA's ptxas does. We steer it through PTX intrinsics, launch bounds and compiler flags, and check the result. Optimize's simulator lets us do that without a GPU in the loop.

Can you really train on the edge device?

For many devices, yes. QLoRA fine-tunes of 1–8B models fit within the memory of a Jetson Orin, an Apple M-series machine or a workstation GPU. Edge plans the job within the device's memory budget and tells you the time and the expected eval gain before it starts.

Do you support AMD or non-NVIDIA hardware?

AMD MI300-class GPUs through ROCm, with kernels written in Triton where possible. Edge covers Apple, Qualcomm and ARM targets. The kernel simulator currently models NVIDIA GPUs.

Get started

What's the number you need to hit?

Tokens per second, time to first token, a memory budget, a cost per million tokens. Bring it and the model. We'll tell you in 30 minutes whether it's reachable and what it would take.

Book an inference reviewhello@vitalops.ai