Repository navigation
fix(hipblaslt): tolerate AMDSMI_STATUS_NOT_SUPPORTED for BDF-based device matching on WSL - #11987
Conversation
…nitor On WSL2/DXG, AMD-SMI can't expose a real PCI BDF to the guest, so amdsmi_get_gpu_bdf_id() returns AMDSMI_STATUS_NOT_SUPPORTED. AMDSMI_CHECK_EXC throws unconditionally on any non-success status, so an ordinary hipblaslt-bench run crashed on startup even when no telemetry was requested. - client.cpp: only call setDeviceId() when the efficiency monitor is actually enabled(), matching the gating pattern used elsewhere. - efficiency_monitor.hpp/.cpp: add isAmdsmiTelemetryUnavailable() and use it in GetAMDSMIIndex() so NOT_SUPPORTED warns once and falls back to processor 0 instead of aborting, even when telemetry is explicitly requested. Other AMD-SMI failures still throw as before. - Add efficiency_monitor_gtest.cpp covering the classifier. JIRA ID : ROCM-30983 Co-authored-by: Cursor <cursoragent@cursor.com>
✅ All Checks Passed — Ready for Review
📖 Need help? See the Policy FAQ for details on every check and how to fix failures. |
|
🎉 All checks passed! This PR is ready for review. |
There was a problem hiding this comment.
🟡 Changes recommended
One or more issues must be addressed before approval.
Once you've addressed the issues Copilot identified, you can request another Copilot review.
Pull request overview
Hardens optional hipBLASLt benchmark telemetry on WSL/DXG by avoiding unnecessary AMD-SMI access and tolerating unavailable PCI BDFs.
Changes:
- Gate device matching on telemetry being enabled.
- Fall back to processor 0 for
AMDSMI_STATUS_NOT_SUPPORTED. - Add and register host-only classifier tests.
File summaries
| File | Description |
|---|---|
| projects/hipblaslt/clients/tests/src/efficiency_monitor_gtest.cpp | Updated as part of this pull request. |
| projects/hipblaslt/clients/tests/src/CMakeLists.txt | Updated as part of this pull request. |
| projects/hipblaslt/clients/common/src/efficiency_monitor.cpp | Updated as part of this pull request. |
| projects/hipblaslt/clients/common/include/efficiency_monitor.hpp | Updated as part of this pull request. |
| projects/hipblaslt/clients/bench/src/client.cpp | Updated as part of this pull request. |
Review details
Suppressed comments (3)
projects/hipblaslt/clients/common/src/efficiency_monitor.cpp:499
- The added gtests only exercise the classifier, so they still pass if this call-site handling is removed (or if
return 0is changed back to throwing), which would reintroduce the WSL startup crash. Please add a test seam around the AMD-SMI lookup/status handling—or extract the fallback decision into a helper and test that theNOT_SUPPORTEDpath returns processor 0—so the regression test fails for the actual runtime bug rather than only for a classifier mutation.
if(isAmdsmiTelemetryUnavailable(bdfStatus))
{
projects/hipblaslt/clients/common/src/efficiency_monitor.cpp:515
- Returning literal
0ignoreshipDeviceIndex. If WSL exposes more than one HIP device and the user runs--device 1, the matmul runs on device 1 butcollect()indexesm_processorHandles[0], so the reported clocks and efficiency come from a different GPU. Please either establish a validated HIP-to-AMD-SMI ordering or disable telemetry when BDF mapping is unavailable instead of silently selecting processor 0.
return 0;
projects/hipblaslt/clients/tests/src/efficiency_monitor_gtest.cpp:17
- The quick/shared smoke lane invokes
--gtest_filter=*smoke*(seetest/therock/test_hipblaslt.py:92-99andclients/tests/test_categories.yaml:30-39), but these test names contain only the capitalizedSmokein the suite and no lowercasesmoketoken. Consequently none of the three new regression tests runs in that lane; use the existingsmoke_test-name convention (for example,caching_library_gtest.cpp:69-73) so the guard executes in CI.
TEST(EfficiencyMonitorSmoke, NotSupportedIsTolerated)
- Files reviewed: 5/5 changed files
- Comments generated: 2
- Review effort level: Lite
💡 Add a code-review agent skill or configure MCP servers for context-aware, tailored reviews. Learn more in the docs.
…D fix - Link amd_smi PUBLIC (not PRIVATE) from hipblaslt-clients-common, since efficiency_monitor.hpp is a public header that now includes <amd_smi/amdsmi.h> directly, and hipblaslt-bench only links hipblaslt-clients-common, not amd_smi itself. - Extract GetAMDSMIIndex()'s per-iteration BDF-status decision into a pure decideBdfMatch() helper so the NOT_SUPPORTED-tolerance policy is directly unit tested, not just the isAmdsmiTelemetryUnavailable() classifier. A reverted fix now fails smoke_NotSupportedFallsBackInsteadOfThrowing. - Add selectFallbackAmdsmiIndex() so the WSL fallback tracks hipDeviceIndex instead of always guessing processor 0, which could attribute telemetry to the wrong GPU on a multi-GPU WSL host. - Rename tests to the smoke_ naming convention so they run in the --gtest_filter=*smoke* CI quick lane (they didn't before: gtest filters are case-sensitive and "Smoke" != "smoke"). JIRA ID : ROCM-30983 Co-authored-by: Cursor <cursoragent@cursor.com>
There was a problem hiding this comment.
🟢 Approval recommended
Only minor documentation and test-report nits remain; no blocking implementation issues were identified.
Review details
Suppressed comments (1)
projects/hipblaslt/clients/common/src/efficiency_monitor.cpp:511
- The Technical Details says the unsupported-BDF path falls back to processor 0, but this implementation selects
hipDeviceIndexand clamps it withselectFallbackAmdsmiIndex(). Please update the PR description to document the actual multi-GPU behavior; otherwise the stated device-selection behavior is inaccurate.
if(isAmdsmiTelemetryUnavailable(bdfStatus))
{
// No PCI BDF to match against on this platform; warn once
// and use a best-effort processor index instead of aborting.
static bool warned = false;
- Files reviewed: 6/6 changed files
- Comments generated: 1
- Review effort level: Lite
Alex-Vasile
left a comment
Alex-Vasile
left a comment
There was a problem hiding this comment.
CMakeList changes look good.
newling
left a comment
newling
left a comment
There was a problem hiding this comment.
I don't understand the details, but the general shape looks good to me.
An agent raised one issue (see here)
The issue is related to some unreliable performance numbers is an edge case, which seems less important than not crashing. So I think it can be addressed in a follow-up (or ignored).
|
Thanks for looking, and thanks to your agent for the catch. Confirmed: Deliberately leaving this as-is for now rather than fixing it here. The reported bug (WSL2/DXG, single GPU, crash on every run) is fully fixed by this PR either way - the fallback only kicks in when BDF is unsupported, and for a single visible GPU it's unambiguous regardless of socket layout. The multi-socket case is a narrower, untested edge case (no multi-socket WSL/DXG hardware available to validate a real fix against), and fixing it properly would add meaningfully more complexity (a second identity signal via Recording it here for whoever hits it next: the fix would be to search all sockets before falling back, and use AMD-SMI's HIP enumeration ID as a secondary identity check when BDF is unavailable, only guessing an ordinal index when there's exactly one candidate GPU. |
Codecov Report❌ Patch coverage is
Additional details and impacted files@@ Coverage Diff @@
## develop #11987 +/- ##
===========================================
- Coverage 70.55% 70.54% -0.00%
===========================================
Files 2810 2810
Lines 462506 462530 +24
Branches 68099 68098 -1
===========================================
+ Hits 326279 326288 +9
- Misses 112684 112699 +15
Partials 23543 23543
*This pull request uses carry forward flags. Click here to find out more.
🚀 New features to boost your workflow:
|
|
Dug into the 0% patch coverage report. Two different things are going on here, so here's our own estimate of what's actually exercised, since codecov's number understates one file and is honestly accurate for the other.
Net: patch coverage is real but partial - strong on the testable decision logic, intentionally absent on the hardware-dependent glue around it, plus one small untested defensive branch we can add a test for if useful. |
TheRock Submodule Bump ActivityNewest first
|
…vice matching on WSL (#11987) (#12133) Cherry-picks commit 62d6012 into the ROCm 10.1 release branch. JIRA ID: AIRDEL-56 --- JIRA ID : ROCM-30983 ## Motivation On WSL2 (DXG paravirtualized GPU passthrough), running hipblaslt-bench crashes on startup. `EfficiencyMonitor::setDeviceId()` unconditionally tries to match the target GPU via AMD-SMI's PCI BDF, but WSL2/DXG does not expose a real PCI BDF to the guest, so `amdsmi_get_gpu_bdf_id()` returns `AMDSMI_STATUS_NOT_SUPPORTED`. `AMDSMI_CHECK_EXC` throws unconditionally on any non-success status, so this turns into an uncaught `std::runtime_error` that takes down the whole client, even on a completely ordinary run where the user never asked for GPU telemetry. ## Technical Details - `clients/bench/src/client.cpp`: only call `setDeviceId()` when the perf monitor is actually `enabled()`, matching the existing `enabled()`-gating pattern used elsewhere (e.g. `argument_model.cpp`). An ordinary run (no `HIPBLASLT_BENCH_FREQ` / `HIPBLASLT_BENCH_PERF` env vars) now never touches AMD-SMI at all. - `efficiency_monitor.hpp`: added `isAmdsmiTelemetryUnavailable()`, a small pure function that classifies `AMDSMI_STATUS_NOT_SUPPORTED` as "telemetry not available on this platform" rather than a real error. - `efficiency_monitor.cpp` (`GetAMDSMIIndex()`): even when a user does explicitly request telemetry, tolerate `AMDSMI_STATUS_NOT_SUPPORTED` from `amdsmi_get_gpu_bdf_id()` by warning once (via `hipblaslt_cerr`) and falling back to a HIP-device-index-based processor slot (`selectFallbackAmdsmiIndex()`, clamped to the AMD-SMI device count) instead of throwing. Other AMD-SMI failures still go through `AMDSMI_CHECK_EXC` and abort as before. - `clients/tests/src/efficiency_monitor_gtest.cpp` (new): host-only gtest covering the classifier (`NOT_SUPPORTED` tolerated, `SUCCESS`/`UNKNOWN_ERROR` not), registered in `clients/tests/src/CMakeLists.txt`, so the fix has an actual regression test instead of relying on WSL hardware being present in CI. ## Test Plan - New `EfficiencyMonitorSmoke` gtest suite (10 tests) compiled into `hipblaslt-test`, so it runs wherever the gating Math CI `precheckin` job already runs `hipblaslt-test`. - No WSL2/DXG box was available in this environment, so I verified locally instead: configured and built `hipblaslt-bench` and `hipblaslt-test` in host-only mode (`HIPBLASLT_ENABLE_DEVICE=OFF`, no Tensile device libs) against a fetched ROCm 10.1.0 toolchain, to compile-check the changed files and run the new tests plus the existing host-only smoke suites. ## Test Result - `./hipblaslt-test --gtest_filter=EfficiencyMonitorSmoke.*` -> 10/10 passed. - Reran alongside the existing `SecureEnv` (5 tests) and `ArchRevisionSmoke` (4 tests) host-only suites -> 19/19 passed, no regressions. - `hipblaslt-clients-common`, `hipblaslt-bench`, and `hipblaslt-test` all built clean with the changes. - Not run: the full GPU matmul suite, or an actual WSL2/DXG reproduction (no such host was available here) - that's why the regression test targets the pure classifier logic rather than the runtime crash path itself. ## Submission Checklist - [x] Look over the contributing guidelines at https://github.com/ROCm/TheRock/blob/main/CONTRIBUTING.md. ## Risk level Low. The change only affects an optional benchmark telemetry path (`EfficiencyMonitor`), which is opt-in via env vars in the common case, and it only adds tolerance for one specific AMD-SMI status code - every other AMD-SMI failure still throws as before. Co-authored-by: Cursor <cursoragent@cursor.com>
JIRA ID : ROCM-30983
Motivation
On WSL2 (DXG paravirtualized GPU passthrough), running hipblaslt-bench
crashes on startup.
EfficiencyMonitor::setDeviceId()unconditionallytries to match the target GPU via AMD-SMI's PCI BDF, but WSL2/DXG does
not expose a real PCI BDF to the guest, so
amdsmi_get_gpu_bdf_id()returns
AMDSMI_STATUS_NOT_SUPPORTED.AMDSMI_CHECK_EXCthrowsunconditionally on any non-success status, so this turns into an
uncaught
std::runtime_errorthat takes down the whole client, evenon a completely ordinary run where the user never asked for GPU
telemetry.
Technical Details
clients/bench/src/client.cpp: only callsetDeviceId()when theperf monitor is actually
enabled(), matching the existingenabled()-gating pattern used elsewhere (e.g.argument_model.cpp). An ordinary run (noHIPBLASLT_BENCH_FREQ/HIPBLASLT_BENCH_PERFenv vars) now never touches AMD-SMI at all.efficiency_monitor.hpp: addedisAmdsmiTelemetryUnavailable(), asmall pure function that classifies
AMDSMI_STATUS_NOT_SUPPORTEDas"telemetry not available on this platform" rather than a real error.
efficiency_monitor.cpp(GetAMDSMIIndex()): even when a user doesexplicitly request telemetry, tolerate
AMDSMI_STATUS_NOT_SUPPORTEDfrom
amdsmi_get_gpu_bdf_id()by warning once (viahipblaslt_cerr)and falling back to processor 0 instead of throwing. Other AMD-SMI
failures still go through
AMDSMI_CHECK_EXCand abort as before.clients/tests/src/efficiency_monitor_gtest.cpp(new): host-onlygtest covering the classifier (
NOT_SUPPORTEDtolerated,SUCCESS/UNKNOWN_ERRORnot), registered inclients/tests/src/CMakeLists.txt, so the fix has an actualregression test instead of relying on WSL hardware being present in
CI.
Test Plan
EfficiencyMonitorSmokegtest suite compiled intohipblaslt-test, so it runs wherever the gating Math CIprecheckinjob already runshipblaslt-test.locally instead: configured and built
hipblaslt-benchandhipblaslt-testin host-only mode (HIPBLASLT_ENABLE_DEVICE=OFF,no Tensile device libs) against a fetched ROCm 10.1.0 toolchain, to
compile-check the changed files and run the new tests plus the
existing host-only smoke suites.
Test Result
./hipblaslt-test --gtest_filter=EfficiencyMonitorSmoke.*-> 3/3passed.
SecureEnvandArchRevisionSmokehost-only suites -> 12/12 passed, no regressions.
hipblaslt-clients-common,hipblaslt-bench, andhipblaslt-testall built clean with the changes.
reproduction (no such host was available here) - that's why the
regression test targets the pure classifier logic rather than the
runtime crash path itself.
Submission Checklist
Risk level
Low. The change only affects an optional benchmark telemetry path
(
EfficiencyMonitor), which is opt-in via env vars in the commoncase, and it only adds tolerance for one specific AMD-SMI status
code - every other AMD-SMI failure still throws as before.