fix: resolve lint error by zhyncs · Pull Request #650 · sgl-project/sglang

zhyncs · 2024-07-18T10:10:47Z

Motivation

fix https://github.com/sgl-project/sglang/actions/runs/9989535244/job/27608367620
as titled cc @Ying1123 @wisclmy0611 @hnyls2002

Modification

as titled

Checklist

Ensure pre-commit or other linting tools are used to fix potential lint issues.
Confirm that modifications are covered by complete unit tests. If not, please add more unit tests for correctness.
Modify documentation as needed, such as docstrings or example tutorials.

…me the wedged rank THE GAP. The census heartbeat is a scheduler-thread collective and dies with the hung rank — structurally blind in exactly the case it exists for — and the abort dump spoke only rank-local state: a survivor's Bar1CollectiveAborted could not say where the wedged peer was. THE FIX. The lockstep sentinel's sidecar keeps exchanging while a main thread hangs (proven mid-wedge on-card during the sgl-project#622 hunt). Its last successful gather (every peer's ring position) and its last anatomy exchange (every peer's tail = last op entered) are now RETAINED, and peer_statement() formats them: rank, ring seq, age of the observation, last op, plus a SYNC LOST clause when a peer's sidecar has since died — the listed positions are then its last life sign. The barlink abort raise appends the statement as 'PEER POSITIONS (sgl-project#650)', never-raise guarded. FALSIFIER, red-first (executed record in the test docstring): on the pre-fix tree, module-level peer_statement absent, instance method absent, raise text without PEER POSITIONS — all three checked False by direct import of the pre-fix module. Green: synthetic hang over 3 real gloo processes (one rank stops noting at seq 500, its sidecar keeps participating — the real wedge shape); both survivors' statements name 'rank 2 at ring seq 500' and its last op from the retained tails; disarmed path safe; source invariant pins the raise wiring. Counting basis: 3 new test functions; 7 test items green with the sentinel regression suite (34.8 s); ruff clean; barlink import clean.

…tement in abort dumps The sentinel sidecar's last gather (peer ring positions) and last anatomy exchange (peer last-entered ops) are retained and appended to every Bar1CollectiveAborted as 'PEER POSITIONS (sgl-project#650)', with a SYNC LOST clause when a peer's sidecar has died. Falsifier red-first over 3 real gloo processes; 3 new test functions, 7 items green with the sentinel regression suite.

…e, never __trap The last live instance of the sgl-project#583 context-killing mechanism: wait_ge in the host-staged transport still called __trap() on spin expiry, which destroys the CUDA context and surfaces as sticky 'unspecified launch failure' at whatever unrelated call site runs next. Unreachable in every production configuration (host is explicitly-selected, no-fallback, bar1 pinned) — but a landmine for any host-transport measurement boot. Ported the device transport's sgl-project#583 fix honestly, not minimally: - wait_ge returns bool and writes a per-site code into the abort word on expiry (put reuse-guard=1, reduce peer-wait=2, copyout wait=3); the BARLINK_HOST_TRAP macro and its hipify plumbing are deleted. - Block-uniform exit: shared abortS written by the spinning thread 0, __syncthreads(), all threads return together before any payload loop, fence, or last_block barrier — verified at all three sites. Terminal- abort contract documented: cross-block skew and stale blk_ctr are acceptable residue of a process the host is about to kill by raise. - The transport's THREE sequence-counter families (collective, per-peer send, per-peer recv) are unified into one zero-initialized backing tensor with 2-word rows — the spec's naive grow-to-2 would have made the p2p kernels' abort store corrupt a NEIGHBOR's sequence number, and the unified shape makes check_aborted a single read over every abort word the transport can produce. - HostCollectiveAborted mirrors DeviceCollectiveAborted (rank/world/ code/where, garbage-results contract) and appends the sgl-project#650 peer statement. check_aborted is a plain in-line read (documented choice: this is a measurement vehicle, not the bs=1 hot path) wired twice: the duck-typed _after_transport seam (sgl-project#431) for eager collectives, and abort-gate registration for replay-only collectives — host is capturable, so a graph-resident collective never reaches the seam (the same gap sgl-project#583 closed for the device transport). - The load-bearing reachability claim is pinned: 'host' is in the no-fallback set via the real parallel_state lookup, and the dispatch seam is proven transport-stable mid-run — an explicit SGLANG_BARLINK_TRANSPORT=host boot is the only way this code runs. Falsifiers, red-first (executed against the pre-fix source inside the test): old wait_ge provably trapped and returned void, old seq tensor had no abort-word room; fixed source has zero traps, bool returns, all three consuming sites, distinct codes. nvcc compile smoke rc=0 (sm_86). GPU-gated on-card falsifier written and documented (put-kernel forced timeout must set the word, leave the context ALIVE, and raise named) — skips at the desk by design. Counting basis: 35 passed + 1 GPU-skip + 19 subtests in the new file; full battery with abort/stub-drift/sentinel regression suites 81 passed + 1 skipped + 19 subtests (97.9 s, my rerun); ruff and imports clean.

…port wait_ge wait_ge in the host transport writes a per-site abort code and returns instead of trapping the context; block-uniform exit via shared flag + barrier at all three sites; HostCollectiveAborted mirrors the structured device/bar1 shape and carries the sgl-project#650 peer statement. Also closes the capturable-transport abort-gate wiring gap for host, and unifies the abort/seq words into one zero-initialized backing tensor (the spec's grow-to-2 layout would have corrupted p2p slice neighbors). Falsifiers red-first against the pre-fix source; 35 passed + 1 GPU-skip + 19 subtests new; 81 passed full battery.

…stead of clobbering it The re-issue this ticket asked for already existed. A cutover retracts every resident; the retracted request is re-queued through _add_request_to_queue, which calls _prefetch_kvcache, which re-runs prefetch_from_storage -- a fresh radix walk under the CURRENT binding generation, allocating a fresh host span from the currently bound pool. Building a second requeue path would have been redundant machinery on top of a working one. What did not exist was a guard on the registration. prefetch_from_storage ended with an unconditional `self.ongoing_prefetch[req_id] = ...`, so when the post-cutover registration arrived while the pre-cutover one was still ongoing, the incumbent was overwritten outright: its host_indices were lost to every owner (nothing else holds them) and its anchor_lock_params lock ref was never decremented, leaving that node PROTECTED for good. The next cutover's drop_prefix_tree_returning_rows then orphans exactly those device rows -- one request's allocation per cutover, which is the sgl-project#938 forensic shape, and it needs no in-flight write-through to explain it. RETIRE, DO NOT REAP. Freeing the displaced span at the registration site is not available: whether it is safe to release is answered by can_terminate_prefetch, which checks pool_transfers_done and is COLLECTIVE, and HiCacheController.terminate_prefetch is only mark_terminate() -- a flag, not a join -- so the transfer thread may still be writing into that span (sgl-project#913 IMA family). Running that collective there is not available either: whether a clobber happens is RANK-LOCAL, and the site already sits inside the sgl-project#580 participation region. So the record is retired -- terminated, moved out of the req_id slot, nothing freed -- and drain_retired_prefetch reaps it from the per-round site, where collectives are legal. That also DEFINES the ordering instead of leaving it to timing: the re-issue owns the slot, so a late completion resolves through check_prefetch_progress to the new record and cannot reach the displaced one, which is reachable only from the reap. Neither insert nor free happens on a retired record anywhere else. UNIFORMITY IS ENFORCED, NOT INFERRED. It is tempting to argue the retired list is rank-uniform (retirement follows a registration that only happens under the sgl-project#580 positive consensus), but this module refuses arguments of that shape elsewhere -- _prefetch_done_for RAISES rather than assume a request is in the replicated set -- and a wrong inference here hangs three ranks. Both halves are enforced instead: candidates are taken in sorted req_id order (canonical, never insertion order), and the ranks AGREE on the candidate before anyone reaps it, via one all_reduce contributing [d, -d] under MIN so min(d) and max(d) arrive in a single collective. Reaping proceeds only on min == max != 0. A rank that has not retired it yet contributes 0, agreement fails, and the record waits a round. A latecomer delays a reap; it never wedges one. The collective is UNCONDITIONAL -- gating it on a non-empty local list would be the rank-local participation predicate this whole design exists to avoid. Empty contributes zeros; the common case is one two-element all_reduce per round. Not micro-optimised into an existing round collective: correctness first, measure before optimising. The re-fetch budget is a REPORT, not a gate (sgl-project#939 RE-FETCH BUDGET SPENT ... recomputing). Refusing a re-issue would be a rank-local decision taken inside the participation region, which is the one thing this path may not do; the line exists so a cutover cadence faster than a fetch completes is named instead of surfacing only as an unexplained cached=0. Red-first, single process (13 behavioural + wiring pins): retire does not free, the span returns to its MINTING pool and never to the bound one, the drain empties the list with no further event from the request at all (otherwise this trades a clobber for unbounded growth), a completion cannot touch a retired record, and the registration retires before it assigns. Mutants killed: retire frees immediately (5 red), the clobber restored (1 red). MULTI-RANK, and I had this wrong first. I recorded "the agreement only bites multi-rank, which a CPU test can't exercise" -- false, and the user caught it: three gloo processes on localhost exercise it fine, as sgl-project#630/sgl-project#650/sgl-project#653/sgl-project#899 already do here. Only a SINGLE-PROCESS test cannot. The new gloo falsifier runs the real drain and the real _all_reduce_attn_groups over three real ranks with DIVERGENT retired lists. With the agreement every rank reports rounds [0, 1] -- nobody reaps while rank 2 is behind, everybody reaps once it catches up. Removing it does NOT hang, which is worse and is why this needed measuring rather than predicting: gloo pairs ranks 0/1's round-1 all_reduce with rank 2's round-2 one -- same group, same shape, nothing to distinguish them -- so no rank blocks, every rank reports success, and the ranks silently reap in different rounds. That is the sgl-project#580 signature, not a deadlock, and the next collective carrying a differently-shaped payload is the one that corrupts. The test asserts on that divergence (and accepts a timeout too, since another backend may block). Verified to kill the production mutant only once BOTH agreement gates are removed -- the second, `digest != agreed_min`, independently covers the all-but-one-rank case, so a partial mutation leaves it alive. Also fixed here, and it was mine: test_prefetch_participation_desync_580's cache stand-in drives prefetch_from_storage as a real bound method and had no _retire_ongoing_prefetch, so the new call broke 4 of its tests. Carried as the REAL bound method rather than a stub -- production instances always have it, so a defensive getattr would have hidden a genuine missing-method bug. Tests: test/registered/unit/mem_cache/ plus the gloo module -- 2875 passed, 3 failed (test_acceptance_emitters_758 RefillTiming x2, test_kv_row_ownership_822 TestCallSiteAdapters), all three identical by name and count on the pristine file at HEAD, i.e. pre-existing. ruff clean on both new modules and unchanged on the two production files (unified_radix_cache 0/0, scheduler 103/103 pre-existing); black clean. The gloo module spawns real ranks, so it belongs in the narrow RANKS lane. gate_partition.tsv is generated and sha256-checked, so it is not hand-edited here; gate_partition_build.py has to be re-run for the classification to land. Stage-A interaction, so a reader does not misdiagnose it: a retired-but-unreaped record still holds its lock ref, so sgl-project#938 PROTECTED RESIDUE ORPHANED can show it TRANSIENTLY. Healthy is a counter that RETURNS across reap rounds; monotone growth is still a leak.

…he silent hop Boot 2 of window-flip-0828 wedged silently for 10+ minutes: PP0 and PP1 blocked in gloo waitSend under _pp_commit_comm_work's naked p2p_work.work.wait() (PP1 via _pp_commit_pending_req_work), PP2 in _do_recv -- a closed three-arc cycle on a group whose own timeout is two hours. The sgl-project#753 comment in this file already described the exact shape (boot v7pp9); the ordering was fixed then, the wait stayed naked. The bound routes through the sgl-project#630/sgl-project#829 canon (hicache_collective.bounded_wait -> ParkedWait): the unbounded wait() parks on a thread and the deadline is on the JOIN. Deliberate deviation from the order's 'pass the deadline into work.wait(timeout)': sgl-project#829 retracted that design in this tree -- an expired timed Work.wait closes the gloo pair, and hicache_collective.py names _pp_commit_comm_work as a measured victim (34 of 262 boot logs). Budget 120 s (SGLANG_PP_RING_COMMIT_BUDGET_S; <= 0 = documented escape hatch, byte-for-byte pre-sgl-project#973), reasoned against the canon's constants: 12.6x the longest healthy cutover, 4x the sgl-project#789 budget, 5x under the HiCache bound, 60x under the group timeout. On expiry: RingCommitTimeout with a sgl-project#650-style peer statement naming the silent hop; transport failure is NOT converted (sgl-project#734 stays distinguishable). Helpers are module-level, not methods -- the method form broke ~12 one-method-at-a-time stand-in holders (measured: 796 5->2, 801 14->3), recorded in a code comment. Wait-site audit: :5253 bounded-now; :3270 bounded by delegation (proven by test); _do_recv/:7592 and :5545 bounded-already via the sgl-project#789 gate; parallel_state.py recv_object stays unbounded and is NAMED as its own posten (needs the resumable PpChainReceiver treatment, not a terminal bound mid-protocol). Tests: test_ring_commit_bounded_973.py, 5 arms on real 3-process gloo -- pre-sgl-project#973 naked commit HANGS (distinguished from raised via fsynced progress marker), bounded commit raises within budget with the peer statement, healthy paired traffic unchanged, escape-hatch mutant hangs again (can-fail), :3270 reachability arm. Gates: BEFORE == AFTER byte-identical on 796(5)/801(14)/797(31)/630(5+14)/829(7) plus an 18-module sibling sweep; extraction count probe OK on every run. ruff/codespell: 0 new findings.

…lt ein TRIGGER. Aktionsarm ist S. Reiner Lese-Pass, kein Boot, kein Code. Beide Quellen je Zeile. == PFLICHT-VORFRAGE BEANTWORTET: es ist ein UNGEDECKTER TRANSPORT == Das haengende all_reduce laeuft auf `grp = self.tp_cpu_group` (scheduler.py:5964) -- eine GLOO-CPU-Gruppe. Die sgl-project#650/sgl-project#653-Klassen (Bar1CollectiveAborted / HostCollectiveAborted / Bar1CollectiveStalled) liegen ausschliesslich in `distributed/device_communicators/barlink_*.py` plus den zwei debug_utils-Lesern: sie decken den DEVICE/BAR1-Pfad. Also NICHT "Deckung da, Schwellwert/Leser defekt", sondern: der Pfad hat diese Deckung nie gehabt. Auf dem gloo-Pfad gibt es weder Deadline noch Peer-Positions-Bericht. == CENSUS-FAEHIGKEIT: IDENTITAET IST DA, und mehr == `_history.append((family, nbytes))` (:246) -- jede Eintragung traegt die KOLLEKTIV-FAMILIE, nicht nur Zaehlstaende. `Divergence(family, counts)` (:459) vergleicht bereits NACH FAMILIE. Und der Modulkopf beschreibt zwei Berichtspfade: 1. `compare_across_ranks` -- Detektor, laeuft auf gloo `tp_cpu_group`, replizierte Kadenz, feuert waehrend die Raenge noch gesund sind. 2. `format_local_census` -- die WEDGE-FESTE Haelfte: NIMMT KEIN KOLLEKTIV, jeder Rang kippt seine EIGENEN Zaehler ins Log, die drei Dumps werden hinterher gediffed. Die Identitaet fehlt also NICHT. Das ist nicht die Luecke. == DIE LUECKE, praezise == Der Modulkopf sagt woertlich, der Detektor laufe auf dem CPU-Pfad, "NEVER the device/BAR1 path, SO A WEDGED DEVICE GROUP DOES NOT DISABLE THE INSTRUMENT THAT IS SUPPOSED TO EXPLAIN IT". Genau diese Immunitaet greift bei uns NICHT: unser Wedge IST auf der CPU-Gruppe. Der Detektor braucht ein `all_gather_object` auf ebendieser Gruppe -- er kann eine Uneinigkeit der Gruppe, die er selbst benutzt, nicht melden. Die Schutz-Annahme ist um genau einen Transport verschoben. Und die wedge-feste Haelfte (`format_local_census`, kollektiv-frei) EXISTIERT fuer diesen Fall -- aber sie wird nie AUSGELOEST, weil auf dem gloo-Pfad nichts abbricht: kein Deadline, kein Abort, kein Ausloeser. DAMIT: die Luecke ist ein TRIGGER, keine Faehigkeit. == AKTIONSARM-DESIGN, S == Diskriminator: NICHT "Follower wartet lange" (Normalzustand) und NICHT "PP0 macht keinen Fortschritt" (PP0 wartet selbst) -- sondern eine BEGRENZTE Wartezeit am gloo-Kollektiv, nach deren Ablauf jeder Rang `format_local_census()` ruft (kollektiv-frei, schon gebaut) und die Gruppe per CRASH/STOP stirbt. Die Beider-Seiten-Diagnose faellt dabei GRATIS an: jeder Rang druckt seine eigene Familien-Historie, und der bestehende Hinweistext sagt bereits, was damit zu tun ist -- "Diff this line against the peers' to find where the sequences diverged". DEADLINE-BEGRUENDUNG gegen die zwei Lehren: * sgl-project#630 (wait() treiben, nie is_completed() pollen): der bounded wait muss ueber das async work-handle mit Timeout laufen, NICHT ueber eine Poll-Schleife. * sgl-project#615 (Build-Fenster-Dehnung): die Deadline darf nicht in ein legitimes JIT-Cold-Build-Fenster feuern (bis SGLANG_BARLINK_BUILD_WINDOW_CAP_S=900s). Sie muss dasselbe Fenster konsultieren, das barlink schon fuehrt -- sonst ersetzt sie einen Wedge durch einen Fehlalarm beim ersten Boot. KEIN timeout-und-weiter: Ablauf = Tod der Gruppe, per Rang-Gesetz. == VERDIKT: S == Wiederverwendet wird durchweg Vorhandenes (Census-Familien-Historie, kollektiv-freier Dump, barlink-Build-Fenster). Neu ist nur der bounded wait plus sein Ausloeser. Kein neues Kollektiv auf dem Pfad -- das Gesetz bleibt gewahrt, weil die Diagnose-Haelfte ausdruecklich kollektiv-frei ist. OFFEN und ehrlich: die Deadline-ZAHL ist nicht herleitbar aus dem, was ich gemessen habe -- sie braucht die legitime Maximal-Wartezeit unter Last, und die kenne ich nur als "DE-06 dauerte 95,6 s im gesunden Lauf". Das ist die eine Groesse, die vor dem Bau gemessen gehoert. BELEG-STUFE: DESK-BEWIESEN (Transport-Zuordnung, Deckungs-Scope, Census- Faehigkeit -- alle am Pin gelesen, Modulkopf woertlich zitiert).

fix: resolve lint error

cd2a7c3

zhyncs requested review from Ying1123, hnyls2002 and wisclmy0611 July 18, 2024 10:10

Ying1123 merged commit 9c5cac2 into sgl-project:main Jul 18, 2024

zhyncs deleted the fix-lint branch July 18, 2024 10:42

timethink pushed a commit to timethink/sglang that referenced this pull request Mar 9, 2025

fix: resolve lint error (sgl-project#650)

42ebef2

Navigation Menu

Uh oh!

fix: resolve lint error - #650

fix: resolve lint error#650
Ying1123 merged 1 commit into
sgl-project:mainfrom
zhyncs:fix-lint

zhyncs commented Jul 18, 2024

Uh oh!

Reviewers

Assignees

Labels

Projects

Milestone

Development

Uh oh!

2 participants

Uh oh!

Conversation

zhyncs commented Jul 18, 2024

Motivation

Modification

Checklist

Uh oh!

Reviewers

Assignees

Labels

Projects

Milestone

Development

Uh oh!

2 participants