[Bindless][UR][E2E] Fix/improve Vulkan tests. Fix L0 BI aspect query.

sycl/test-e2e/bindless_images/vulkan_interop/buffer_usm.cpp

@@ -2,7 +2,7 @@
 // REQUIRES: vulkan
 
 // RUN: %{build} %link-vulkan -o %t.out %if target-spir %{ -Wno-ignored-attributes %}
-// RUN: %{run} env NEOReadDebugKeys=1 UseBindlessMode=1 UseExternalAllocatorForSshAndDsh=1 %t.out
+// RUN: %{run} %t.out
 
 /**
  * This test does not use any image specific APIs.

sycl/test-e2e/bindless_images/vulkan_interop/depth_format.cpp

@@ -1,3 +1,4 @@
+// REQUIRES: aspect-ext_oneapi_bindless_images
 // REQUIRES: aspect-ext_oneapi_external_memory_import || (windows && level_zero && aspect-ext_oneapi_bindless_images)
 // REQUIRES: vulkan
 
@@ -128,9 +129,8 @@ bool runTest(const sycl::device &syclDevice, sycl::range<2> dims,
   {
     vkInputImage = vkutil::createImage(imgType, imgInFormat, imgExtent,
                                        VK_IMAGE_USAGE_TRANSFER_SRC_BIT |
-                                           VK_IMAGE_USAGE_TRANSFER_DST_BIT |
-                                           VK_IMAGE_USAGE_STORAGE_BIT,
-                                       1);
+                                           VK_IMAGE_USAGE_TRANSFER_DST_BIT,
+                                       1 /*mipLevels*/);
     VkMemoryRequirements memRequirements;
     auto inputImageMemoryTypeIndex = vkutil::getImageMemoryTypeIndex(
         vkInputImage, VK_MEMORY_PROPERTY_DEVICE_LOCAL_BIT, memRequirements);
