[SYCL][Bindless] Refactor existing bindless aspects for missing L0 bindless functionality (#14323)

DBDuncan · web-flow · commit 67fa0848e4e6 · 2024-07-23T17:38:08.000Z
Add device aspects query if a device supports image arrays, unique
addressing per dimension and support for sampling 1D usm images.

Remove export semaphore, export image memory and fetch 3D usm image
aspects as they are not supported on any backend.

Update Level Zero and CUDA aspect results to reflect current support
status with latest drivers.
diff --git a/llvm/include/llvm/SYCLLowerIR/DeviceConfigFile.td b/llvm/include/llvm/SYCLLowerIR/DeviceConfigFile.td
@@ -54,20 +54,21 @@ def AspectExt_oneapi_bindless_images_shared_usm : Aspect<"ext_oneapi_bindless_im
 def AspectExt_oneapi_bindless_images_1d_usm : Aspect<"ext_oneapi_bindless_images_1d_usm">;
 def AspectExt_oneapi_bindless_images_2d_usm : Aspect<"ext_oneapi_bindless_images_2d_usm">;
 def AspectExt_oneapi_interop_memory_import : Aspect<"ext_oneapi_interop_memory_import">;
-def AspectExt_oneapi_interop_memory_export : Aspect<"ext_oneapi_interop_memory_export">;
 def AspectExt_oneapi_interop_semaphore_import : Aspect<"ext_oneapi_interop_semaphore_import">;
-def AspectExt_oneapi_interop_semaphore_export : Aspect<"ext_oneapi_interop_semaphore_export">;
 def AspectExt_oneapi_mipmap : Aspect<"ext_oneapi_mipmap">;
 def AspectExt_oneapi_mipmap_anisotropy : Aspect<"ext_oneapi_mipmap_anisotropy">;
 def AspectExt_oneapi_mipmap_level_reference : Aspect<"ext_oneapi_mipmap_level_reference">;
 def AspectExt_oneapi_bindless_sampled_image_fetch_1d_usm : Aspect<"ext_oneapi_bindless_sampled_image_fetch_1d_usm">;
 def AspectExt_oneapi_bindless_sampled_image_fetch_1d : Aspect<"ext_oneapi_bindless_sampled_image_fetch_1d">;
 def AspectExt_oneapi_bindless_sampled_image_fetch_2d_usm : Aspect<"ext_oneapi_bindless_sampled_image_fetch_2d_usm">;
 def AspectExt_oneapi_bindless_sampled_image_fetch_2d : Aspect<"ext_oneapi_bindless_sampled_image_fetch_2d">;
-def AspectExt_oneapi_bindless_sampled_image_fetch_3d_usm : Aspect<"ext_oneapi_bindless_sampled_image_fetch_3d_usm">;
 def AspectExt_oneapi_bindless_sampled_image_fetch_3d : Aspect<"ext_oneapi_bindless_sampled_image_fetch_3d">;
 def AspectExt_oneapi_cubemap : Aspect<"ext_oneapi_cubemap">;
 def AspectExt_oneapi_cubemap_seamless_filtering : Aspect<"ext_oneapi_cubemap_seamless_filtering">;
+def AspectExt_oneapi_image_array : Aspect<"ext_oneapi_image_array">;
+def AspectExt_oneapi_unique_addressing_per_dim : Aspect<"ext_oneapi_unique_addressing_per_dim">;
+def AspectExt_oneapi_bindless_images_sample_1d_usm : Aspect<"ext_oneapi_bindless_images_sample_1d_usm">;
+def AspectExt_oneapi_bindless_images_sample_2d_usm : Aspect<"ext_oneapi_bindless_images_sample_2d_usm">;
 def AspectExt_intel_esimd : Aspect<"ext_intel_esimd">;
 def AspectExt_oneapi_ballot_group : Aspect<"ext_oneapi_ballot_group">;
 def AspectExt_oneapi_fixed_size_group : Aspect<"ext_oneapi_fixed_size_group">;
@@ -128,13 +129,16 @@ def : TargetInfo<"__TestAspectList",
     AspectExt_intel_device_id, AspectExt_intel_memory_clock_rate, AspectExt_intel_memory_bus_width, AspectEmulated,
     AspectExt_intel_legacy_image, AspectExt_oneapi_bindless_images,
     AspectExt_oneapi_bindless_images_shared_usm, AspectExt_oneapi_bindless_images_1d_usm, AspectExt_oneapi_bindless_images_2d_usm,
-    AspectExt_oneapi_interop_memory_import, AspectExt_oneapi_interop_memory_export,
-    AspectExt_oneapi_interop_semaphore_import, AspectExt_oneapi_interop_semaphore_export,
-    AspectExt_oneapi_mipmap, AspectExt_oneapi_mipmap_anisotropy, AspectExt_oneapi_mipmap_level_reference, AspectExt_oneapi_cubemap, 
+    AspectExt_oneapi_interop_memory_import, AspectExt_oneapi_interop_semaphore_import,
+    AspectExt_oneapi_mipmap, AspectExt_oneapi_mipmap_anisotropy, AspectExt_oneapi_mipmap_level_reference,
+    AspectExt_oneapi_bindless_sampled_image_fetch_3d, AspectExt_oneapi_cubemap, 
     AspectExt_oneapi_cubemap_seamless_filtering, 
+    AspectExt_oneapi_image_array,
+    AspectExt_oneapi_unique_addressing_per_dim,
+    AspectExt_oneapi_bindless_images_sample_1d_usm,
+    AspectExt_oneapi_bindless_images_sample_2d_usm,
     AspectExt_oneapi_bindless_sampled_image_fetch_1d_usm, AspectExt_oneapi_bindless_sampled_image_fetch_1d,
     AspectExt_oneapi_bindless_sampled_image_fetch_2d_usm, AspectExt_oneapi_bindless_sampled_image_fetch_2d,
-    AspectExt_oneapi_bindless_sampled_image_fetch_3d_usm, AspectExt_oneapi_bindless_sampled_image_fetch_3d,
     AspectExt_intel_esimd,
     AspectExt_oneapi_ballot_group, AspectExt_oneapi_fixed_size_group, AspectExt_oneapi_opportunistic_group,
     AspectExt_oneapi_tangle_group, AspectExt_intel_matrix, AspectExt_oneapi_is_composite, AspectExt_oneapi_is_component, 
@@ -216,7 +220,9 @@ defvar CudaMinAspects = !listconcat(AllUSMAspects, [AspectGpu, AspectFp64, Aspec
 defvar CudaBindlessImagesAspects = [AspectExt_oneapi_bindless_images, AspectExt_oneapi_bindless_images_shared_usm,
     AspectExt_oneapi_bindless_images_1d_usm, AspectExt_oneapi_bindless_images_2d_usm, AspectExt_oneapi_interop_memory_import,
     AspectExt_oneapi_interop_semaphore_import, AspectExt_oneapi_mipmap, AspectExt_oneapi_mipmap_anisotropy,
-    AspectExt_oneapi_mipmap_level_reference, AspectExt_oneapi_cubemap, AspectExt_oneapi_cubemap_seamless_filtering];
+    AspectExt_oneapi_mipmap_level_reference, AspectExt_oneapi_cubemap, AspectExt_oneapi_cubemap_seamless_filtering,
+    AspectExt_oneapi_image_array, AspectExt_oneapi_unique_addressing_per_dim, AspectExt_oneapi_bindless_images_sample_2d_usm,
+    AspectExt_oneapi_bindless_images_sample_2d_usm];
 
 def : CudaTargetInfo<"nvidia_gpu_sm_50", !listconcat(CudaMinAspects, CudaBindlessImagesAspects)>;
 def : CudaTargetInfo<"nvidia_gpu_sm_52", !listconcat(CudaMinAspects, CudaBindlessImagesAspects)>;
diff --git a/sycl/doc/extensions/experimental/sycl_ext_oneapi_bindless_images.asciidoc b/sycl/doc/extensions/experimental/sycl_ext_oneapi_bindless_images.asciidoc
@@ -135,8 +135,8 @@ The device aspects for these queries are:
 |======================
 |Device descriptor |Description
 |`aspect::ext_oneapi_bindless_images` | Indicates if the device supports 
-creation of bindless images backed by the `image_mem` and `image_mem_handle` 
-APIs.
+bindless images. This includes creating bindless images backed by the
+`image_mem` and `image_mem_handle` APIs.
 |`aspect::ext_oneapi_bindless_images_shared_usm` | Indicates if the device 
 supports the creation of bindless images backed by shared USM memory.
 |`aspect::ext_oneapi_bindless_images_1d_usm` | Indicates if the device supports 
@@ -714,6 +714,8 @@ represents the backend's default addressing mode. On CUDA this is `Wrap`, i.e.
 `addressing[3]` defines the addressing mode per texture dimension. A 
 `bindless_image_sampler` can be constructed with a singular 
 `sycl::addressing_mode`, where this parameter will define all dimensions.
+Not all devices may support unique addressing per dimension.  We provide device
+aspect queries for this in <<querying_unique_addressing_support>>
 
 `mipmap_filtering` dictates the method in which sampling between mipmap 
 levels is performed.
@@ -1069,6 +1071,9 @@ not all devices may support fetching of sampled image data depending on the
 dimension or backing memory type. We provide device aspect queries for this in 
 <<querying_sampled_image_fetch_support>>.
 
+Additionally, not all devices may support sampling of USM images. We provide
+device aspect queries for this in <<querying_usm_sample_support>>
+
 The user is required to pass a `DataT` template parameter, which specifies the
 return type of the `fetch_image` and `sample_image` functions. If `DataT` is 
 not a recognized standard type, as defined in <<recognized_standard_types>>, 
@@ -1171,14 +1176,41 @@ The device aspect descriptors for these queries are:
 |`aspect::ext_oneapi_bindless_sampled_image_fetch_2d` | 
  Indicates if the device is capable of fetching non-USM backed 2D 
  sampled image data.
-|`aspect::ext_oneapi_bindless_sampled_image_fetch_3d_usm` | 
- Indicates if the device is capable of fetching USM backed 3D 
- sampled image data.
 |`aspect::ext_oneapi_bindless_sampled_image_fetch_3d` | 
  Indicates if the device is capable of fetching non-USM backed 3D 
  sampled image data.
 |======================
 
+=== Querying USM sampling support [[querying_usm_sample_support]]
+
+We provide the following device queries to query support for sampling USM
+images.
+
+The device aspect descriptors for these queries are:
+
+[frame="none",options="header"]
+|======================
+|Device descriptor | Description
+|`aspect::ext_oneapi_bindless_images_sample_1d_usm` | Indicates if the device
+supports the sampling of 1D bindless images backed by USM.
+|`aspect::ext_oneapi_bindless_images_sample_2d_usm` | Indicates if the device
+supports the sampling of 2D bindless images backed by USM.
+|======================
+
+=== Querying unique addressing support [[querying_unique_addressing_support]]
+
+We provide the following device queries to query support information for
+unique addressing for each image dimension.
+
+The device aspect descriptor for this query is:
+
+[frame="none",options="header"]
+|======================
+|Device descriptor |Description
+|`aspect::ext_oneapi_unique_addressing_per_dim` | Indicates if the device 
+supports unique addressing per dimension when sampling.
+|======================
+
 == Mipmapped images
 
 So far, we have described how to create and operate on standard bindless images.
@@ -1312,6 +1344,20 @@ every index has the same dimensionality, size, and data type.
 Image arrays may also be referred to as layered images, and the array indices 
 may be referred to layers.
 
+=== Querying image array support
+
+We provide the following device aspect to retrieve support information for
+image arrays.
+
+The device aspect descriptor for this query is:
+
+[frame="none",options="header"]
+|======================
+|Device descriptor |Description
+|`aspect::ext_oneapi_image_array` | Indicates if the device supports
+image arrays.
+|======================
+
 === Allocation of image arrays
 
 Image arrays are allocated in a similar manner to standard images.
@@ -1547,12 +1593,8 @@ The device aspect descriptors for these queries are:
 |Device descriptor |Description
 |`aspect::ext_oneapi_interop_memory_import` | Indicates if the device supports 
 importing external memory resources.
-|`aspect::ext_oneapi_interop_memory_export` | Indicates if the device supports 
-exporting internal memory resources.
 |`aspect::ext_oneapi_interop_semaphore_import`` | Indicates if the device 
 supports importing external semaphore resources.
-|`aspect::ext_oneapi_interop_semaphore_export` | Indicates if the device 
-supports exporting internal event resources.
 |======================
 
 
@@ -2811,4 +2853,10 @@ These features still need to be handled:
                     funcs.
 |5.14|2024-07-17| - Rename `destroy_external_semaphore` to 
                     `release_external_semaphore`.
+|5.15|2024-07-19 - Add missing device queries for image arrays, sampling USM 
+                   images and unique addressing per dimension.
+                 - Remove aspects for semaphore export, memory export and fetch 
+                   3D USM images as they are not supported on any platform.
+                 - Refine the description of `ext_oneapi_bindless_images` aspect
+                   to indicate support for bindless image APIs.
 |======================
diff --git a/sycl/include/sycl/detail/pi.h b/sycl/include/sycl/detail/pi.h
@@ -206,9 +206,18 @@
 // 19.61 Rename piextDestroyExternalSemaphore to piextReleaseExternalSemaphore
 // 20.62 Changed the signature of piextMemImageCopy to take 2 image and format
 //       descriptors.
+// 20.63 Added device queries
+//         - PI_EXT_ONEAPI_DEVICE_INFO_IMAGE_ARRAY_SUPPORT
+//         - PI_EXT_ONEAPI_DEVICE_INFO_BINDLESS_UNIQUE_ADDRESSING_PER_DIM
+//         - PI_EXT_ONEAPI_DEVICE_INFO_BINDLESS_SAMPLE_1D_USM
+//         - PI_EXT_ONEAPI_DEVICE_INFO_BINDLESS_SAMPLE_2D_USM
+//       Removed device queries
+//         - PI_EXT_ONEAPI_DEVICE_INFO_INTEROP_MEMORY_EXPORT_SUPPORT
+//         - PI_EXT_ONEAPI_DEVICE_INFO_INTEROP_SEMAPHORE_EXPORT_SUPPORT
+//         - PI_EXT_ONEAPI_DEVICE_INFO_BINDLESS_SAMPLED_IMAGE_FETCH_3D_USM
 
 #define _PI_H_VERSION_MAJOR 20
-#define _PI_H_VERSION_MINOR 62
+#define _PI_H_VERSION_MINOR 63
 
 #define _PI_STRING_HELPER(a) #a
 #define _PI_CONCAT(a, b) _PI_STRING_HELPER(a.b)
@@ -492,9 +501,7 @@ typedef enum {
   PI_EXT_ONEAPI_DEVICE_INFO_MIPMAP_MAX_ANISOTROPY = 0x2010A,
   PI_EXT_ONEAPI_DEVICE_INFO_MIPMAP_LEVEL_REFERENCE_SUPPORT = 0x2010B,
   PI_EXT_ONEAPI_DEVICE_INFO_INTEROP_MEMORY_IMPORT_SUPPORT = 0x2010C,
-  PI_EXT_ONEAPI_DEVICE_INFO_INTEROP_MEMORY_EXPORT_SUPPORT = 0x2010D,
   PI_EXT_ONEAPI_DEVICE_INFO_INTEROP_SEMAPHORE_IMPORT_SUPPORT = 0x2010E,
-  PI_EXT_ONEAPI_DEVICE_INFO_INTEROP_SEMAPHORE_EXPORT_SUPPORT = 0x2010F,
 
   PI_EXT_ONEAPI_DEVICE_INFO_MATRIX_COMBINATIONS = 0x20110,
 
@@ -515,7 +522,6 @@ typedef enum {
   PI_EXT_ONEAPI_DEVICE_INFO_BINDLESS_SAMPLED_IMAGE_FETCH_1D = 0x20118,
   PI_EXT_ONEAPI_DEVICE_INFO_BINDLESS_SAMPLED_IMAGE_FETCH_2D_USM = 0x20119,
   PI_EXT_ONEAPI_DEVICE_INFO_BINDLESS_SAMPLED_IMAGE_FETCH_2D = 0x2011A,
-  PI_EXT_ONEAPI_DEVICE_INFO_BINDLESS_SAMPLED_IMAGE_FETCH_3D_USM = 0x2011B,
   PI_EXT_ONEAPI_DEVICE_INFO_BINDLESS_SAMPLED_IMAGE_FETCH_3D = 0x2011C,
 
   // Timestamp enqueue
@@ -529,6 +535,12 @@ typedef enum {
 
   // Return whether cluster launch is supported by device
   PI_EXT_ONEAPI_DEVICE_INFO_CLUSTER_LAUNCH = 0x2021,
+
+  // Bindless image arrays, unique addressing and USM sampling
+  PI_EXT_ONEAPI_DEVICE_INFO_IMAGE_ARRAY_SUPPORT = 0x20122,
+  PI_EXT_ONEAPI_DEVICE_INFO_BINDLESS_UNIQUE_ADDRESSING_PER_DIM = 0x20123,
+  PI_EXT_ONEAPI_DEVICE_INFO_BINDLESS_SAMPLE_1D_USM = 0x20124,
+  PI_EXT_ONEAPI_DEVICE_INFO_BINDLESS_SAMPLE_2D_USM = 0x20125,
 } _pi_device_info;
 
 typedef enum {
diff --git a/sycl/include/sycl/device_aspect_macros.hpp b/sycl/include/sycl/device_aspect_macros.hpp
@@ -238,21 +238,11 @@
 #define __SYCL_ALL_DEVICES_HAVE_ext_oneapi_interop_memory_import__ 0
 #endif
 
-#ifndef __SYCL_ALL_DEVICES_HAVE_ext_oneapi_interop_memory_export__
-//__SYCL_ASPECT(ext_oneapi_interop_memory_export, 47)
-#define __SYCL_ALL_DEVICES_HAVE_ext_oneapi_interop_memory_export__ 0
-#endif
-
 #ifndef __SYCL_ALL_DEVICES_HAVE_ext_oneapi_interop_semaphore_import__
 //__SYCL_ASPECT(ext_oneapi_interop_semaphore_import, 48)
 #define __SYCL_ALL_DEVICES_HAVE_ext_oneapi_interop_semaphore_import__ 0
 #endif
 
-#ifndef __SYCL_ALL_DEVICES_HAVE_ext_oneapi_interop_semaphore_export__
-//__SYCL_ASPECT(ext_oneapi_interop_semaphore_export, 49)
-#define __SYCL_ALL_DEVICES_HAVE_ext_oneapi_interop_semaphore_export__ 0
-#endif
-
 #ifndef __SYCL_ALL_DEVICES_HAVE_ext_oneapi_mipmap__
 //__SYCL_ASPECT(ext_oneapi_mipmap, 50)
 #define __SYCL_ALL_DEVICES_HAVE_ext_oneapi_mipmap__ 0
@@ -355,12 +345,6 @@
 #define __SYCL_ALL_DEVICES_HAVE_ext_oneapi_bindless_sampled_image_fetch_2d__ 0
 #endif
 
-#ifndef __SYCL_ALL_DEVICES_HAVE_ext_oneapi_bindless_sampled_image_fetch_3d_usm__
-//__SYCL_ASPECT(ext_oneapi_bindless_sampled_image_fetch_3d_usm, 71)
-#define __SYCL_ALL_DEVICES_HAVE_ext_oneapi_bindless_sampled_image_fetch_3d_usm__ \
-  0
-#endif
-
 #ifndef __SYCL_ALL_DEVICES_HAVE_ext_oneapi_bindless_sampled_image_fetch_3d__
 //__SYCL_ASPECT(ext_oneapi_bindless_sampled_image_fetch_3d, 72)
 #define __SYCL_ALL_DEVICES_HAVE_ext_oneapi_bindless_sampled_image_fetch_3d__ 0
@@ -381,6 +365,26 @@
 #define __SYCL_ALL_DEVICES_HAVE_ext_oneapi_cuda_cluster_group__ 0
 #endif
 
+#ifndef __SYCL_ALL_DEVICES_HAVE_ext_oneapi_image_array__
+//__SYCL_ASPECT(ext_oneapi_image_array, 76)
+#define __SYCL_ALL_DEVICES_HAVE_ext_oneapi_image_array__ 0
+#endif
+
+#ifndef __SYCL_ALL_DEVICES_HAVE_ext_oneapi_unique_addressing_per_dim__
+//__SYCL_ASPECT(ext_oneapi_unique_addressing_per_dim, 77)
+#define __SYCL_ALL_DEVICES_HAVE_ext_oneapi_unique_addressing_per_dim__ 0
+#endif
+
+#ifndef __SYCL_ALL_DEVICES_HAVE_ext_oneapi_bindless_images_sample_1d_usm__
+//__SYCL_ASPECT(ext_oneapi_bindless_images_sample_1d_usm, 78)
+#define __SYCL_ALL_DEVICES_HAVE_ext_oneapi_bindless_images_sample_1d_usm__ 0
+#endif
+
+#ifndef __SYCL_ALL_DEVICES_HAVE_ext_oneapi_bindless_images_sample_2d_usm__
+//__SYCL_ASPECT(ext_oneapi_bindless_images_sample_2d_usm, 79)
+#define __SYCL_ALL_DEVICES_HAVE_ext_oneapi_bindless_images_sample_2d_usm__ 0
+#endif
+
 #ifndef __SYCL_ANY_DEVICE_HAS_host__
 // __SYCL_ASPECT(host, 0)
 #define __SYCL_ANY_DEVICE_HAS_host__ 0
@@ -611,21 +615,11 @@
 #define __SYCL_ANY_DEVICE_HAS_ext_oneapi_interop_memory_import__ 0
 #endif
 
-#ifndef __SYCL_ANY_DEVICE_HAS_ext_oneapi_interop_memory_export__
-//__SYCL_ASPECT(ext_oneapi_interop_memory_export, 47)
-#define __SYCL_ANY_DEVICE_HAS_ext_oneapi_interop_memory_export__ 0
-#endif
-
 #ifndef __SYCL_ANY_DEVICE_HAS_ext_oneapi_interop_semaphore_import__
 //__SYCL_ASPECT(ext_oneapi_interop_semaphore_import, 48)
 #define __SYCL_ANY_DEVICE_HAS_ext_oneapi_interop_semaphore_import__ 0
 #endif
 
-#ifndef __SYCL_ANY_DEVICE_HAS_ext_oneapi_interop_semaphore_export__
-//__SYCL_ASPECT(ext_oneapi_interop_semaphore_export, 49)
-#define __SYCL_ANY_DEVICE_HAS_ext_oneapi_interop_semaphore_export__ 0
-#endif
-
 #ifndef __SYCL_ANY_DEVICE_HAS_ext_oneapi_mipmap__
 //__SYCL_ASPECT(ext_oneapi_mipmap, 50)
 #define __SYCL_ANY_DEVICE_HAS_ext_oneapi_mipmap__ 0
@@ -726,11 +720,6 @@
 #define __SYCL_ANY_DEVICE_HAS_ext_oneapi_bindless_sampled_image_fetch_2d__ 0
 #endif
 
-#ifndef __SYCL_ANY_DEVICE_HAS_ext_oneapi_bindless_sampled_image_fetch_3d_usm__
-//__SYCL_ASPECT(ext_oneapi_bindless_sampled_image_fetch_3d_usm, 71)
-#define __SYCL_ANY_DEVICE_HAS_ext_oneapi_bindless_sampled_image_fetch_3d_usm__ 0
-#endif
-
 #ifndef __SYCL_ANY_DEVICE_HAS_ext_oneapi_bindless_sampled_image_fetch_3d__
 //__SYCL_ASPECT(ext_oneapi_bindless_sampled_image_fetch_3d, 72)
 #define __SYCL_ANY_DEVICE_HAS_ext_oneapi_bindless_sampled_image_fetch_3d__ 0
@@ -750,3 +739,23 @@
 // __SYCL_ASPECT(ext_oneapi_cuda_cluster_group, 75)
 #define __SYCL_ANY_DEVICE_HAS_ext_oneapi_cuda_cluster_group__ 0
 #endif
+
+#ifndef __SYCL_ANY_DEVICE_HAS_ext_oneapi_image_array__
+//__SYCL_ASPECT(ext_oneapi_image_array, 76)
+#define __SYCL_ANY_DEVICE_HAS_ext_oneapi_image_array__ 0
+#endif
+
+#ifndef __SYCL_ANY_DEVICE_HAS_ext_oneapi_unique_addressing_per_dim__
+//__SYCL_ASPECT(ext_oneapi_unique_addressing_per_dim, 77)
+#define __SYCL_ANY_DEVICE_HAS_ext_oneapi_unique_addressing_per_dim__ 0
+#endif
+
+#ifndef __SYCL_ANY_DEVICE_HAS_ext_oneapi_bindless_images_sample_1d_usm__
+//__SYCL_ASPECT(ext_oneapi_bindless_images_sample_1d_usm, 78)
+#define __SYCL_ANY_DEVICE_HAS_ext_oneapi_bindless_images_sample_1d_usm__ 0
+#endif
+
+#ifndef __SYCL_ANY_DEVICE_HAS_ext_oneapi_bindless_images_sample_2d_usm__
+//__SYCL_ASPECT(ext_oneapi_bindless_images_sample_2d_usm, 79)
+#define __SYCL_ANY_DEVICE_HAS_ext_oneapi_bindless_images_sample_2d_usm__ 0
+#endif
diff --git a/sycl/include/sycl/info/aspects.def b/sycl/include/sycl/info/aspects.def
@@ -40,9 +40,7 @@ __SYCL_ASPECT(ext_oneapi_bindless_images_shared_usm, 43)
 __SYCL_ASPECT(ext_oneapi_bindless_images_1d_usm, 44)
 __SYCL_ASPECT(ext_oneapi_bindless_images_2d_usm, 45)
 __SYCL_ASPECT(ext_oneapi_interop_memory_import, 46)
-__SYCL_ASPECT(ext_oneapi_interop_memory_export, 47)
 __SYCL_ASPECT(ext_oneapi_interop_semaphore_import, 48)
-__SYCL_ASPECT(ext_oneapi_interop_semaphore_export, 49)
 __SYCL_ASPECT(ext_oneapi_mipmap, 50)
 __SYCL_ASPECT(ext_oneapi_mipmap_anisotropy, 51)
 __SYCL_ASPECT(ext_oneapi_mipmap_level_reference, 52)
@@ -63,8 +61,11 @@ __SYCL_ASPECT(ext_oneapi_bindless_sampled_image_fetch_1d_usm, 67)
 __SYCL_ASPECT(ext_oneapi_bindless_sampled_image_fetch_1d, 68)
 __SYCL_ASPECT(ext_oneapi_bindless_sampled_image_fetch_2d_usm, 69)
 __SYCL_ASPECT(ext_oneapi_bindless_sampled_image_fetch_2d, 70)
-__SYCL_ASPECT(ext_oneapi_bindless_sampled_image_fetch_3d_usm, 71)
 __SYCL_ASPECT(ext_oneapi_bindless_sampled_image_fetch_3d, 72)
 __SYCL_ASPECT(ext_oneapi_queue_profiling_tag, 73)
 __SYCL_ASPECT(ext_oneapi_virtual_mem, 74)
 __SYCL_ASPECT(ext_oneapi_cuda_cluster_group, 75)
+__SYCL_ASPECT(ext_oneapi_image_array, 76)
+__SYCL_ASPECT(ext_oneapi_unique_addressing_per_dim, 77)
+__SYCL_ASPECT(ext_oneapi_bindless_images_sample_1d_usm, 78)
+__SYCL_ASPECT(ext_oneapi_bindless_images_sample_2d_usm, 79)
diff --git a/sycl/plugins/unified_runtime/CMakeLists.txt b/sycl/plugins/unified_runtime/CMakeLists.txt
@@ -100,12 +100,12 @@ if(SYCL_PI_UR_USE_FETCH_CONTENT)
   endfunction()
 
   set(UNIFIED_RUNTIME_REPO "https://github.com/oneapi-src/unified-runtime.git")
-  # commit d2b086a7170145f255e8702f512f6157b047cbe5
-  # Author: Callum Fare <callum@codeplay.com>
-  # Date:   Mon Jul 22 11:11:32 2024 +0100
-  #     Merge pull request #1803 from cppchedy/chedy/new-ur-copy-api
-  #     [CUDA][L0][Bindless] update design of urBindlessImagesImageCopyExp
-  set(UNIFIED_RUNTIME_TAG d2b086a7170145f255e8702f512f6157b047cbe5)
+  # commit f11caf98857bb22800d09aedec8bd607643ab67b
+  # Author: Kenneth Benzie (Benie) <k.benzie@codeplay.com>
+  # Date:   Tue Jul 23 15:28:10 2024 +0100
+  #     Merge pull request #1795 from DBDuncan/duncan/bindless-aspects
+  #     [Bindless][Exp] Update device_info enum elements and update Level Zero and CUDA backend device_info supported features
+  set(UNIFIED_RUNTIME_TAG f11caf98857bb22800d09aedec8bd607643ab67b)
 
   fetch_adapter_source(level_zero
     ${UNIFIED_RUNTIME_REPO}
diff --git a/sycl/plugins/unified_runtime/pi2ur.hpp b/sycl/plugins/unified_runtime/pi2ur.hpp
diff --git a/sycl/source/detail/device_impl.cpp b/sycl/source/detail/device_impl.cpp
diff --git a/sycl/test-e2e/bindless_images/image_get_info.cpp b/sycl/test-e2e/bindless_images/image_get_info.cpp
