From dbd39aa68eb0027c9730cc757d6f3ff0e2b5235e Mon Sep 17 00:00:00 2001
From: Wenju He <wenju.he@intel.com>
Date: Fri, 26 Jul 2024 13:53:28 +0800
Subject: [PATCH] [NFC][Bindless] Add a test of DX12 interop without semaphore

This allows testing of DX12 interop in L0 backend that doesn't support
semaphore importing yet.
Fix getDX12Adapter to increment index when software adapter is seen.
---
 .../dx12_interop/read_write_unsampled.cpp     | 27 ++++++++++++-------
 .../dx12_interop/read_write_unsampled.h       |  2 +-
 .../read_write_unsampled_semaphore.cpp        |  8 ++++++
 3 files changed, 26 insertions(+), 11 deletions(-)
 create mode 100644 sycl/test-e2e/bindless_images/dx12_interop/read_write_unsampled_semaphore.cpp

diff --git a/sycl/test-e2e/bindless_images/dx12_interop/read_write_unsampled.cpp b/sycl/test-e2e/bindless_images/dx12_interop/read_write_unsampled.cpp
index eacf3a40ebfd5..d142aa611b4e1 100644
--- a/sycl/test-e2e/bindless_images/dx12_interop/read_write_unsampled.cpp
+++ b/sycl/test-e2e/bindless_images/dx12_interop/read_write_unsampled.cpp
@@ -84,12 +84,14 @@ void DX12InteropTest::initDX12Resources() {
   ThrowIfFailed(m_dx12Device->CreateFence(
       m_sharedFenceValue, D3D12_FENCE_FLAG_SHARED, IID_PPV_ARGS(&m_dx12Fence)));
 
+#ifdef TEST_SEMAPHORE_IMPORT
   ThrowIfFailed(m_dx12Device->CreateSharedHandle(m_dx12Fence.Get(), nullptr,
                                                  GENERIC_ALL, nullptr,
                                                  &m_sharedSemaphoreHandle));
 
   // Import our shared DX12 fence resource to SYCL.
   importDX12SharedSemaphoreHandle();
+#endif
 
   // Create an event handle to use for synchronization.
   m_dx12FenceEvent = CreateEvent(nullptr, FALSE, FALSE, nullptr);
@@ -126,11 +128,12 @@ void DX12InteropTest::importDX12SharedSemaphoreHandle() {
 }
 
 void DX12InteropTest::callSYCLKernel() {
-
+#ifdef TEST_SEMAPHORE_IMPORT
   // Wait for imported semaphore. This semaphore was signalled at the
   // end of `populateDX12Texture`.
   m_syclQueue.ext_oneapi_wait_external_semaphore(m_syclInteropSemaphoreHandle,
                                                  m_sharedFenceValue);
+#endif
 
   // We can't capture the image handle through `this` in the lambda.
   // If we do the kernel will crash.
@@ -159,6 +162,7 @@ void DX12InteropTest::callSYCLKernel() {
     exit(-1);
   }
 
+#ifdef TEST_SEMAPHORE_IMPORT
   // Increment the fence value.
   m_sharedFenceValue++;
 
@@ -170,6 +174,9 @@ void DX12InteropTest::callSYCLKernel() {
 
   // Use DX12 to wait for the semaphore signalled by SYCL above.
   waitDX12Fence();
+#else
+  m_syclQueue.wait();
+#endif
 }
 
 void DX12InteropTest::populateDX12Texture() {
@@ -417,7 +424,8 @@ void DX12InteropTest::cleanupDX12() {
   waitDX12Fence();
 
   // Clean up opened handles
-  CloseHandle(m_sharedSemaphoreHandle);
+  if (m_sharedSemaphoreHandle != INVALID_HANDLE_VALUE)
+    CloseHandle(m_sharedSemaphoreHandle);
   CloseHandle(m_sharedMemoryHandle);
   CloseHandle(m_dx12FenceEvent);
 
@@ -436,16 +444,15 @@ void DX12InteropTest::getDX12Adapter(IDXGIFactory2 *pFactory,
     DXGI_ADAPTER_DESC1 desc;
     adapter->GetDesc1(&desc);
 
-    if (desc.Flags & DXGI_ADAPTER_FLAG_SOFTWARE) {
+    if (!(desc.Flags & DXGI_ADAPTER_FLAG_SOFTWARE)) {
       // We don't want a software adapter.
-      continue;
-    }
 
-    // Check to see if the adapter supports Direct3D 12, but don't create the
-    // actual device yet.
-    if (SUCCEEDED(D3D12CreateDevice(adapter.Get(), D3D_FEATURE_LEVEL_12_0,
-                                    _uuidof(ID3D12Device), nullptr))) {
-      break;
+      // Check to see if the adapter supports Direct3D 12, but don't create the
+      // actual device yet.
+      if (SUCCEEDED(D3D12CreateDevice(adapter.Get(), D3D_FEATURE_LEVEL_12_0,
+                                      _uuidof(ID3D12Device), nullptr))) {
+        break;
+      }
     }
 
     // Increment adapter index and find the next adapter.
diff --git a/sycl/test-e2e/bindless_images/dx12_interop/read_write_unsampled.h b/sycl/test-e2e/bindless_images/dx12_interop/read_write_unsampled.h
index fab2512d3bbdb..6cdc3906e5785 100644
--- a/sycl/test-e2e/bindless_images/dx12_interop/read_write_unsampled.h
+++ b/sycl/test-e2e/bindless_images/dx12_interop/read_write_unsampled.h
@@ -81,7 +81,7 @@ class DX12InteropTest {
   // Shared handles and values
   uint64_t m_sharedFenceValue;
   HANDLE m_sharedMemoryHandle;
-  HANDLE m_sharedSemaphoreHandle;
+  HANDLE m_sharedSemaphoreHandle = INVALID_HANDLE_VALUE;
 
   // SYCL Objects
   sycl::queue m_syclQueue;
diff --git a/sycl/test-e2e/bindless_images/dx12_interop/read_write_unsampled_semaphore.cpp b/sycl/test-e2e/bindless_images/dx12_interop/read_write_unsampled_semaphore.cpp
new file mode 100644
index 0000000000000..0a699e4d22438
--- /dev/null
+++ b/sycl/test-e2e/bindless_images/dx12_interop/read_write_unsampled_semaphore.cpp
@@ -0,0 +1,8 @@
+// REQUIRES: cuda
+// REQUIRES: windows
+
+// RUN: %{build} -l d3d12 -l dxgi -l dxguid -o %t.out
+// RUN: %t.out
+
+#define TEST_SEMAPHORE_IMPORT
+#include "read_write_unsampled.cpp"