Skip to content

Do not fork an OpenMP team to run one iteration - #806

Open
ZacharyZcR wants to merge 2 commits into
JustVugg:devfrom
ZacharyZcR:perf/omp-grain-guard
Open

Do not fork an OpenMP team to run one iteration#806
ZacharyZcR wants to merge 2 commits into
JustVugg:devfrom
ZacharyZcR:perf/omp-grain-guard

Conversation

@ZacharyZcR

@ZacharyZcR ZacharyZcR commented Aug 3, 2026

Copy link
Copy Markdown
Contributor

Draft.

The change

The DSA indexer's per-position top-k is #pragma omp parallel for over s < S. At decode S is 1 β€” the region has exactly one iteration, and OpenMP still forks and joins the whole team to run it. Once per layer, per token.

It happens whether or not the body does any work: both early continue branches are taken on short contexts, and the fork precedes them.

#pragma omp parallel for schedule(dynamic,1) if(S > 1)

Prefill (S > 1) is untouched. A one-iteration parallel region is overhead by construction β€” this needs no benchmark to justify.

But it came out of one, and that part is worth reading

Tiny oracle model (5 layers, hidden 128), 15 runs per configuration, median with quartiles:

OMP_NUM_THREADS tok/s p25 p75
1 17109 15573 18119
4 10749 10650 11125
24 1529 913 1935

The distributions do not overlap. One thread is 11Γ— faster than 24, and the spread widens with the team β€” Β±8% at one thread, Β±34% at twenty-four. That is the same tail the profiler reports as p90 sitting 27% above p50.

A small model on a 24-core host, so the ratio does not transfer to a 744B run. What transfers is the shape: colibri hands every parallel region the full team regardless of how much work it contains, and there is no notion of a minimum grain anywhere in the tree. PyTorch's at::parallel_for takes a grain_size for precisely this reason; vLLM's CPU backend carries its own thread-count controls (VLLM_CPU_OMP_THREADS_BIND, VLLM_CPU_NUM_OF_RESERVED_CPU).

What this PR deliberately does not do

Fix that properly. A per-region threshold is only honest if it is measured on the model people actually run, and this machine cannot do that: the tiny replay is 0.015 s, and repeated runs of an unchanged binary land anywhere from 356 to 3893 tok/s. That noise swamps any single-region effect β€” I tried, got a plausible-looking +135% median, and threw it away because the quartiles overlapped almost completely.

So this takes only the case that needs no threshold at all: a loop that cannot be parallel. The rest wants the lab machine and a real model.

The benchmark is included

make -C c bench-omp-grain is the tool those numbers come from, and it is in this PR. It runs a region whose body does nothing measurable 200k times and reports a median of samples, so the figure is stable to a few percent instead of the 356..3893 tok/s the engine-level replay gives on an unchanged binary.

On this host (Core Ultra 9 285K, 24 cores, no SMT), with the engine's own OMP settings:

threads fork/join per region
4 881 ns
8 1500 ns
16 2498 ns
24 12767 ns

12.8 Β΅s for a region with one iteration β€” 1.0 ms/token on a 78-layer model paying it once per layer, and 149Γ— the cost of the same loop inline.

A lead this does NOT establish

The 16 β†’ 24 jump is steeper than the trend, and this CPU is hybrid (P-cores + E-cores). That is the effect omp_tune.h already avoids on Apple Silicon, by counting only performance cores β€” #707 measured E-cores costing 4.2% of decode there. The Linux path has no equivalent: it dedupes SMT via thread_siblings_list and takes every physical core, E-cores included.

So Intel hybrid parts may have exactly the same bug. But this host runs under WSL2, which exposes neither /sys/devices/system/cpu/types/ nor cpu_capacity, so P/E membership is not observable here and the jump cannot be attributed. I am flagging it, not claiming it. Anyone on bare metal with a hybrid CPU can settle it with the included benchmark in a minute.

Verification

Identical output, not merely green:

  • teacher-forcing 32/32 positions, before and after
  • the REPLAY token sequence is byte-for-byte the same before and after
  • four engines build warning-free; make test-c passes

Side note that made this measurable at all

