[SYCL][Docs] Fix sycl_ext_oneapi_peer_access implementation and extension #19787

steffenlarsen · 2025-08-13T12:53:39Z

This commit fixes the following bugs in the specification and extension:

Changes the exceptions from repeat calls to enabling and disabling to undefined behavior. The implementation did not properly issue the specified exception before.
Implement the exception from attempting to enable access between devices that do not support peer access between them.
Make the access support query return false for backends that do not support it.
Specify and implement the relaxation that the access can be enabled when both devices are the same, even if the backend doesn't support P2P.
Specify that devices need to be from the same platform, rather than with the same backend.
Add the missing feature test macro.

…sion This commit fixes the following bugs in the specification and extension: * Changes the exceptions from repeat calls to enabling and disabling to undefined behavior. The implementation did not properly issue the specified exception before. * Implement the exception from attempting to enable access between devices that do not support peer access between them. * Make the access support query return false for backends that do not support it. * Specify and implement the relaxation that the access can be enabled when both devices are the same, even if the backend doesn't support P2P. * Specify that devices need to be from the same platform, rather than with the same backend. * Add the missing feature test macro. Signed-off-by: Larsen, Steffen <[email protected]>

steffenlarsen · 2025-08-13T12:54:46Z

sycl/source/device.cpp

