Repository navigation
test(rccl): check config parsers keep the caller's strtok state - #12648
thomas-huber wants to merge 2 commits into
Conversation
a217f56 to
5e4b885
Compare
✅ All Policy Checks Passed
📖 Need help? See the Policy FAQ for details on every check and how to fix failures. |
a19c143 to
4f48764
Compare
![amd-argus[bot]](https://avatars.githubusercontent.com/in/4740020?s=60&v=4){kind=small}
|
This pull request has been inactive for 25 days and will be marked as stale. If you would like to keep this PR open, please:
This PR will be automatically closed in 5 days if no further activity occurs. |
Runs the NCCL_CTA_POLICY parser, ncclStrToCpuset() and ncclStrListToCpuset() on a second thread while the calling thread is between strtok() calls on its own string, then checks the caller's next token. strtok() has one process-wide save pointer: before the switch to strtok_r() the parser leaves it at the end of its own string, so the caller gets NULL instead of "b". Upstream fix: NVIDIA/nccl@35512d6f6b73 ("Squash commit for Fix strtok thread safety")
Follow-up to review feedback: - Assert that the parser's free() went through the deferral hook. If the hook stopped engaging, a clobbered strtok pointer would point into freed memory while the tests still passed. - Note that ncclStrListToCpuset reads "1,3" as a CPU list, not a mask.
02d842a to
80ca602
Compare
|
Rebased onto The tests are unchanged. On the new base they pass 3/3, and fail 3/3 with the parsers' |
PR Pundit review(Testing PR Pundit Code Review agent. Not all comments may be relevant. This is a one-time automated pass: please address what is useful, no need to request a re-review. Please reply to each point, whether it applies or not: we harvest these replies, and both kinds are enormously helpful for tuning the agent. Reach out to @peymanr for suggestions.)
1. 🟠 Should fix: Cpuset-parser concurrency regressions report under init microtest binary: this PR adds cpuset tests to
|
Motivation
NCCL 2.32.3 fixed config parsers that used
strtok(), whose single process-wide state let a parse on one thread corrupt another thread's tokenization in progress (NVIDIA/nccl#2361). RCCL picks up the fix with the 2.32.3 sync in #12321. This PR adds a regression test that fails without the fix and passes with it.Technical Details
InitMicrotestcases intest/host/init-test.cc(rccl-UnitTestsMicroInit) run theNCCL_CTA_POLICYparser,ncclStrToCpuset()andncclStrListToCpuset()on a second thread while the calling thread is betweenstrtok()calls on"a,b", then check that the caller's next token is still"b". The file now also includescpuset.h; thencclOsCpu*helpers its parsers call come fromfakes/os_fakes.cc.strtok()calls leave the shared save pointer at the end of their string, so the caller getsNULL. Theirfree()is deferred through the existingg_microFreeseam, so this is a clean assertion failure, not a use-after-free.GetEnvCtaPolicy_TwoNamedModes_AccumulatesBothandGetEnvCtaPolicy_UnknownTokenAmongValid_IgnoresOnlyTheUnknowntests.InitMicrotest.*intest_categories_micro_init.yaml, used by all threerccl-UnitTestsMicroInitvariants, which the host unit-test workflow runs.developwith the NCCL 2.32.3 sync (sync(rccl): Sync with NCCL v2.32.3-1 #12321). Upstream fix: NVIDIA/nccl@35512d6f6b73.Issue Tracking
JIRA ID: AICOMRCCL-2712
Test Plan
test/host/build/rccl-UnitTestsMicroInit --gtest_filter='InitMicrotest.*_ConcurrentParse_KeepsCallerStrtokState'(built withtest/host/run_host_tests.sh build) ondevelop(42e7f5d9cb8b), and on the same tree with the fix reverted (the parsers'strtok_r()calls incpuset.handinit.ccback tostrtok()), three runs each, in therocm/dev-ubuntu-22.04:7.2image the host unit-test workflow uses.run_host_tests.shbuildandrunphases pass on the PR head.b0299c14dd71) againstdevelopbefore the sync (5e6a51f28579) on MI355X (gfx950) and MI300X (gfx942), ROCm 7.0.2, with the same results.Test Result
(null)instead of"b").Submission Checklist