These numbers come from tests/test_inefficiency.py and its glm_tiny fixture. Those 8 tests had never executed on any platform β€” they looked for c/glm.exe, a name that has not existed since the glm β†’ colibri rename (fixed in #804). With that fix and a generated fixture, 5 of them run for the first time and pass. They are the throughput floor, the disk-wait ceiling, the PROFILE phase assertions and the determinism check β€” i.e. the harness this kind of work needs.

The DSA indexer's per-position top-k is `#pragma omp parallel for` over
`s < S`. At decode S is 1, so the region has exactly one iteration -- and
OpenMP still forks and joins the whole team to run it, once per layer, per
token. It happens whether or not the body does any work: both early
`continue` branches are taken on short contexts, and the fork precedes them.

`if(S > 1)` makes the single-iteration case run inline. Prefill (S > 1) is
untouched.

This is not a measurement-driven change -- a one-iteration parallel region
is overhead by construction, no benchmark needed to say so. But it came out
of one, and the measurement is worth recording because it points at
something larger.

Running the tiny oracle model (5 layers, hidden 128) at different team
sizes, 15 runs each, median with quartiles:

    OMP_NUM_THREADS=1     17109 tok/s   (p25 15573, p75 18119)
    OMP_NUM_THREADS=4     10749 tok/s   (p25 10650, p75 11125)
    OMP_NUM_THREADS=24     1529 tok/s   (p25   913, p75  1935)

The distributions do not overlap. One thread is 11x faster than 24, and the
spread widens with the team -- +-8% at one thread, +-34% at twenty-four,
which is the same tail the profiler reports as p90 27% above p50.

That is a small model on a 24-core host, so the ratio does not transfer to
a 744B run. What does transfer is the shape: colibri hands every parallel
region the full team regardless of how much work it contains, and there is
no notion of a minimum grain anywhere in the tree. PyTorch's at::parallel_for
takes a grain_size for exactly this reason, and vLLM's CPU backend carries
its own thread-count controls.

Fixing that properly means a per-region threshold, and a threshold is only
honest if it is measured on the model people actually run -- which needs the
lab machine, not a 0.015s tiny-model replay whose noise (356..3893 tok/s on
identical binaries) swamps any single-region effect. This commit takes only
the case that needs no threshold at all: a loop that cannot be parallel.

Verified identical, not just green: teacher-forcing is 32/32 positions and
the REPLAY token sequence is byte-for-byte the same before and after. Four
engines build warning-free, make test-c passes.
@ThefloorMiner

Copy link
Copy Markdown

A per-region threshold is only honest if it is measured on the model people actually run, and this machine cannot do that.

Mine can, and it is running your branch now. 4Γ— RTX A6000 + EPYC 7402P (24c/48t), GLM-5.2 744B int4, .coli_usage snapshot/restored between runs, three runs per side, plus a byte-for-byte output diff. I will post the numbers here.

Two things I can say before that, because one of them looks like it contradicts you and does not.

Your tiny-model result and my 744B result point in opposite directions, and both are right

You measured OMP_NUM_THREADS=1 at 11Γ— faster than 24. I measured the opposite on the real model β€” same host class, same engine:

OMP_NUM_THREADS routed CPU read decode per thread
4 6.48 GB/s 1.28 tok/s 1.62 GB/s
8 12.50 1.97 1.56
16 23.97 2.63 1.49
24 (physical) 30.10 2.90 1.25
48 (logical) 17.35 2.06 0.36

Near-linear to the physical core count, then a collapse at SMT.

The two results are the same finding seen from opposite ends of the grain axis. On glm_tiny (5 layers, hidden 128) a region has almost nothing in it, so fork/join is the entire cost and any team beyond one thread is pure loss. On GLM-5.2 one expert region is 3 Γ— 6144 Γ— 2048 MACs and there are up to 8 of them per layer β€” real work, and the team pays for itself right up to the point where SMT siblings start fighting over the same load/store units.

So your sentence β€”

colibri hands every parallel region the full team regardless of how much work it contains, and there is no notion of a minimum grain anywhere in the tree

β€” is exactly right, and having the two extremes measured makes it much harder to argue with than either number alone. A fixed team size cannot be correct for both, and today there is only one.

What I will report, and what I can run next

For this PR specifically: median of 3 before/after, the orchestration counter from PROF=1's P0-EXEC line (the closest thing the engine has to a fork/join proxy), and whether the greedy output is byte-identical. Worth noting for your determinism check: I verified last night that this engine is token-identical across six runs at temperature 0 on 4 GPUs, so an output diff here is a real signal rather than multi-GPU reduction noise.

On the larger question you deliberately left alone β€” the per-region threshold β€” I would rather measure than speculate, and I think the honest experiment is not a grain_size sweep but the phase counters. P0-EXEC already separates routed CPU, routed GPU critical, router and orchestration; on this host orchestration is 2.45 s of a 29.8 s run, i.e. 8.2 %, and that is the envelope any threshold work has to fit inside. If you want a specific region instrumented I can add a counter and run it against the 744B β€” say which one.

Unrelated but worth having: my #805 sets OMP_NUM_THREADS from physical cores on Linux and macOS, which today only happens on Windows. It is orthogonal to yours β€” you are removing a region that should not fork, I am fixing how large the team is when it does β€” but they touch the same subsystem, so flagging it rather than have you find out from a conflict.

`make -C c bench-omp-grain` measures what a parallel region costs before its
body runs: a loop whose iterations do nothing measurable, so the number IS
the fork/join. It is the tool behind the `if(S > 1)` clause -- and behind the
claim that a one-iteration region is not free.

On this host (Core Ultra 9 285K, 24 cores, no SMT):

    threads   fork/join per region
       4         881 ns
       8        1500 ns
      16        2498 ns
      24       12767 ns        (OMP_WAIT_POLICY=active OMP_PROC_BIND=close)

At 24 threads that is 12.8 us of overhead for a region with one iteration --
1.0 ms/token for a 78-layer model paying it once per layer, and 149x the cost
of running the same loop inline.

Why a benchmark and not just the engine: the tiny-model replay is 15 ms end
to end, and repeated runs of an UNCHANGED binary spread from 356 to 3893
tok/s. Nothing about a single parallel region survives that. This runs the
region 200k times and reports a median of samples, so the number is stable
to a few percent and comparable across machines.

One thing it is NOT evidence for. The jump between 16 and 24 threads is
steeper than the trend, and this CPU is hybrid (P-cores + E-cores) -- the
same effect omp_tune.h already avoids on Apple Silicon by counting only
performance cores (JustVugg#707, -4.2% decode from letting E-cores into the team).
The Linux path has no equivalent: it dedupes SMT via thread_siblings_list
and takes every physical core, E-cores included. That may well be the same
bug on Intel hybrid parts, but this host runs under WSL2, which exposes
neither /sys/devices/system/cpu/types/ nor cpu_capacity, so P/E membership
is not observable here and the jump cannot be attributed. Someone on bare
metal with a hybrid CPU can settle it with this benchmark in a minute.
@ThefloorMiner

Copy link
Copy Markdown

Measured on the 744B, as promised. Your change is free, and your tiny-model ratio does not transfer β€” which I think strengthens the PR rather than weakening it.

4Γ— RTX A6000 + EPYC 7402P, GLM-5.2 744B int4, CTX=32768, 64 tokens greedy, OMP_NUM_THREADS=24, .coli_usage restored byte-for-byte before every run, three runs per side:

run 1 run 2 run 3 median orchestration
dev (7fb1159) 1.81 1.91 1.92 1.91 tok/s 2.230 / 2.231 / 2.205 s
this PR 1.90 1.90 1.90 1.90 tok/s 2.218 / 2.204 / 2.227 s

No measurable difference, and the orchestration counter does not move either. Greedy output is byte-identical between the two branches β€” worth stating plainly because I separately verified this engine is token-identical across six runs at temperature 0 on 4 GPUs, so that identity is a real result rather than an untested assumption.

Why I read this as supporting the PR

You wrote that the region "needs no benchmark to justify" and that the 11Γ— is a shape, not a transferable ratio. Both hold up. On glm_tiny the DSA indexer's s < S region is essentially all fork/join; on GLM-5.2 the same region is a rounding error next to 75 MoE layers of int4 GEMV. A change that is free on the big model and large on the small one is exactly what "there is no notion of a minimum grain" predicts.

It is also the cheapest kind of correctness: a one-iteration parallel region cannot be right, and the measurement confirms it costs nothing to remove.

Where the grain question actually bites on a real model

Since you deliberately left the general threshold alone for want of a lab machine β€” here is what mine says, so the next person has a number to aim at.

PROF=1 on the tuned configuration, 29.8 s for 64 tokens:

expert-matmul  13.75 s  (46.2 %)
attention       7.80 s  (26.2 %)
disk wait       3.94 s  (13.3 %)
other           4.26 s  (14.3 %)   of which router 1.81 s, orchestration 2.45 s

Orchestration is 8.2 % of wall time on a 744B run. That is the entire envelope any grain-size work has to fit inside β€” it bounds the prize, and it is small. The thing that is not small on this host is one line up:

P0-EXEC: routed CPU 10.967s / 19.67 GB/s

19.67 GB/s of CPU-side expert reads on 8-channel DDR4-2400 that sustains ~85. I swept the team size to find out whether that was a broken parallelisation or a per-core ceiling:

OMP_NUM_THREADS routed CPU decode per thread
4 6.48 GB/s 1.28 tok/s 1.62 GB/s
16 23.97 2.63 1.49
24 (physical) 30.10 2.90 1.25
48 (logical) 17.35 2.06 0.36

Near-linear to the physical core count, then SMT collapse. So the parallelisation is fine and ~1.4 GB/s per Zen 2 core is the real ceiling β€” the team is not too large in general, it is too large by exactly the SMT factor. That is #805, and it is orthogonal to yours.

Happy to run any specific region you want instrumented against the 744B; the harness is scripted and the host is set up for it.

@ZacharyZcR
ZacharyZcR marked this pull request as ready for review August 3, 2026 21:11
@ZacharyZcR

Copy link
Copy Markdown
Contributor Author

This is the measurement I could not make, and it lands harder than the PR did on its own. Thank you for running it properly β€” snapshot/restore of .coli_usage, three runs per side, and a byte diff on the output is exactly the protocol this needed.

Taking your two results in turn.

Your 744B numbers correct my framing, not just complete it

You are right that the 11Γ— does not transfer, and I should not have let that number stand as close to the change as it did. On glm_tiny the region is fork/join and nothing else; on GLM-5.2 it is a rounding error beside 75 layers of int4 GEMV. Same defect, two ends of the grain axis β€” and your reading is the better one: a change that is free on the big model and large on the small one is what "there is no minimum grain anywhere in the tree" predicts. A fixed team size cannot be right for both.

So the honest claim for this PR is the narrow one: a one-iteration parallel region cannot be correct, and removing it costs nothing measurable on the model people actually run. Your table is what licenses that second half. I have no objection to the PR being judged on that basis alone.

On the general threshold β€” I think your data argues against pursuing it

You measured orchestration at 2.45 s of 29.8 s, 8.2 %, and it does not move between branches. That is the whole envelope, and the realisable fraction of it is smaller still, since most of that 8.2 % is fork/join amortised over regions that genuinely deserve a team. A grain_size mechanism would add a tuning knob and a per-region policy surface to chase single-digit percentages of single-digit percentages.

I would rather this be recorded as a bounded negative result than turned into work: the grain-threshold prize on a 744B decode is ≀ 8.2 % and realistically far less. If you would rather have it instrumented than argued, the one region I would still spend a counter on is the router β€” 1.81 s, called once per layer per token, small per call, high frequency, which is the shape most likely to be paying fork/join it cannot repay. If that one also comes back flat, the question is closed and worth closing loudly.

Everything else on your profile says the same thing about where the time is:

expert-matmul  13.75 s  (46.2 %)
attention       7.80 s  (26.2 %)
disk wait       3.94 s  (13.3 %)

#805 is the one with the actual headroom

routed CPU 19.67 GB/s on 8-channel DDR4-2400 that sustains ~85, with 24 physical β†’ 30.10 GB/s and 48 logical β†’ 17.35 GB/s, is a much larger and much better-evidenced finding than anything in this PR. ~1.4 GB/s per Zen 2 core being the ceiling, and the team being too large by exactly the SMT factor, is a clean result.

No conflict from my side: this PR deletes a region that should never have forked and touches nothing about team sizing. I will not go near OMP_NUM_THREADS selection β€” that is yours. If #805 lands first I will rebase.

One thing worth stating plainly for anyone reading later, since you verified it independently: token-identical across six runs at temperature 0 on 4 GPUs. That makes an output diff a real signal here rather than noise, which is what let your byte-for-byte comparison mean anything. That is a useful property to have on the record.

Sign up for free to join this conversation on GitHub. Already have an account? Sign in to comment

Labels

None yet

Projects

None yet

Development

Successfully merging this pull request may close these issues.

2 participants