Skip to content

Commit 380453d

Browse files
isaacaulthjabirdDBDuncanprzemektmaloncppchedy
authored
[SYCL][Bindless][2/4] Add experimental implementation of SYCL bindless images extension (#10112)
# Experimental Implementation of SYCL Bindless Images Extension This commit stands as the second commit of four to make code review easier, implementing revision 4 of the [bindless images extension proposal](#9842). ## Scope This PR covers changes made to the PI and the UR. This includes - Extending PI with extension functions - Updating UR FetchContent commit and implementing [UR bindless images experimental features](https://oneapi-src.github.io/unified-runtime/core/EXP-BINDLESS-IMAGES.html) on the CUDA adaptor ## Following Split PRs - [3/4] Implement the user-facing SYCL extension - [4/4] Add tests ## Authors Co-authored-by: Isaac Ault <[email protected]> Co-authored-by: Hugh Bird <[email protected]> Co-authored-by: Duncan Brawley <[email protected]> Co-authored-by: Przemek Malon <[email protected]> Co-authored-by: Chedy Najjar <[email protected]> Co-authored-by: Sean Stirling <[email protected]> Co-authored-by: Peter Zuzek <[email protected]>
1 parent a63f07c commit 380453d

28 files changed

+3448
-490
lines changed

‎sycl/include/sycl/detail/pi.def

Lines changed: 24 additions & 0 deletions
Original file line numberDiff line numberDiff line change
@@ -174,4 +174,28 @@ _PI_API(piextCommandBufferMemBufferRead)
174174
_PI_API(piextCommandBufferMemBufferReadRect)
175175
_PI_API(piextEnqueueCommandBuffer)
176176

177+
_PI_API(piextUSMPitchedAlloc)
178+
179+
// Bindless Images
180+
_PI_API(piextMemUnsampledImageHandleDestroy)
181+
_PI_API(piextMemSampledImageHandleDestroy)
182+
_PI_API(piextBindlessImageSamplerCreate)
183+
_PI_API(piextMemImageAllocate)
184+
_PI_API(piextMemImageFree)
185+
_PI_API(piextMemUnsampledImageCreate)
186+
_PI_API(piextMemSampledImageCreate)
187+
_PI_API(piextMemImageCopy)
188+
_PI_API(piextMemImageGetInfo)
189+
_PI_API(piextMemMipmapGetLevel)
190+
_PI_API(piextMemMipmapFree)
191+
192+
// Interop
193+
_PI_API(piextMemImportOpaqueFD)
194+
_PI_API(piextMemReleaseInterop)
195+
_PI_API(piextMemMapExternalArray)
196+
_PI_API(piextImportExternalSemaphoreOpaqueFD)
197+
_PI_API(piextDestroyExternalSemaphore)
198+
_PI_API(piextWaitExternalSemaphore)
199+
_PI_API(piextSignalExternalSemaphore)
200+
177201
#undef _PI_API

‎sycl/include/sycl/detail/pi.h

Lines changed: 309 additions & 1 deletion
Large diffs are not rendered by default.

‎sycl/include/sycl/detail/pi.hpp

Lines changed: 7 additions & 0 deletions
Original file line numberDiff line numberDiff line change
@@ -156,6 +156,13 @@ using PiExtSyncPoint = ::pi_ext_sync_point;
156156
using PiExtCommandBuffer = ::pi_ext_command_buffer;
157157
using PiExtCommandBufferDesc = ::pi_ext_command_buffer_desc;
158158
using PiPeerAttr = ::pi_peer_attr;
159+
using PiImageHandle = ::pi_image_handle;
160+
using PiImageMemHandle = ::pi_image_mem_handle;
161+
using PiImageCopyFlags = ::pi_image_copy_flags;
162+
using PiInteropMemHandle = ::pi_interop_mem_handle;
163+
using PiInteropSemaphoreHandle = ::pi_interop_semaphore_handle;
164+
using PiImageOffset = ::pi_image_offset_struct;
165+
using PiImageRegion = ::pi_image_region_struct;
159166

160167
__SYCL_EXPORT void contextSetExtendedDeleter(const sycl::context &constext,
161168
pi_context_extended_deleter func,

‎sycl/plugins/cuda/CMakeLists.txt

Lines changed: 2 additions & 0 deletions
Original file line numberDiff line numberDiff line change
@@ -64,6 +64,8 @@ add_sycl_plugin(cuda
6464
"../unified_runtime/ur/adapters/cuda/enqueue.cpp"
6565
"../unified_runtime/ur/adapters/cuda/event.cpp"
6666
"../unified_runtime/ur/adapters/cuda/event.hpp"
67+
"../unified_runtime/ur/adapters/cuda/image.cpp"
68+
"../unified_runtime/ur/adapters/cuda/image.hpp"
6769
"../unified_runtime/ur/adapters/cuda/kernel.cpp"
6870
"../unified_runtime/ur/adapters/cuda/kernel.hpp"
6971
"../unified_runtime/ur/adapters/cuda/memory.cpp"

‎sycl/plugins/cuda/pi_cuda.cpp

Lines changed: 29 additions & 0 deletions
Original file line numberDiff line numberDiff line change
@@ -158,6 +158,7 @@ pi_result piPluginInit(pi_plugin *PluginInit) {
158158
_PI_CL(piextUSMHostAlloc, pi2ur::piextUSMHostAlloc)
159159
_PI_CL(piextUSMDeviceAlloc, pi2ur::piextUSMDeviceAlloc)
160160
_PI_CL(piextUSMSharedAlloc, pi2ur::piextUSMSharedAlloc)
161+
_PI_CL(piextUSMPitchedAlloc, pi2ur::piextUSMPitchedAlloc)
161162
_PI_CL(piextUSMFree, pi2ur::piextUSMFree)
162163
_PI_CL(piextUSMEnqueueMemset, pi2ur::piextUSMEnqueueMemset)
163164
_PI_CL(piextUSMEnqueueMemcpy, pi2ur::piextUSMEnqueueMemcpy)
@@ -196,10 +197,38 @@ pi_result piPluginInit(pi_plugin *PluginInit) {
196197
_PI_CL(piextCommandBufferMemBufferCopyRect,
197198
pi2ur::piextCommandBufferMemBufferCopyRect)
198199
_PI_CL(piextEnqueueCommandBuffer, pi2ur::piextEnqueueCommandBuffer)
200+
199201
// Peer to Peer
200202
_PI_CL(piextEnablePeerAccess, pi2ur::piextEnablePeerAccess)
201203
_PI_CL(piextDisablePeerAccess, pi2ur::piextDisablePeerAccess)
202204
_PI_CL(piextPeerAccessGetInfo, pi2ur::piextPeerAccessGetInfo)
205+
206+
// Bindless Images
207+
_PI_CL(piextMemUnsampledImageHandleDestroy,
208+
pi2ur::piextMemUnsampledImageHandleDestroy)
209+
_PI_CL(piextMemSampledImageHandleDestroy,
210+
pi2ur::piextMemSampledImageHandleDestroy)
211+
_PI_CL(piextMemImageAllocate, pi2ur::piextMemImageAllocate)
212+
_PI_CL(piextMemImageFree, pi2ur::piextMemImageFree)
213+
_PI_CL(piextMemUnsampledImageCreate, pi2ur::piextMemUnsampledImageCreate)
214+
_PI_CL(piextMemSampledImageCreate, pi2ur::piextMemSampledImageCreate)
215+
_PI_CL(piextBindlessImageSamplerCreate,
216+
pi2ur::piextBindlessImageSamplerCreate)
217+
_PI_CL(piextMemImageCopy, pi2ur::piextMemImageCopy)
218+
_PI_CL(piextMemImageGetInfo, pi2ur::piextMemImageGetInfo)
219+
220+
_PI_CL(piextMemMipmapGetLevel, pi2ur::piextMemMipmapGetLevel)
221+
_PI_CL(piextMemMipmapFree, pi2ur::piextMemMipmapFree)
222+
223+
_PI_CL(piextMemImportOpaqueFD, pi2ur::piextMemImportOpaqueFD)
224+
_PI_CL(piextMemReleaseInterop, pi2ur::piextMemReleaseInterop)
225+
_PI_CL(piextMemMapExternalArray, pi2ur::piextMemMapExternalArray)
226+
_PI_CL(piextImportExternalSemaphoreOpaqueFD,
227+
pi2ur::piextImportExternalSemaphoreOpaqueFD)
228+
_PI_CL(piextDestroyExternalSemaphore, pi2ur::piextDestroyExternalSemaphore)
229+
_PI_CL(piextWaitExternalSemaphore, pi2ur::piextWaitExternalSemaphore)
230+
_PI_CL(piextSignalExternalSemaphore, pi2ur::piextSignalExternalSemaphore)
231+
203232
#undef _PI_CL
204233

205234
return PI_SUCCESS;

‎sycl/plugins/esimd_emulator/pi_esimd_emulator.cpp

Lines changed: 102 additions & 0 deletions
Original file line numberDiff line numberDiff line change
@@ -1826,6 +1826,12 @@ pi_result piEnqueueMemImageWrite(pi_queue, pi_mem, pi_bool, pi_image_offset,
18261826
DIE_NO_IMPLEMENTATION;
18271827
}
18281828

1829+
pi_result piextBindlessImageSamplerCreate(pi_context,
1830+
const pi_sampler_properties *, float,
1831+
float, float, pi_sampler *) {
1832+
DIE_NO_IMPLEMENTATION;
1833+
}
1834+
18291835
pi_result piEnqueueMemImageCopy(pi_queue, pi_mem, pi_mem, pi_image_offset,
18301836
pi_image_offset, pi_image_region, pi_uint32,
18311837
const pi_event *, pi_event *) {
@@ -1968,6 +1974,12 @@ pi_result piextUSMSharedAlloc(void **ResultPtr, pi_context Context,
19681974
return PI_SUCCESS;
19691975
}
19701976

1977+
pi_result piextUSMPitchedAlloc(void **, size_t *, pi_context, pi_device,
1978+
pi_usm_mem_properties *, size_t, size_t,
1979+
unsigned int) {
1980+
DIE_NO_IMPLEMENTATION;
1981+
}
1982+
19711983
pi_result piextUSMFree(pi_context Context, void *Ptr) {
19721984
if (Context == nullptr) {
19731985
return PI_ERROR_INVALID_CONTEXT;
@@ -2267,6 +2279,96 @@ pi_result piextPeerAccessGetInfo(pi_device command_device,
22672279
return ReturnValue(pi_int32{0});
22682280
}
22692281

2282+
pi_result piextMemUnsampledImageHandleDestroy(pi_context, pi_device,
2283+
pi_image_handle) {
2284+
DIE_NO_IMPLEMENTATION;
2285+
}
2286+
2287+
pi_result piextMemSampledImageHandleDestroy(pi_context, pi_device,
2288+
pi_image_handle) {
2289+
DIE_NO_IMPLEMENTATION;
2290+
}
2291+
2292+
pi_result piextMemImageAllocate(pi_context, pi_device, pi_image_format *,
2293+
pi_image_desc *, pi_image_mem_handle *) {
2294+
DIE_NO_IMPLEMENTATION;
2295+
}
2296+
2297+
pi_result piextMemMipmapGetLevel(pi_context, pi_device, pi_image_mem_handle,
2298+
unsigned int, pi_image_mem_handle *) {
2299+
DIE_NO_IMPLEMENTATION;
2300+
}
2301+
2302+
pi_result piextMemImageFree(pi_context, pi_device, pi_image_mem_handle) {
2303+
DIE_NO_IMPLEMENTATION;
2304+
}
2305+
2306+
pi_result piextMemMipmapFree(pi_context, pi_device, pi_image_mem_handle) {
2307+
DIE_NO_IMPLEMENTATION;
2308+
}
2309+
2310+
pi_result piextMemUnsampledImageCreate(pi_context, pi_device,
2311+
pi_image_mem_handle, pi_image_format *,
2312+
pi_image_desc *, pi_mem *,
2313+
pi_image_handle *) {
2314+
DIE_NO_IMPLEMENTATION;
2315+
}
2316+
2317+
pi_result piextMemSampledImageCreate(pi_context, pi_device, pi_image_mem_handle,
2318+
pi_image_format *, pi_image_desc *,
2319+
pi_sampler, pi_mem *, pi_image_handle *) {
2320+
DIE_NO_IMPLEMENTATION;
2321+
}
2322+
2323+
pi_result piextMemImageCopy(pi_queue, void *, void *, const pi_image_format *,
2324+
const pi_image_desc *, const pi_image_copy_flags,
2325+
pi_image_offset, pi_image_offset, pi_image_region,
2326+
pi_image_region, pi_uint32, const pi_event *,
2327+
pi_event *) {
2328+
DIE_NO_IMPLEMENTATION;
2329+
}
2330+
2331+
pi_result piextMemImageGetInfo(const pi_image_mem_handle, pi_image_info, void *,
2332+
size_t *) {
2333+
DIE_NO_IMPLEMENTATION;
2334+
}
2335+
2336+
pi_result piextMemImportOpaqueFD(pi_context, pi_device, size_t, int,
2337+
pi_interop_mem_handle *) {
2338+
DIE_NO_IMPLEMENTATION;
2339+
}
2340+
2341+
pi_result piextMemMapExternalArray(pi_context, pi_device, pi_image_format *,
2342+
pi_image_desc *, pi_interop_mem_handle,
2343+
pi_image_mem_handle *) {
2344+
DIE_NO_IMPLEMENTATION;
2345+
}
2346+
2347+
pi_result piextMemReleaseInterop(pi_context, pi_device, pi_interop_mem_handle) {
2348+
DIE_NO_IMPLEMENTATION;
2349+
}
2350+
2351+
pi_result piextImportExternalSemaphoreOpaqueFD(pi_context, pi_device, int,
2352+
pi_interop_semaphore_handle *) {
2353+
DIE_NO_IMPLEMENTATION;
2354+
}
2355+
2356+
pi_result piextDestroyExternalSemaphore(pi_context, pi_device,
2357+
pi_interop_semaphore_handle) {
2358+
DIE_NO_IMPLEMENTATION;
2359+
}
2360+
2361+
pi_result piextWaitExternalSemaphore(pi_queue, pi_interop_semaphore_handle,
2362+
pi_uint32, const pi_event *, pi_event *) {
2363+
DIE_NO_IMPLEMENTATION;
2364+
}
2365+
2366+
pi_result piextSignalExternalSemaphore(pi_queue, pi_interop_semaphore_handle,
2367+
pi_uint32, const pi_event *,
2368+
pi_event *) {
2369+
DIE_NO_IMPLEMENTATION;
2370+
}
2371+
22702372
#ifdef _WIN32
22712373
#define __SYCL_PLUGIN_DLL_NAME "pi_esimd_emulator.dll"
22722374
#include "../common_win_pi_trace/common_win_pi_trace.hpp"

‎sycl/plugins/level_zero/CMakeLists.txt

Lines changed: 4 additions & 2 deletions
Original file line numberDiff line numberDiff line change
@@ -107,8 +107,9 @@ add_sycl_plugin(level_zero
107107
"../unified_runtime/ur/adapters/level_zero/context.hpp"
108108
"../unified_runtime/ur/adapters/level_zero/device.hpp"
109109
"../unified_runtime/ur/adapters/level_zero/event.hpp"
110-
"../unified_runtime/ur/adapters/level_zero/memory.hpp"
110+
"../unified_runtime/ur/adapters/level_zero/image.hpp"
111111
"../unified_runtime/ur/adapters/level_zero/kernel.hpp"
112+
"../unified_runtime/ur/adapters/level_zero/memory.hpp"
112113
"../unified_runtime/ur/adapters/level_zero/platform.hpp"
113114
"../unified_runtime/ur/adapters/level_zero/program.hpp"
114115
"../unified_runtime/ur/adapters/level_zero/queue.hpp"
@@ -120,8 +121,9 @@ add_sycl_plugin(level_zero
120121
"../unified_runtime/ur/adapters/level_zero/context.cpp"
121122
"../unified_runtime/ur/adapters/level_zero/device.cpp"
122123
"../unified_runtime/ur/adapters/level_zero/event.cpp"
123-
"../unified_runtime/ur/adapters/level_zero/memory.cpp"
124+
"../unified_runtime/ur/adapters/level_zero/image.cpp"
124125
"../unified_runtime/ur/adapters/level_zero/kernel.cpp"
126+
"../unified_runtime/ur/adapters/level_zero/memory.cpp"
125127
"../unified_runtime/ur/adapters/level_zero/platform.cpp"
126128
"../unified_runtime/ur/adapters/level_zero/program.cpp"
127129
"../unified_runtime/ur/adapters/level_zero/queue.cpp"

0 commit comments

Comments
 (0)