intel · ProGTX · Jun 16, 2025
@@ -2074,6 +2074,7 @@ enum class external_mem_handle_type {
   opaque_fd = 0,
   win32_nt_handle = 1,
   win32_nt_dx12_resource = 2,
+  dma_buf = 3,
 };
 
 // Descriptor templated on specific resource type
@@ -2095,10 +2096,14 @@ for Windows NT resource handles.
 The user must populate the `external_mem_descriptor` with the appropriate 
 `ResourceType` values, a `handle_type`, and the size of the external memory in 
 bytes, before they can then import that memory into SYCL through 
-`import_external_memory`. Note that some handle types can only be used in 
-combination with certain resource types, for example the `opaque_fd` handle type
-is only used on Linux systems and is only compatible with the `resource_fd` 
-resource type.
+`import_external_memory`.
+
+Note that some handle types can only be used in 
+combination with certain resource types, for example the `opaque_fd`
+and `dma_buf` handle types are only used on Linux systems
+and are only compatible with the `resource_fd` resource type.
+The handle types supported by the device can be queried using the
+`supports_importing_handle_type` device query.
 
 ```cpp
 namespace sycl::ext::oneapi::experimental {
@@ -2139,6 +2144,10 @@ void *map_external_linear_memory(
     uint64_t size, uint64_t offset,
     const sycl::queue &syclQueue);
 }
+
+bool supports_importing_handle_type(
+    external_mem_handle_type externMemHandleType,
+    const sycl::device &syclDevice);
 ```
 
 The resulting `external_mem` can then be mapped, where the resulting type 
@@ -2728,3 +2737,5 @@ This query should be added in a later revision of the proposal.
 |6.10|2025-05-09| - Add `unmap_external_image_memory` and 
                     `unmap_external_linear_memory` APIs.
                   - Clarify how and when external memory should be unmapped.
+|6.11|2025-06-16| - Add `external_mem_handle_type::dma_buf`
+                  - Add `supports_importing_handle_type`
@@ -333,6 +333,16 @@ inline void unmap_external_image_memory(image_mem_handle mappedImageMem,
                               syclQueue.get_context());
 }
 
+/**
+ * @brief  Check if the device supports importing a handle of a specific type
+ * @param  externMemHandleType Type of external memory handle
+ * @param  syclDevice The device where we want to import memory
+ * @return true if the device supports importing the specified handle type
+ */
+__SYCL_EXPORT bool
+supports_importing_handle_type(external_mem_handle_type externMemHandleType,
+                               const sycl::device &syclDevice);
+
 /**
  *  @brief   Create an image and return the device image handle
  *

@@ -21,6 +21,7 @@ enum class external_mem_handle_type {
   opaque_fd = 0,
   win32_nt_handle = 1,
   win32_nt_dx12_resource = 2,
+  dma_buf = 3,
 };
 
 // Types of external semaphore handles

@@ -389,6 +389,34 @@ create_image(void *devPtr, size_t pitch, const bindless_image_sampler &sampler,
                       syclQueue.get_context());
 }
 
+namespace detail {
+
+/**
+ * Converts SYCL external_mem_handle_type to the corresponding UR type.
+ *
+ * Note that this function does a simple conversion
+ * and doesn't check the result validity for any specific scenario.
+ */
+constexpr ur_exp_external_mem_type_t
+to_ur_type(external_mem_handle_type externalMemHandleType) {
+  switch (externalMemHandleType) {
+  case external_mem_handle_type::opaque_fd:
+    return UR_EXP_EXTERNAL_MEM_TYPE_OPAQUE_FD;
+  case external_mem_handle_type::dma_buf:
+    return UR_EXP_EXTERNAL_MEM_TYPE_DMA_BUF;
+  case external_mem_handle_type::win32_nt_handle:
+    return UR_EXP_EXTERNAL_MEM_TYPE_WIN32_NT;
+  case external_mem_handle_type::win32_nt_dx12_resource:
+    return UR_EXP_EXTERNAL_MEM_TYPE_WIN32_NT_DX12_RESOURCE;
+  default:
+    // This ensures that all cases have to be handled
+    assert(false && "Invalid memory handle type");
+    return UR_EXP_EXTERNAL_MEM_TYPE_OPAQUE_FD; // Fallback
+  }
+}
+
+} // namespace detail
+
 template <>
 __SYCL_EXPORT external_mem import_external_memory<resource_fd>(
     external_mem_descriptor<resource_fd> externalMemDesc,
@@ -403,15 +431,18 @@ __SYCL_EXPORT external_mem import_external_memory<resource_fd>(
   urExternalMemDescriptor.stype = UR_STRUCTURE_TYPE_EXP_EXTERNAL_MEM_DESC;
   urExternalMemDescriptor.pNext = &urFileDescriptor;
 
-  // For `resource_fd` external memory type, the handle type is always
-  // `OPAQUE_FD`. No need for a switch statement like we have for win32
-  // resources.
+  const auto urHandleType = detail::to_ur_type(externalMemDesc.handle_type);
+  if ((urHandleType != UR_EXP_EXTERNAL_MEM_TYPE_OPAQUE_FD) &&
+      (urHandleType != UR_EXP_EXTERNAL_MEM_TYPE_DMA_BUF)) {
+    throw sycl::exception(sycl::make_error_code(sycl::errc::invalid),
+                          "Invalid memory handle type");
+  }
+
   Adapter
       ->call<sycl::errc::invalid,
              sycl::detail::UrApiKind::urBindlessImagesImportExternalMemoryExp>(
-          urCtx, urDevice, externalMemDesc.size_in_bytes,
-          UR_EXP_EXTERNAL_MEM_TYPE_OPAQUE_FD, &urExternalMemDescriptor,
-          &urExternalMem);
+          urCtx, urDevice, externalMemDesc.size_in_bytes, urHandleType,
+          &urExternalMemDescriptor, &urExternalMem);
 
   return external_mem{urExternalMem};
 }
