sgl-project/sglang · error · NotImplementedError

topk kernels only support streaming implementation: {_impl=}

Error message

topk kernels only support streaming implementation: {_impl=}

What it means

NotImplementedError from gate_topk: only _impl == "streaming" has a code path. Any other value for the internal _impl selector falls into the else branch and raises, because no alternative kernel implementation is compiled in this build.

Source

Thrown at python/sglang/kernels/ops/moe/gate_topk.py:163

        BLOCK_SIZE_N = 32
        BLOCK_SIZE_M = 32
        grid = (triton.cdiv(n_rows, BLOCK_SIZE_M),)
        _streaming_topk_kernel[grid](
            x_ptr=x,
            stride_xm=x.stride(0),
            values_ptr=values,
            indices_ptr=indices,
            M=n_rows,
            N=n_cols,
            N_PAD=triton.cdiv(n_cols, BLOCK_SIZE_N) * BLOCK_SIZE_N,
            K=k,
            K_POW2=triton.next_power_of_2(k),
            BLOCK_SIZE_M=BLOCK_SIZE_M,
            BLOCK_SIZE_N=BLOCK_SIZE_N,
            RETURN_VALUES=return_values,
        )
    else:
        raise NotImplementedError(
            f"topk kernels only support streaming implementation: {_impl=}"
        )

    if return_values:
        return values, indices
    return indices

View on GitHub (pinned to 0132848349)

Solutions

  1. Pass _impl="streaming" or rely on the default
  2. If you intended a new implementation, implement the branch in gate_topk before enabling the flag
  3. Remove the impl parameter from the call if the streaming path is fine

Example fix

// before
ids = gate_topk(x, k=8, _impl="naive")
// after
ids = gate_topk(x, k=8)  # defaults to the streaming implementation
Defensive patterns

Strategy: validation

Validate before calling

assert _impl in ("streaming", None)

Prevention

When it happens

Trigger: Calling gate_topk with _impl set to something other than "streaming" (e.g. "naive", "sorted", or an experimental name) from the routing forward.

Common situations: Adding a new topk implementation flag without wiring a kernel; copy-pasted code passing an outdated impl name; feature-flag plumbing that defaults to an unimplemented variant.

Related errors


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