@@ -141,9 +141,8 @@ bool runTest(const sycl::device &syclDevice, sycl::range<2> dims,
 
     vkOutputImage = vkutil::createImage(imgType, imgOutFormat, imgExtent,
                                         VK_IMAGE_USAGE_TRANSFER_SRC_BIT |
-                                            VK_IMAGE_USAGE_TRANSFER_DST_BIT |
-                                            VK_IMAGE_USAGE_STORAGE_BIT,
-                                        1);
+                                            VK_IMAGE_USAGE_TRANSFER_DST_BIT,
+                                        1 /*mipLevels*/);
     VkMemoryRequirements outputMemRequirements;
     auto outputImageMemoryTypeIndex = vkutil::getImageMemoryTypeIndex(
         vkOutputImage, VK_MEMORY_PROPERTY_DEVICE_LOCAL_BIT,
@@ -373,7 +372,7 @@ int main() {
     return EXIT_FAILURE;
   }
 
-  auto testPassed = runTest(syclDevice, {16, 16}, {16, 16});
+  auto testPassed = runTest(syclDevice, {128, 128}, {16, 16});
 
   if (vkutil::cleanup() != VK_SUCCESS) {
     std::cerr << "Cleanup failed!\n";

sycl/test-e2e/bindless_images/vulkan_interop/mipmaps.cpp

@@ -1,3 +1,4 @@
+// REQUIRES: aspect-ext_oneapi_bindless_images
 // REQUIRES: aspect-ext_oneapi_external_memory_import
 // REQUIRES: aspect-ext_oneapi_mipmap
 // REQUIRES: vulkan
@@ -267,8 +268,7 @@ bool run_test(sycl::range<NDims> dims, sycl::range<NDims> localSize,
   // Create input image memory
   auto inputImage = vkutil::createImage(imgType, format, {width, height, depth},
                                         VK_IMAGE_USAGE_TRANSFER_SRC_BIT |
-                                            VK_IMAGE_USAGE_TRANSFER_DST_BIT |
-                                            VK_IMAGE_USAGE_STORAGE_BIT,
+                                            VK_IMAGE_USAGE_TRANSFER_DST_BIT,
                                         mipLevels);
   VkMemoryRequirements memRequirements;
   auto inputImageMemoryTypeIndex = vkutil::getImageMemoryTypeIndex(

sycl/test-e2e/bindless_images/vulkan_interop/sampled_images.cpp

@@ -1,3 +1,4 @@
+// REQUIRES: aspect-ext_oneapi_bindless_images
 // REQUIRES: aspect-ext_oneapi_external_memory_import || (windows && level_zero && aspect-ext_oneapi_bindless_images)
 // REQUIRES: vulkan
 
@@ -302,14 +303,13 @@ bool run_test(sycl::range<NDims> dims, sycl::range<NDims> localSize,
 
   printString("Creating input image\n");
   // Create input image memory
-  auto inputImage = vkutil::createImage(imgType, format, {width, height, depth},
-                                        VK_IMAGE_USAGE_TRANSFER_SRC_BIT |
-                                            VK_IMAGE_USAGE_TRANSFER_DST_BIT |
-                                            VK_IMAGE_USAGE_STORAGE_BIT,
-                                        1 /*mipLevels*/
+  auto inputImage = vkutil::createImage(
+      imgType, format, {width, height, depth},
+      VK_IMAGE_USAGE_TRANSFER_SRC_BIT | VK_IMAGE_USAGE_TRANSFER_DST_BIT,
+      1 /*mipLevels*/
 #ifdef ENABLE_LINEAR_TILING
-                                        ,
-                                        true /*linearTiling*/
+      ,
+      true /*linearTiling*/
 #endif
   );
   VkMemoryRequirements memRequirements;

sycl/test-e2e/bindless_images/vulkan_interop/sampled_images_semaphore.cpp

@@ -1,3 +1,4 @@
+// REQUIRES: aspect-ext_oneapi_bindless_images
 // REQUIRES: target-nvidia || (windows && level_zero && aspect-ext_oneapi_bindless_images)
 // REQUIRES: vulkan
 

sycl/test-e2e/bindless_images/vulkan_interop/unsampled_images.cpp

@@ -1,3 +1,4 @@
+// REQUIRES: aspect-ext_oneapi_bindless_images
 // REQUIRES: aspect-ext_oneapi_external_memory_import || (windows && level_zero && aspect-ext_oneapi_bindless_images)
 // REQUIRES: vulkan
 

sycl/test-e2e/bindless_images/vulkan_interop/unsampled_images_semaphore.cpp

@@ -1,3 +1,4 @@
+// REQUIRES: aspect-ext_oneapi_bindless_images
 // REQUIRES: aspect-ext_oneapi_external_semaphore_import
 // REQUIRES: aspect-ext_oneapi_external_memory_import
 // REQUIRES: vulkan

sycl/test-e2e/bindless_images/vulkan_interop/unsampled_images_timeline_semaphore.cpp

@@ -1,3 +1,4 @@
+// REQUIRES: aspect-ext_oneapi_bindless_images
 // REQUIRES: aspect-ext_oneapi_external_memory_import
 // REQUIRES: aspect-ext_oneapi_external_semaphore_import
 // REQUIRES: vulkan

sycl/test-e2e/bindless_images/vulkan_interop/vulkan_common.hpp

@@ -61,8 +61,11 @@ static VkCommandPool vk_transferCmdPool;
 static VkCommandBuffer vk_computeCmdBuffer;
 static VkCommandBuffer vk_transferCmdBuffers[2];
 
+static bool supportsDedicatedAllocation = false;
 static bool requiresDedicatedAllocation = false;
 
+static bool supportsExternalSemaphore = false;
+
 // A static debug callback function that relays messages from the Vulkan
 // validation layer to the terminal.
 static VKAPI_ATTR VkBool32 VKAPI_CALL
@@ -137,17 +140,35 @@ VkResult setupInstance() {
   std::vector<const char *> requiredInstanceExtensions = {
       VK_EXT_DEBUG_UTILS_EXTENSION_NAME,
       VK_KHR_GET_PHYSICAL_DEVICE_PROPERTIES_2_EXTENSION_NAME,
-      VK_KHR_EXTERNAL_MEMORY_CAPABILITIES_EXTENSION_NAME,
-      VK_KHR_EXTERNAL_SEMAPHORE_CAPABILITIES_EXTENSION_NAME};
+      VK_KHR_EXTERNAL_MEMORY_CAPABILITIES_EXTENSION_NAME};
+
+  std::vector<const char *> optionalInstanceExtensions = {
+      VK_KHR_EXTERNAL_SEMAPHORE_CAPABILITIES_EXTENSION_NAME,
+      VK_KHR_DEDICATED_ALLOCATION_EXTENSION_NAME};
 
   // Make sure that our required instance extensions are supported by the
   // running Vulkan instance.
   for (int i = 0; i < requiredInstanceExtensions.size(); ++i) {
     std::string requiredExtension = requiredInstanceExtensions[i];
     if (std::find(supportedInstanceExtensions.begin(),
                   supportedInstanceExtensions.end(),
-                  requiredExtension) == supportedInstanceExtensions.end())
+                  requiredExtension) == supportedInstanceExtensions.end()) {
       return VK_ERROR_EXTENSION_NOT_PRESENT;
+    }
+  }
+
+  // Add any optional instance extensions that are supported by the
+  // running Vulkan instance.
+  for (int i = 0; i < optionalInstanceExtensions.size(); ++i) {
+    std::string optionalExtension = optionalInstanceExtensions[i];
+    if (std::find(supportedInstanceExtensions.begin(),
+                  supportedInstanceExtensions.end(),
+                  optionalExtension) != supportedInstanceExtensions.end()) {
+      requiredInstanceExtensions.push_back(optionalInstanceExtensions[i]);
+      if (optionalExtension == VK_KHR_DEDICATED_ALLOCATION_EXTENSION_NAME) {
+        supportsDedicatedAllocation = true;
+      }
+    }
   }
 
   // Create the vulkan instance with our required extensions and layers.
@@ -227,16 +248,25 @@ VkResult setupDevice(const sycl::device &dev) {
   static constexpr const char *requiredExtensions[] = {
       VK_KHR_GET_MEMORY_REQUIREMENTS_2_EXTENSION_NAME,
       VK_KHR_EXTERNAL_MEMORY_EXTENSION_NAME,
-      VK_KHR_EXTERNAL_SEMAPHORE_EXTENSION_NAME,
 #ifdef _WIN32
       VK_KHR_EXTERNAL_MEMORY_WIN32_EXTENSION_NAME,
+#else
+      VK_KHR_EXTERNAL_MEMORY_FD_EXTENSION_NAME,
+#endif
+  };
+
+  static constexpr const char *optionalExtensions[] = {
+      VK_KHR_EXTERNAL_SEMAPHORE_EXTENSION_NAME,
+#ifdef _WIN32
       VK_KHR_EXTERNAL_SEMAPHORE_WIN32_EXTENSION_NAME,
 #else
       VK_KHR_EXTERNAL_SEMAPHORE_FD_EXTENSION_NAME,
-      VK_KHR_EXTERNAL_MEMORY_FD_EXTENSION_NAME,
 #endif
   };
 
+  std::vector<const char *> enabledDeviceExtensions(
+      std::begin(requiredExtensions), std::end(requiredExtensions));
+
   const auto UUID = dev.get_info<sycl::ext::intel::info::device::uuid>();
 
   // From all physical devices, find the first one with a matching UUID
@@ -259,6 +289,7 @@ VkResult setupDevice(const sycl::device &dev) {
       continue;
     }
 
+    // Check if the device supports the required extensions.
     std::vector<VkExtensionProperties> supportedDeviceExtensions;
     getSupportedDeviceExtensions(supportedDeviceExtensions, vk_physical_device);
     const bool hasRequiredExtensions = std::all_of(
@@ -271,10 +302,29 @@ VkResult setupDevice(const sycl::device &dev) {
                                  });
           return (it != std::end(supportedDeviceExtensions));
         });
+    // Skip this device if it does not support all required extensions.
     if (!hasRequiredExtensions) {
       continue;
     }
 
+    // Check if the device supports the optional extensions, if so add them to
+    // the list of enabled device extensions.
+    for (const char *optionalExt : optionalExtensions) {
+      auto it = std::find_if(std::begin(supportedDeviceExtensions),
+                             std::end(supportedDeviceExtensions),
+                             [&](const VkExtensionProperties &ext) {
+                               return (ext.extensionName ==
+                                       std::string_view(optionalExt));
+                             });
+      if (it != std::end(supportedDeviceExtensions)) {
+        enabledDeviceExtensions.push_back(optionalExt);
+        if (std::string_view(optionalExt) ==
+            VK_KHR_EXTERNAL_SEMAPHORE_EXTENSION_NAME) {
+          supportsExternalSemaphore = true;
+        }
+      }
+    }
+
     foundDevice = true;
     std::cout << "Found suitable Vulkan device: "
               << devProps2.properties.deviceName << std::endl;
@@ -348,9 +398,8 @@ VkResult setupDevice(const sycl::device &dev) {
   dci.pQueueCreateInfos = qcis.data();
   dci.queueCreateInfoCount = qcis.size();
   dci.pEnabledFeatures = &deviceFeatures;
-  dci.enabledExtensionCount =
-      sizeof(requiredExtensions) / sizeof(requiredExtensions[0]);
-  dci.ppEnabledExtensionNames = &requiredExtensions[0];
+  dci.enabledExtensionCount = enabledDeviceExtensions.size();
+  dci.ppEnabledExtensionNames = enabledDeviceExtensions.data();
 
   VK_CHECK_CALL_RET(
       vkCreateDevice(vk_physical_device, &dci, nullptr, &vk_device));
@@ -371,13 +420,15 @@ VkResult setupDevice(const sycl::device &dev) {
         << "Could not get func pointer to \"vkGetMemoryWin32HandleKHR\"!\n";
     return VK_ERROR_UNKNOWN;
   }
-  vk_getSemaphoreWin32HandleKHR =
-      (PFN_vkGetSemaphoreWin32HandleKHR)vkGetDeviceProcAddr(
-          vk_device, "vkGetSemaphoreWin32HandleKHR");
-  if (!vk_getSemaphoreWin32HandleKHR) {
-    std::cerr
-        << "Could not get func pointer to \"vkGetSemaphoreWin32HandleKHR\"!\n";
-    return VK_ERROR_UNKNOWN;
+  if (supportsExternalSemaphore) {
+    vk_getSemaphoreWin32HandleKHR =
+        (PFN_vkGetSemaphoreWin32HandleKHR)vkGetDeviceProcAddr(
+            vk_device, "vkGetSemaphoreWin32HandleKHR");
+    if (!vk_getSemaphoreWin32HandleKHR) {
+      std::cerr << "Could not get func pointer to "
+                   "\"vkGetSemaphoreWin32HandleKHR\"!\n";
+      return VK_ERROR_UNKNOWN;
+    }
   }
 #else
   vk_getMemoryFdKHR =
@@ -386,11 +437,13 @@ VkResult setupDevice(const sycl::device &dev) {
     std::cerr << "Could not get func pointer to \"vkGetMemoryFdKHR\"!\n";
     return VK_ERROR_UNKNOWN;
   }
-  vk_getSemaphoreFdKHR = (PFN_vkGetSemaphoreFdKHR)vkGetDeviceProcAddr(
-      vk_device, "vkGetSemaphoreFdKHR");
-  if (!vk_getSemaphoreFdKHR) {
-    std::cerr << "Could not get func pointer to \"vkGetSemaphoreFdKHR\"!\n";
-    return VK_ERROR_UNKNOWN;
+  if (supportsExternalSemaphore) {
+    vk_getSemaphoreFdKHR = (PFN_vkGetSemaphoreFdKHR)vkGetDeviceProcAddr(
+        vk_device, "vkGetSemaphoreFdKHR");
+    if (!vk_getSemaphoreFdKHR) {
+      std::cerr << "Could not get func pointer to \"vkGetSemaphoreFdKHR\"!\n";
+      return VK_ERROR_UNKNOWN;
+    }
   }
 #endif
 
@@ -580,10 +633,11 @@ VkDeviceMemory allocateDeviceMemory(size_t size, uint32_t memoryTypeIndex,
 #else
     emai.handleTypes = VK_EXTERNAL_MEMORY_HANDLE_TYPE_OPAQUE_FD_BIT;
 #endif
-    if (requiresDedicatedAllocation)
+    if (requiresDedicatedAllocation) {
       dedicatedInfo.pNext = &emai;
-    else
+    } else {
       mai.pNext = &emai;
+    }
   }
 
   VkDeviceMemory memory;
@@ -601,12 +655,15 @@ property flags passed.
 */
 uint32_t getImageMemoryTypeIndex(VkImage image, VkMemoryPropertyFlags flags,
                                  VkMemoryRequirements &memRequirements) {
-  VkMemoryDedicatedRequirements dedicatedRequirements{};
-  dedicatedRequirements.sType = VK_STRUCTURE_TYPE_MEMORY_DEDICATED_REQUIREMENTS;
-
   VkMemoryRequirements2 memoryRequirements2{};
   memoryRequirements2.sType = VK_STRUCTURE_TYPE_MEMORY_REQUIREMENTS_2;
-  memoryRequirements2.pNext = &dedicatedRequirements;
+
+  VkMemoryDedicatedRequirements dedicatedRequirements{};
+  if (supportsDedicatedAllocation) {
+    dedicatedRequirements.sType =
+        VK_STRUCTURE_TYPE_MEMORY_DEDICATED_REQUIREMENTS;
+    memoryRequirements2.pNext = &dedicatedRequirements;
+  }
 
   VkImageMemoryRequirementsInfo2 imageRequirementsInfo{};
   imageRequirementsInfo.sType =
@@ -616,8 +673,9 @@ uint32_t getImageMemoryTypeIndex(VkImage image, VkMemoryPropertyFlags flags,
   vk_getImageMemoryRequirements2(vk_device, &imageRequirementsInfo,
                                  &memoryRequirements2);
 
-  if (dedicatedRequirements.requiresDedicatedAllocation)
+  if (dedicatedRequirements.requiresDedicatedAllocation) {
     requiresDedicatedAllocation = true;
+  }
 
   VkPhysicalDeviceMemoryProperties memProperties;
   vkGetPhysicalDeviceMemoryProperties(vk_physical_device, &memProperties);
@@ -715,6 +773,11 @@ HANDLE getSemaphoreWin32Handle(VkSemaphore semaphore) {
   sghwi.semaphore = semaphore;
   sghwi.handleType = VK_EXTERNAL_SEMAPHORE_HANDLE_TYPE_OPAQUE_WIN32_BIT;
 
+  if (!supportsExternalSemaphore) {
+    std::cerr << "External semaphore support is not enabled!\n";
+    return 0;
+  }
+
   if (vk_getSemaphoreWin32HandleKHR != nullptr) {
     VK_CHECK_CALL(vk_getSemaphoreWin32HandleKHR(vk_device, &sghwi, &retHandle));
   } else {
@@ -757,6 +820,12 @@ int getSemaphoreOpaqueFD(VkSemaphore semaphore) {
   sgfi.handleType = VK_EXTERNAL_SEMAPHORE_HANDLE_TYPE_OPAQUE_FD_BIT;
 
   int fd = 0;
+
+  if (!supportsExternalSemaphore) {
+    std::cerr << "External semaphore support is not enabled!\n";
+    return 0;
+  }
+
   if (vk_getSemaphoreFdKHR != nullptr) {
     VK_CHECK_CALL(vk_getSemaphoreFdKHR(vk_device, &sgfi, &fd));
   } else {
@@ -805,11 +874,9 @@ struct vulkan_image_test_resources_t {
 
   vulkan_image_test_resources_t(VkImageType imgType, VkFormat format,
                                 VkExtent3D ext, const size_t imageSizeBytes) {
-    vkImage = vkutil::createImage(imgType, format, ext,
-                                  VK_IMAGE_USAGE_TRANSFER_SRC_BIT |
-                                      VK_IMAGE_USAGE_TRANSFER_DST_BIT |
-                                      VK_IMAGE_USAGE_STORAGE_BIT,
-                                  1);
+    vkImage = vkutil::createImage(
+        imgType, format, ext,
+        VK_IMAGE_USAGE_TRANSFER_SRC_BIT | VK_IMAGE_USAGE_TRANSFER_DST_BIT, 1);
     VkMemoryRequirements memRequirements;
     auto inputImageMemoryTypeIndex = vkutil::getImageMemoryTypeIndex(
         vkImage, VK_MEMORY_PROPERTY_DEVICE_LOCAL_BIT, memRequirements);

unified-runtime/source/adapters/level_zero/device.cpp

@@ -1087,21 +1087,18 @@ ur_result_t urDeviceGetInfo(
   case UR_DEVICE_INFO_COMMAND_BUFFER_SUBGRAPH_SUPPORT_EXP:
     return ReturnValue(false);
   case UR_DEVICE_INFO_BINDLESS_IMAGES_SUPPORT_EXP: {
-    return ReturnValue(Device->isIntelDG2OrNewer() &&
-                       Device->ZeDeviceImageProperties->maxImageDims1D > 0 &&
-                       Device->ZeDeviceImageProperties->maxImageDims2D > 0 &&
-                       Device->ZeDeviceImageProperties->maxImageDims3D > 0);
+    return ReturnValue(Device->Platform->ZeBindlessImagesExtensionSupported);
   }
   case UR_DEVICE_INFO_BINDLESS_IMAGES_SHARED_USM_SUPPORT_EXP: {
     // On L0 bindless images can not be backed by shared (managed) USM.
     return ReturnValue(false);
   }
   case UR_DEVICE_INFO_BINDLESS_IMAGES_1D_USM_SUPPORT_EXP: {
-    return ReturnValue(Device->isIntelDG2OrNewer() &&
+    return ReturnValue(Device->Platform->ZeBindlessImagesExtensionSupported &&
                        Device->ZeDeviceImageProperties->maxImageDims1D > 0);
   }
   case UR_DEVICE_INFO_BINDLESS_IMAGES_2D_USM_SUPPORT_EXP: {
-    return ReturnValue(Device->isIntelDG2OrNewer() &&
+    return ReturnValue(Device->Platform->ZeBindlessImagesExtensionSupported &&
                        Device->ZeDeviceImageProperties->maxImageDims2D > 0);
   }
   case UR_DEVICE_INFO_IMAGE_PITCH_ALIGN_EXP: