[SYCL][HIP][CUDA] Report errc::nd_range for oversized work-group size

[zjin-lcf]

Summary

• Throw a proper errc::nd_range exception when an oversized work-group size is requested on HIP/CUDA, instead of a generic/opaque failure.
• Re-enable Regression/invalid_reqd_wg_size_correct_exception on HIP.

Test plan

• Basic/gpu_max_wgs_error passes on gfx942
• Regression/invalid_reqd_wg_size_correct_exception passes on gfx942

Relates to #22300

[Copilot]

Pull request overview

This PR improves SYCL runtime diagnostics for oversized work-group sizes on the CUDA/HIP backends by translating the underlying launch failure into a consistent sycl::errc::nd_range exception (matching the existing OpenCL behavior), and re-enables a previously-disabled HIP regression test.

Changes:

• Add explicit CUDA/HIP handling in handleInvalidWorkGroupSize() to throw errc::nd_range when local_size total exceeds UR_DEVICE_INFO_MAX_WORK_GROUP_SIZE.
• Re-enable Regression/invalid_reqd_wg_size_correct_exception on HIP by removing UNSUPPORTED: hip.
• Re-enable Basic/gpu_max_wgs_error on HIP by removing UNSUPPORTED: hip.

Reviewed changes

Copilot reviewed 3 out of 3 changed files in this pull request and generated 1 comment.

File	Description
sycl/test-e2e/Regression/invalid_reqd_wg_size_correct_exception.cpp	Re-enables the regression test on HIP.
sycl/test-e2e/Basic/gpu_max_wgs_error.cpp	Re-enables the work-group-size diagnostic test on HIP (still disabled on CUDA).
sycl/source/detail/error_handling/error_handling.cpp	Adds CUDA/HIP-specific translation of oversized work-group size failures into errc::nd_range.

[zjin-lcf]

Not sure Copilot can continue the review.

Do you recommend the Runtime fix below ?

In handleOutOfResources(), inside the existing CUDA branch, check the work-group total against the device max before emitting the register
message:

  if (Backend == sycl::backend::ext_oneapi_cuda) {
    const size_t TotalNumberOfWIs = /* ... */;
    size_t MaxWGSize = 0;
    Adapter.call<UrApiKind::urDeviceGetInfo>(
        Device, UR_DEVICE_INFO_MAX_WORK_GROUP_SIZE, sizeof(size_t),
        &MaxWGSize, nullptr);
    if (TotalNumberOfWIs > MaxWGSize)
      throw sycl::exception(
          make_error_code(errc::nd_range),
          "Total number of work-items in a work-group cannot exceed " +
              std::to_string(MaxWGSize));
    // ... existing register-exhaustion check (only a real register issue now) ...
  }

• The driver returns OUT_OF_RESOURCES when registers are exhausted and INVALID_WORK_GROUP_SIZE otherwise, so the "too many work-items" check now needs to live in both handlers. The PR already added it to handleInvalidWorkGroupSize(); this adds the symmetric case.
• Result: CUDA emits the exact message the test asserts for the max-size input, so you just delete // UNSUPPORTED: cuda. Cleanest, gives consistent cross-backend diagnostics.