@@ -438,16 +469,9 @@ __SYCL_EXPORT external_mem import_external_memory<resource_win32_handle>(
   urExternalMemDescriptor.stype = UR_STRUCTURE_TYPE_EXP_EXTERNAL_MEM_DESC;
   urExternalMemDescriptor.pNext = &urWin32Handle;
 
-  // Select appropriate memory handle type.
-  ur_exp_external_mem_type_t urHandleType;
-  switch (externalMemDesc.handle_type) {
-  case external_mem_handle_type::win32_nt_handle:
-    urHandleType = UR_EXP_EXTERNAL_MEM_TYPE_WIN32_NT;
-    break;
-  case external_mem_handle_type::win32_nt_dx12_resource:
-    urHandleType = UR_EXP_EXTERNAL_MEM_TYPE_WIN32_NT_DX12_RESOURCE;
-    break;
-  default:
+  const auto urHandleType = detail::to_ur_type(externalMemDesc.handle_type);
+  if ((urHandleType != UR_EXP_EXTERNAL_MEM_TYPE_WIN32_NT) &&
+      (urHandleType != UR_EXP_EXTERNAL_MEM_TYPE_WIN32_NT_DX12_RESOURCE)) {
     throw sycl::exception(sycl::make_error_code(sycl::errc::invalid),
                           "Invalid memory handle type");
   }
@@ -560,6 +584,19 @@ __SYCL_EXPORT void unmap_external_image_memory(
   free_image_mem(mappedImageMem, imageType, syclDevice, syclContext);
 }
 
+__SYCL_EXPORT bool
+supports_importing_handle_type(external_mem_handle_type externMemHandleType,
+                               const sycl::device &syclDevice) {
+  auto [urDevice, Adapter] = get_ur_handles(syclDevice);
+  const auto urHandleType = detail::to_ur_type(externMemHandleType);
+  ur_bool_t supportsExternalHandleType{0};
+  Adapter->call<
+      sycl::errc::invalid,
+      sycl::detail::UrApiKind::urBindlessImagesSupportsImportingHandleTypeExp>(
+      urDevice, urHandleType, &supportsExternalHandleType);
+  return static_cast<bool>(supportsExternalHandleType);
+}
+
 template <>
 __SYCL_EXPORT external_semaphore import_external_semaphore(
     external_semaphore_descriptor<resource_fd> externalSemaphoreDesc,

@@ -377,5 +377,12 @@ inline auto get_ur_handles(const sycl::device &syclDevice,
       sycl::detail::getSyclObjImpl(syclDevice)->getHandleRef();
   return std::tuple{urDevice, urCtx, Adapter};
 }
+inline auto get_ur_handles(const sycl::device &syclDevice) {
+  auto &implDevice = *sycl::detail::getSyclObjImpl(syclDevice);
+  ur_device_handle_t urDevice = implDevice.getHandleRef();
+  const sycl::detail::Adapter *Adapter = implDevice.getAdapter().get();
+  return std::tuple{urDevice, Adapter};
+}
+
 } // namespace _V1
 } // namespace sycl
@@ -50,7 +50,7 @@ int main() {
       sycl::ext::oneapi::experimental::resource_fd>
       output_ext_mem_desc{
           external_output_image_file_descriptor,
-          sycl::ext::oneapi::experimental::external_mem_handle_type::opaque_fd,
+          sycl::ext::oneapi::experimental::external_mem_handle_type::dma_buf,
           img_size_in_bytes};
 
   // An external API semaphore will signal this semaphore before our SYCL

@@ -23,31 +23,22 @@
 
 namespace syclexp = sycl::ext::oneapi::experimental;
 
-template <typename InteropMemHandleT>
-void runSycl(const sycl::device &syclDevice, sycl::range<1> globalSize,
+template <syclexp::external_mem_handle_type ExtMemHandleTypeV,
+          typename InteropMemHandleT>
+void runSycl(sycl::queue &syclQueue, sycl::range<1> globalSize,
              sycl::range<1> localSize, InteropMemHandleT extMemInHandle,
              InteropMemHandleT extMemOutHandle) {
-
-  sycl::queue syclQueue{syclDevice};
-
   const size_t bufferSizeBytes = globalSize.size() * sizeof(uint32_t);
 
-#ifdef _WIN32
-  syclexp::external_mem_descriptor<syclexp::resource_win32_handle> extMemInDesc{
-      extMemInHandle, syclexp::external_mem_handle_type::win32_nt_handle,
-      bufferSizeBytes};
-  syclexp::external_mem_descriptor<syclexp::resource_win32_handle>
-      extMemOutDesc{extMemOutHandle,
-                    syclexp::external_mem_handle_type::win32_nt_handle,
-                    bufferSizeBytes};
-#else
-  syclexp::external_mem_descriptor<syclexp::resource_fd> extMemInDesc{
-      extMemInHandle, syclexp::external_mem_handle_type::opaque_fd,
-      bufferSizeBytes};
-  syclexp::external_mem_descriptor<syclexp::resource_fd> extMemOutDesc{
-      extMemOutHandle, syclexp::external_mem_handle_type::opaque_fd,
-      bufferSizeBytes};
-#endif
+  using ResourceT =
+      std::conditional_t<(ExtMemHandleTypeV ==
+                          syclexp::external_mem_handle_type::win32_nt_handle),
+                         syclexp::resource_win32_handle, syclexp::resource_fd>;
+
+  syclexp::external_mem_descriptor<ResourceT> extMemInDesc{
+      extMemInHandle, ExtMemHandleTypeV, bufferSizeBytes};
+  syclexp::external_mem_descriptor<ResourceT> extMemOutDesc{
+      extMemOutHandle, ExtMemHandleTypeV, bufferSizeBytes};
 
   // Extension: create interop memory handles.
   syclexp::external_mem externalMemIn =
@@ -64,13 +55,13 @@ void runSycl(const sycl::device &syclDevice, sycl::range<1> globalSize,
 
   try {
     syclQueue.submit([&](sycl::handler &cgh) {
-      cgh.parallel_for<class TestVkBufferUSMInterop>(
-          sycl::nd_range<1>{globalSize, localSize}, [=](sycl::nd_item<1> it) {
-            size_t index = it.get_global_id(0);
+      cgh.parallel_for(sycl::nd_range<1>{globalSize, localSize},
+                       [=](sycl::nd_item<1> it) {
+                         size_t index = it.get_global_id(0);
 
-            uint32_t bufferValue = memIn[index];
-            memOut[index] = bufferValue * 2;
-          });
+                         uint32_t bufferValue = memIn[index];
+                         memOut[index] = bufferValue * 2;
+                       });
     });
 
     // Wait for kernel completion before destroying external objects.
@@ -91,8 +82,25 @@ void runSycl(const sycl::device &syclDevice, sycl::range<1> globalSize,
   }
 }
 
+template <syclexp::external_mem_handle_type ExtMemHandleTypeV>
 bool runTest(const sycl::device &syclDevice, sycl::range<1> bufferSize,
              sycl::range<1> localSize) {
+  sycl::queue syclQueue{syclDevice};
+  if constexpr (ExtMemHandleTypeV ==
+                syclexp::external_mem_handle_type::dma_buf) {
+    if (!supportsDmaBuf) {
+      std::cout
+          << "dma_buf test skipped because Vulkan driver does not support it\n";
+      return true;
+    }
+    if (!syclexp::supports_importing_handle_type(ExtMemHandleTypeV,
+                                                 syclDevice)) {
+      std::cout
+          << "dma_buf test skipped because SYCL backend does not support it\n";
+      return true;
+    }
+  }
+
   const size_t bufferSizeElems = bufferSize[0];
   const size_t bufferSizeBytes = bufferSizeElems * sizeof(uint32_t);
 
@@ -194,17 +202,26 @@ bool runTest(const sycl::device &syclDevice, sycl::range<1> bufferSize,
 
   printString("Getting memory interop handles\n");
   // Get memory interop handles.
+  const auto get_memory_handle = [](VkDeviceMemory vulkanDeviceMem) {
 #ifdef _WIN32
-  auto bufferMemIn = vkutil::getMemoryWin32Handle(vkInputBufferMemory);
-  auto bufferMemOut = vkutil::getMemoryWin32Handle(vkOutputBufferMemory);
+    return vkutil::getMemoryWin32Handle(vulkanDeviceMem);
 #else
-  auto bufferMemIn = vkutil::getMemoryOpaqueFD(vkInputBufferMemory);
-  auto bufferMemOut = vkutil::getMemoryOpaqueFD(vkOutputBufferMemory);
+    if constexpr (ExtMemHandleTypeV ==
+                  syclexp::external_mem_handle_type::dma_buf) {
+      return vkutil::getMemoryOpaqueFD<
+          VK_EXTERNAL_MEMORY_HANDLE_TYPE_DMA_BUF_BIT_EXT>(vulkanDeviceMem);
+    } else {
+      return vkutil::getMemoryOpaqueFD(vulkanDeviceMem);
+    }
 #endif
+  };
+  auto bufferMemIn = get_memory_handle(vkInputBufferMemory);
+  auto bufferMemOut = get_memory_handle(vkOutputBufferMemory);
 
   // Call into SYCL to read from input buffer, and populate the output buffer.
   printString("Calling into SYCL with interop memory handles\n");
-  runSycl(syclDevice, bufferSize, localSize, bufferMemIn, bufferMemOut);
+  runSycl<ExtMemHandleTypeV>(syclQueue, bufferSize, localSize, bufferMemIn,
+                             bufferMemOut);
 
   // Copy device buffer memory to temporary staging buffer, and back to host.
   printString("Copying buffer memory to host\n");
@@ -309,14 +326,34 @@ int main() {
     return EXIT_FAILURE;
   }
 
-  auto testPassed = runTest(syclDevice, {1024}, {256});
+  const auto globalSize = sycl::range<1>{1024};
+  const auto localSize = sycl::range<1>{256};
+#ifdef _WIN32
+  const bool opaqueTestPassed =
+      runTest<syclexp::external_mem_handle_type::win32_nt_handle>(
+          syclDevice, globalSize, localSize);
+  constexpr bool dmaBufTestPassed = true;
+#else
+  const bool opaqueTestPassed =
+      runTest<syclexp::external_mem_handle_type::opaque_fd>(
+          syclDevice, globalSize, localSize);
+  if (!opaqueTestPassed) {
+    std::cout << "opaque_fd test failed!\n";
+  }
+  const bool dmaBufTestPassed =
+      runTest<syclexp::external_mem_handle_type::dma_buf>(
+          syclDevice, globalSize, localSize);
+  if (!dmaBufTestPassed) {
+    std::cout << "dma_buf test failed!\n";
+  }
+#endif
 
   if (vkutil::cleanup() != VK_SUCCESS) {
     std::cerr << "Cleanup failed!\n";
     return EXIT_FAILURE;
   }
 
-  if (testPassed) {
+  if (opaqueTestPassed && dmaBufTestPassed) {
     std::cout << "Test passed!\n";
     return EXIT_SUCCESS;
   }