{"record":{"id":"1d48dfe5187b5a12","repo":"sgl-project/sglang","slug":"q-k-and-v-must-be-cuda-tensors","errorCode":null,"errorMessage":"q, k, and v must be CUDA tensors","messagePattern":"q, k, and v must be CUDA tensors","errorType":"validation","errorClass":"ValueError","httpStatus":null,"severity":"error","filePath":"python/sglang/kernels/ops/diffusion/layout/ulysses_qkv_triton.py","lineNumber":66,"sourceCode":"    )\n    output_base = head_slot * (3 * head_size) + dim\n    tl.store(output_ptr + output_base, q, mask=mask)\n    tl.store(output_ptr + output_base + head_size, k, mask=mask)\n    tl.store(output_ptr + output_base + 2 * head_size, v, mask=mask)\n\n\ndef pack_qkv_destination_major(\n    q: torch.Tensor,\n    k: torch.Tensor,\n    v: torch.Tensor,\n    world_size: int,\n    out: torch.Tensor | None = None,\n) -> torch.Tensor:\n    \"\"\"Pack matching ``[rows, global_heads, head_size]`` Q/K/V tensors.\"\"\"\n    if q.dim() != 3 or q.shape != k.shape or q.shape != v.shape:\n        raise ValueError(\"q, k, and v must have the same 3D shape\")\n    if not (q.is_cuda and k.is_cuda and v.is_cuda):\n        raise ValueError(\"q, k, and v must be CUDA tensors\")\n    if not (q.device == k.device == v.device and q.dtype == k.dtype == v.dtype):\n        raise ValueError(\"q, k, and v must have the same device and dtype\")\n    if q.stride(-1) != 1 or k.stride(-1) != 1 or v.stride(-1) != 1:\n        raise ValueError(\"q, k, and v must be contiguous in head_size\")\n    if world_size < 1 or q.shape[1] % world_size != 0:\n        raise ValueError(\"world_size must be positive and divide global_heads\")\n\n    rows, global_heads, head_size = q.shape\n    local_heads = global_heads // world_size\n    expected_shape = (world_size, rows, local_heads, 3 * head_size)\n    if out is not None:\n        if not (\n            out.shape == expected_shape\n            and out.is_contiguous()\n            and out.dtype == q.dtype\n            and out.device == q.device\n        ):\n            raise ValueError(","sourceCodeStart":48,"sourceCodeEnd":84,"githubUrl":"https://github.com/sgl-project/sglang/blob/0132848349585cfe6aae51c4941cbae872505f8a/python/sglang/kernels/ops/diffusion/layout/ulysses_qkv_triton.py#L48-L84","documentation":"pack_qkv_destination_major launches a Triton kernel and therefore requires q, k, and v to all be CUDA tensors. Supplying any CPU tensor (or a tensor on a non-CUDA device) fails this explicit check before kernel launch.","triggerScenarios":"Calling pack_qkv_destination_major with any of q, k, v on CPU — common in unit tests that build tensors with torch.zeros(...) without device='cuda', or when a tensor was moved to CPU for logging and not moved back.","commonSituations":"Writing tests without device='cuda'; mixed-device pipelines where activations were detached().cpu()'d for debugging; running on CPU-only environments where the kernel cannot work at all.","solutions":["Move all three tensors to the GPU: q, k, v = q.cuda(), k.cuda(), v.cuda()","In tests, always construct inputs with device='cuda'","Skip or mark tests xfail on CPU-only machines","Ensure the model runner places attention tensors on the correct device"],"exampleFix":"# before\nq = torch.randn(128, 32, 64)  # CPU\n# after\nq = torch.randn(128, 32, 64, device='cuda', dtype=torch.bfloat16)","handlingStrategy":"validation","validationCode":"assert q.is_cuda and k.is_cuda and v.is_cuda","typeGuard":"def all_cuda(*ts) -> bool:\n    return all(t.is_cuda for t in ts)","tryCatchPattern":null,"preventionTips":["Construct test tensors with device='cuda'","Move tensors back after any .cpu() debugging"],"tags":["cuda","device","ulysses","attention"],"backgroundTag":"tensor-on-wrong-device","analyzedSha":"0132848349585cfe6aae51c4941cbae872505f8a","analyzedAt":"2026-08-28T05:10:05.995Z","schemaVersion":2},"datasetVersion":"2026-08-28T06:17:29.519Z"}