intel / intel/llvm

[SYCL][USM] Trying to use malloc_shared does not raise an exception, but hits an assert instead

Open
#11,747 1 comment 0 reactions 0 assignees View on GitHub
bug confirmed
Dominant language
LLVM
Stars
1.5k
Forks
854
Avg merge
3d 17h
Merged PRs (30d)
137

Description

**Describe the bug**
https://registry.khronos.org/SYCL/specs/sycl-2020/html/sycl-2020.html#sec:usm has the following example SYCL code to demonstrate USM:

```c++
#include
#include
using namespace sycl; // (optional) avoids need for "sycl::" before SYCL names

int main() {
// Create a default queue to enqueue work to the default device
queue myQueue;

// Allocate shared memory bound to the device and context associated to the
// queue Replacing malloc_shared with malloc_host would yield a correct
// program that allocated device-visible memory on the host.
int* data = sycl::malloc_shared(1024, myQueue);

myQueue.parallel_for(1024, [=](id<1> idx) {
// Initialize each buffer element with its own rank number starting at 0
data[idx] = idx;
}); // End of the kernel function

// Explicitly wait for kernel execution since there is no accessor involved
myQueue.wait();

// Print result
for (int i = 0; i < 1024; i++)
std::cout << "data[" << i << "] = " << data[i] << std::endl;

return 0;
}
```

According to https://registry.khronos.org/SYCL/specs/sycl-2020/html/sycl-2020.html#_shared_allocation_functions malloc_shared requires a specific aspect, and should throw a `feature_not_supported` if that feature isn't available.

If the above program runs with a device which does not have the `usm_shared_allocations` aspect and doesn't provide the appropriate functions in the UR interface, then the following assert is hit instead:
```
ur_result_t urUSMSharedAlloc(ur_context_handle_t, ur_device_handle_t, const ur_usm_desc_t*, ur_usm_pool_handle_t, size_t, void**): Assertion `Alignment == 0 || (RetVal == UR_RESULT_SUCCESS && reinterpret_cast(*ppMem) % Alignment == 0)' failed.
```

While this assert is in the unified runtime layer, I think the SYCL runtime should have detected lack of support and thrown an exception before attempting to call out to it.

Looking at `alignedAllocInternal` in `usm_impl.cpp` I think this might also apply to device allocations as well, but I've not tested that.

**To Reproduce**
1. Find a device which does not support malloc_shared and doesn't have the usm_shared_allocations aspect.
2. Build the sample using: `bin/clang++ -fsycl test.cpp -o test_binary`
3. Run the sample using that device.
4. I expect to see an exception being raised (or the program terminating due to an uncaught exception), not an assert being hit.

**Environment (please complete the following information):**

- OS: Linux Mint 21.1 (Based on Ubuntu 22.04)
- Target device and vendor: Custom OpenCL runtime without shared allocation support. This device correctly does not report the usm_shared_allocations aspect (`device.has(aspect::usm_shared_allocations)` returns false)
- DPC++ version: 22c9574b9d4f2803e249c7d416fddaeb88aa8f47 built in debug mode.
- Dependencies version: N/A

**Additional context**
We've seen this when running the SYCL-CTS test suite, but in all cases I've seen it's due to the CTS tests not properly doing feature checks before running the test.

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.