[blas][rocblas] Restore the pointer mode when a rocBLAS call throws

[zjin-lcf]

Summary

Fixes the pointer-mode leak that can produce the intermittent RotTests.RealSinglePrecision failure reported in #486.

The reduction routines (asum, nrm2, dot, dotc, dotu, rotg, rotm, rotmg, iamax, iamin) switch the rocBLAS handle to device pointer mode and switch it back after the call:

rocblas_set_pointer_mode(handle, rocblas_pointer_mode_device);
...
rocblas_native_func(func, err, handle, n, x_, std::abs(incx), res_);
rocblas_set_pointer_mode(handle, rocblas_pointer_mode_host);

rocblas_native_func throws a rocblas_error on any status other than rocblas_status_success, so the reset is skipped. Handles are cached in thread-local storage and reused, so the handle stays in device pointer mode for every subsequent call on that thread.

Routines that pass a host address for their scalar arguments then have that address dereferenced on the device. rot is the most exposed one — it forwards (rocDataType2*)&c and (rocDataType3*)&s and never sets the pointer mode itself, relying entirely on the handle already being in host mode. On a discrete GPU, where host memory is not device-addressable, that is an illegal address.

This replaces the manual set/reset pairs at all 13 sites with an RAII guard, so the mode is restored during stack unwinding as well as on the normal path. The guard restores the previous mode rather than unconditionally forcing host mode, so it is safe to nest.

Validation

On an AMD Instinct MI210 (gfx90a, ROCm 7.1.1, DPC++), provoking a failure inside a guarded region and then calling rot:

	crashes
without the guard	10 / 10
with the guard	0 / 10

Without the guard the rot call fails with hipErrorIllegalAddress (700) and leaves the queue unrecoverable. With it, rot returns correct results and a following asum is correct. Both arms take the identical first exception on the same queue, so the guard is the only difference.

Reproducer:

// step 1: provoke a failure inside a pointer-mode-guarded region
try {
    oneapi::math::blas::column_major::asum(q, n, x, 1, (float*)nullptr).wait();
} catch (const std::exception& e) { /* expected */ }

// step 2: rot passes &c and &s, which are host addresses
const float c = 1.0f, s = 0.0f;          // identity rotation
oneapi::math::blas::column_major::rot(q, n, x, 1, y, 1, c, s).wait();

Also ran the BLAS test suite on the same MI210 with the fix applied: 740 tests pass with no failures.

Caveats

• The reproducer forces the initial rocBLAS failure with a null result pointer. This change explains the cascade given that some earlier reduction call failed; it does not identify what failed first in the CI run behind [BLAS] rocBLAS backend test can fail with HSA_STATUS_ERROR_INVALID_PACKET_FORMAT #486.
• [BLAS] rocBLAS backend test can fail with HSA_STATUS_ERROR_INVALID_PACKET_FORMAT #486 was reported with AdaptiveCpp, which surfaces the fault as HSA_STATUS_ERROR_INVALID_PACKET_FORMAT. I could only verify with DPC++, which reports the same underlying invalid device address as hipErrorIllegalAddress. AdaptiveCpp was not available on the machine used.

Test plan

• Builds against develop with the rocBLAS backend (gfx90a, ROCm 7.1.1)
• Deterministic MI210 reproducer passes with the fix and fails without it
• BLAS functional tests pass on MI210 (740 passed, 0 failed)
• Confirmation on MI210 with AdaptiveCpp, matching the exact error in [BLAS] rocBLAS backend test can fail with HSA_STATUS_ERROR_INVALID_PACKET_FORMAT #486

Made with Cursor