sgl-project/sglang · error · ValueError

q, k, and v must have the same device and dtype

Error message

q, k, and v must have the same device and dtype

What it means

Before packing, the kernel requires q, k, v to live on the same CUDA device and share one dtype. Mixed devices (e.g. cuda:0 vs cuda:1) or mixed dtypes (fp16 q with bf16 k) raise this ValueError.

Source

Thrown at python/sglang/kernels/ops/diffusion/layout/ulysses_qkv_triton.py:68

    tl.store(output_ptr + output_base, q, mask=mask)
    tl.store(output_ptr + output_base + head_size, k, mask=mask)
    tl.store(output_ptr + output_base + 2 * head_size, v, mask=mask)


def pack_qkv_destination_major(
    q: torch.Tensor,
    k: torch.Tensor,
    v: torch.Tensor,
    world_size: int,
    out: torch.Tensor | None = None,
) -> torch.Tensor:
    """Pack matching ``[rows, global_heads, head_size]`` Q/K/V tensors."""
    if q.dim() != 3 or q.shape != k.shape or q.shape != v.shape:
        raise ValueError("q, k, and v must have the same 3D shape")
    if not (q.is_cuda and k.is_cuda and v.is_cuda):
        raise ValueError("q, k, and v must be CUDA tensors")
    if not (q.device == k.device == v.device and q.dtype == k.dtype == v.dtype):
        raise ValueError("q, k, and v must have the same device and dtype")
    if q.stride(-1) != 1 or k.stride(-1) != 1 or v.stride(-1) != 1:
        raise ValueError("q, k, and v must be contiguous in head_size")
    if world_size < 1 or q.shape[1] % world_size != 0:
        raise ValueError("world_size must be positive and divide global_heads")

    rows, global_heads, head_size = q.shape
    local_heads = global_heads // world_size
    expected_shape = (world_size, rows, local_heads, 3 * head_size)
    if out is not None:
        if not (
            out.shape == expected_shape
            and out.is_contiguous()
            and out.dtype == q.dtype
            and out.device == q.device
        ):
            raise ValueError(
                "out must be a contiguous tensor with the expected shape, "
                "device, and dtype"

View on GitHub (pinned to 0132848349)

Solutions

1. Unify dtype: cast k and v (or all three) to a single dtype with .to(q.dtype)
2. Verify q.device == k.device == v.device; move tensors with .to(q.device)
3. Check that model weights and activations were converted consistently (model.to(dtype))
4. In TP setups, ensure the rank's tensors are on its assigned device before packing

Example fix

# before
packed = pack_qkv_destination_major(q_bf16, k_fp16, v_bf16, ws)
# after
k = k.to(q.dtype)
packed = pack_qkv_destination_major(q, k, v, ws)

assert q.device == k.device == v.device and q.dtype == k.dtype == v.dtype
k = k.to(q.dtype); v = v.to(q.dtype)

def same_device_dtype(q, k, v) -> bool:
    return q.device == k.device == v.device and q.dtype == k.dtype == v.dtype

When it happens

Trigger: Passing tensors allocated on different GPUs in multi-GPU setups without proper device placement, or tensors cast to different precisions (e.g. q in bf16 but v still in fp16) during half-precision conversion of a model.

Common situations: Tensor-parallel or multi-GPU diffusion serving where per-rank tensors land on different devices; partially-applied dtype conversions (model.to(torch.bfloat16) missing some buffers); legacy fp16 checkpoints mixed with bf16 activations.

Related errors

• q, k, and v must be CUDA tensors
• Unsupported usp_merge_heads dtype: {dtype}
• scale_shift_table must be CUDA, bf16/fp32, last-dim contiguo
• QKV tensors must be CUDA bfloat16 tensors
• Reserved serve backend names cannot be used: {names}

AI-assisted analysis of sgl-project/sglang@0132848349 (2026-08-28). Data as JSON: /api/errors/40eac09a593ee127. Report an issue: GitHub.