@@ -220,6 +220,9 @@ void device::ext_oneapi_enable_peer_access(const device &peer) {
  ur_device_handle_t Device = impl->getHandleRef();
  ur_device_handle_t Peer = peer.impl->getHandleRef();
  if (Device != Peer) {
+    if (!ext_oneapi_can_access_peer(peer))
+      throw sycl::exception(make_error_code(errc::invalid),
+                            "Peer access is not allowed between the devices.");


@gmlueck - If we worry about the additional overhead of doing this check after requiring the user to do it first, we can make it UB instead.

Shouldn't this be feature_not_supported according to the extension spec?

Are we really worried about performance here? It seems like enabling P2P is something the application will just do once in some initialization code.

Shouldn't this be feature_not_supported according to the extension spec?

Ah, right you are!

Are we really worried about performance here? It seems like enabling P2P is something the application will just do once in some initialization code.

I don't particularly worry about the performance, as you are right that it should mainly be a one-and-done call. It is just redundant when the user is doing it correctly, as they are expected to have already checked the call.

Signed-off-by: Larsen, Steffen <[email protected]>

gmlueck · 2025-08-13T13:58:18Z

sycl/doc/extensions/supported/sycl_ext_oneapi_peer_access.asciidoc

+`errc::feature_not_supported` error code.
+
+Calling this function with `peer` for which access has already been enabled will
+result in undefined behavior.


What will our implementation do in this case?

I looked back at the discussion in #6104, and it seems like the CUDA backend will return an error in this case. (Maybe that ends up throwing an exception?) What is the Level Zero behavior?

Same question for the case below when you disable access that was never enabled.

I believe CUDA will cause a backend error and L0 will just allow it through without issue. The issue with trying to make the L0 case return an error is that the tracking of which P2P pathways are enabled may become somewhat costly, while the user could do it themselves if they are concerned with overlapping enabling.

I'd like @pbalcer to weigh in here, because these things will actually go through UR rather than straight to L0.

My understanding is that the L0 adapter in UR will already have to keep track of which P2P pathways are enabled, in order to avoid sharing all USM allocations with all devices (see #19257).

If we have to track it anyway, I have no objections to moving back to having an exception for this.

gmlueck · 2025-08-13T13:59:33Z

sycl/source/device.cpp

@@ -220,6 +220,9 @@ void device::ext_oneapi_enable_peer_access(const device &peer) {
  ur_device_handle_t Device = impl->getHandleRef();
  ur_device_handle_t Peer = peer.impl->getHandleRef();
  if (Device != Peer) {
+    if (!ext_oneapi_can_access_peer(peer))
+      throw sycl::exception(make_error_code(errc::invalid),
+                            "Peer access is not allowed between the devices.");


Shouldn't this be feature_not_supported according to the extension spec?

Are we really worried about performance here? It seems like enabling P2P is something the application will just do once in some initialization code.

Signed-off-by: Larsen, Steffen <[email protected]>

steffenlarsen requested review from a team as code owners August 13, 2025 12:53

steffenlarsen requested a review from sergey-semenov August 13, 2025 12:53

steffenlarsen had a problem deploying to WindowsCILock August 13, 2025 12:53 — with GitHub Actions Error

steffenlarsen commented Aug 13, 2025

View reviewed changes

fix build error

f2d1f8f

Signed-off-by: Larsen, Steffen <[email protected]>

steffenlarsen temporarily deployed to WindowsCILock August 13, 2025 13:08 — with GitHub Actions Inactive

steffenlarsen temporarily deployed to WindowsCILock August 13, 2025 13:43 — with GitHub Actions Inactive

gmlueck reviewed Aug 13, 2025

View reviewed changes

Stop requiring GPU devices

3f2e183

Signed-off-by: Larsen, Steffen <[email protected]>

steffenlarsen temporarily deployed to WindowsCILock August 13, 2025 14:03 — with GitHub Actions Inactive

steffenlarsen temporarily deployed to WindowsCILock August 13, 2025 14:33 — with GitHub Actions Inactive

Fix errc

f13dc2e

Signed-off-by: Larsen, Steffen <[email protected]>

steffenlarsen temporarily deployed to WindowsCILock August 13, 2025 14:45 — with GitHub Actions Inactive

steffenlarsen temporarily deployed to WindowsCILock August 13, 2025 15:11 — with GitHub Actions Inactive

steffenlarsen mentioned this pull request Aug 14, 2025

memory resident limit to enabled peers #19257

Draft

Remove die-call from OpenCL P2P

21c380f

Signed-off-by: Larsen, Steffen <[email protected]>

steffenlarsen requested a review from a team as a code owner August 14, 2025 03:36

steffenlarsen temporarily deployed to WindowsCILock August 14, 2025 03:36 — with GitHub Actions Inactive

steffenlarsen temporarily deployed to WindowsCILock August 14, 2025 04:17 — with GitHub Actions Inactive

kbenzie approved these changes Aug 14, 2025

View reviewed changes

Provide feedback

Saved searches

Use saved searches to filter your results more quickly

Uh oh!

[SYCL][Docs] Fix sycl_ext_oneapi_peer_access implementation and extension #19787

[SYCL][Docs] Fix sycl_ext_oneapi_peer_access implementation and extension #19787

Uh oh!

steffenlarsen commented Aug 13, 2025

Uh oh!

steffenlarsen Aug 13, 2025

Uh oh!

gmlueck Aug 13, 2025

Uh oh!

steffenlarsen Aug 13, 2025

Uh oh!

gmlueck Aug 13, 2025

Uh oh!

steffenlarsen Aug 13, 2025

Uh oh!

Pennycook Aug 13, 2025

Uh oh!

steffenlarsen Aug 14, 2025

Uh oh!

gmlueck Aug 13, 2025

Uh oh!

Uh oh!

[SYCL][Docs] Fix sycl_ext_oneapi_peer_access implementation and extension #19787

Are you sure you want to change the base?

[SYCL][Docs] Fix sycl_ext_oneapi_peer_access implementation and extension #19787

Uh oh!

Conversation

steffenlarsen commented Aug 13, 2025

Uh oh!

Choose a reason for hiding this comment

Uh oh!

Choose a reason for hiding this comment

Uh oh!

Choose a reason for hiding this comment

Uh oh!

Choose a reason for hiding this comment

Uh oh!

Choose a reason for hiding this comment

Uh oh!

Choose a reason for hiding this comment

Uh oh!

Choose a reason for hiding this comment

Uh oh!

Choose a reason for hiding this comment

Uh oh!

Uh oh!