intel / intel/llvm

[SYCL][UR] pool allocators bypass peer-access API

Open
#21,395 3 comments 0 reactions 0 assignees View on GitHub
bug unified-runtime
Dominant language
LLVM
Stars
1.5k
Forks
854
Avg merge
3d 17h
Merged PRs (30d)
137

Description

### Describe the bug

We have an API extension for whether SYCL GPUs may or may not have peer access defined at https://github.com/intel/llvm/blob/sycl/sycl/doc/extensions/supported/sycl_ext_oneapi_peer_access.asciidoc.

Recently I'd been trying to do just that without knowing about (or using) that API extension. I made a SYCL context across multiple BMG-G31 GPUs and ran into various problems (CMPLRLLVM-73510, URLZA-722). While tracing those issues, it became clear that creating the `sycl::context` objects triggered the creation of pool allocators and what looks like some test allocations. By default, the UR device pool allocator makes all device allocations resident on all devices in the `sycl::context`. I'm not sure what the definition of "resident" is, but the implementation looks like it prepares virtual memory but stops short of allocating a physical page.

I think this behaviour is undesirable in the absence of the user using the API extension for peer access to declare that peer access is desired (and between which GPUs). The existence of multiply-resident allocations across by default all devices in the `sycl::context` may have subtle costs at run time. C++ users expect not to pay for things that they are not using. UR pool allocators should by default respect the explicit declarations of peer access that have or have not been made.

### To reproduce

To reproduce, make a sycl::context over all GPUs on a node and trace, noting many calls to `zeContextMakeMemoryResident`

### Environment

2025.3.0 using L0 and a node with multiple BMG cards

### Additional context

_No response_

Contributor guide

Open the contributing guide

Assessment

This issue has not been assessed yet.

Get new issues in your inbox

A short digest of beginner-friendly GitHub issues.