[SYCL][E2E][Bindless] Add new VK/DX12 interop tests for L0 backend

[wenju-he]

L0 backend recently added support of following VkFormat:
VK_FORMAT_R16G16_SFLOAT
VK_FORMAT_R32_SFLOAT
VK_FORMAT_R16G16B16A16_SFLOAT
VK_FORMAT_R8G8B8A8_UNORM

In order to validate the support for upcoming oneAPI release,

• Update vulkan_interop/sampled_images.cpp to test above formats.
• Update vulkan_interop/unsampled_images.cpp to test above formats.
• Update vulkan_interop/read_write_unsampled.cpp to not test semaphore.
• Update dx12_interop/read_write_unsampled.cpp to test above formats.

[wenju-he]

Note new test sampled_images_l0.cpp has compfail in CUDA AOT (GTX 1060, Windows):

clang++ -fsycl sampled_images_l0.cpp -lD:/iusers/wenjuhe/repo/Vulkan-Loader/build/install/lib/vulkan-1.lib -ID:/iusers/wenjuhe/repo/Vulkan-Headers/build/install/include -w -fsycl-targets=nvptx64-nvidia-cuda
ptxas fatal   : Unresolved extern function '_Z17__spirv_ImageReadIDv3_DF16_yfET_T0_T1_'
llvm-foreach:
ptxas fatal   : Unresolved extern function '_Z17__spirv_ImageReadIDv3_hyfET_T0_T1_'
llvm-foreach:
Assertion failed: !KeyInfoT::isEqual(Val, EmptyKey) && !KeyInfoT::isEqual(Val, TombstoneKey) && "Empty/Tombstone value shouldn't be inserted into map!", file D:\iusers\wenjuhe\llvm\llvm\include\llvm/ADT/DenseMap.h, line 667
PLEASE submit a bug report to https://github.com/intel/llvm/issues and include the crash backtrace, preprocessed source, and associated run script.
Stack dump:
0.      Program arguments: clang++ -fsycl sampled_images_l0.cpp -lD:/iusers/wenjuhe/repo/Vulkan-Loader/build/install/lib/vulkan-1.lib -ID:/iusers/wenjuhe/repo/Vulkan-Headers/build/install/include -w -fsycl-targets=nvptx64-nvidia-cuda
Exception Code: 0x80000003
 #0 0x00007ff62f3a0f05 (D:\iusers\wenjuhe\llvm\build\bin\clang++.exe+0x1130f05)
 #1 0x00007ffefea0bab4 (C:\WINDOWS\System32\ucrtbase.dll+0x7bab4)
 #2 0x00007ffefea0ca81 (C:\WINDOWS\System32\ucrtbase.dll+0x7ca81)
 #3 0x00007ffefea0e50a (C:\WINDOWS\System32\ucrtbase.dll+0x7e50a)
 #4 0x00007ffefea0e401 (C:\WINDOWS\System32\ucrtbase.dll+0x7e401)
 #5 0x00007ffefea0e691 (C:\WINDOWS\System32\ucrtbase.dll+0x7e691)
 #6 0x00007ff62e55b08c (D:\iusers\wenjuhe\llvm\build\bin\clang++.exe+0x2eb08c)
 #7 0x00007ff62fddd4b4 (D:\iusers\wenjuhe\llvm\build\bin\clang++.exe+0x1b6d4b4)
 #8 0x00007ff62fcf8e3b (D:\iusers\wenjuhe\llvm\build\bin\clang++.exe+0x1a88e3b)
 #9 0x00007ff62fcfa36f (D:\iusers\wenjuhe\llvm\build\bin\clang++.exe+0x1a8a36f)
#10 0x00007ff62fcc85ee (D:\iusers\wenjuhe\llvm\build\bin\clang++.exe+0x1a585ee)
#11 0x00007ff62e30cc94 (D:\iusers\wenjuhe\llvm\build\bin\clang++.exe+0x9cc94)
#12 0x00007ff62e31b794 (D:\iusers\wenjuhe\llvm\build\bin\clang++.exe+0xab794)
#13 0x00007ff6324397c8 (D:\iusers\wenjuhe\llvm\build\bin\clang++.exe+0x41c97c8)
#14 0x00007ffeff6a4de0 (C:\WINDOWS\System32\KERNEL32.DLL+0x14de0)
#15 0x00007fff0145e3db (C:\WINDOWS\SYSTEM32\ntdll.dll+0x7e3db)

[ProGTX]

We tested this PR on CUDA on Windows and Linux, seems to pass.

[wenju-he]

Changes in commit 798faa4:

• Fixed vulkan_interop/sampled_images.cpp on DG2 windows.
• Removed VK_FORMAT_R16G16B16_SFLOAT and VK_FORMAT_R8G8B8_UNORM as 3-channels are not supported in SYCL bindless image.
• Update vulkan_interop/unsampled_images.cpp to test above formats.

@ProGTX please review again, thanks.

[wenju-he]

Changes in commit 798faa4:

• Fixed vulkan_interop/sampled_images.cpp on DG2 windows.
• Removed VK_FORMAT_R16G16B16_SFLOAT and VK_FORMAT_R8G8B8_UNORM as 3-channels are not supported in SYCL bindless image.
• Update vulkan_interop/unsampled_images.cpp to test above formats.

@intel/bindless-images-reviewers please review this new commit, thank you.

[wenju-he]

@intel/llvm-gatekeepers please merge, thanks

[DBDuncan]

@wenju-he This PR seems to be causing an off by one error in vulkan_interop/sampled_images.cpp. Getting lots of

Result mismatch! Expected: -5, Actual: -4
Result mismatch! Expected: -4, Actual: -3
Result mismatch! Expected: -3, Actual: -2
Result mismatch! Expected: -2, Actual: -1
Result mismatch! Expected: -1, Actual: 0

[wenju-he]

@wenju-he This PR seems to be causing an off by one error in vulkan_interop/sampled_images.cpp. Getting lots of

Result mismatch! Expected: -5, Actual: -4
Result mismatch! Expected: -4, Actual: -3
Result mismatch! Expected: -3, Actual: -2
Result mismatch! Expected: -2, Actual: -1
Result mismatch! Expected: -1, Actual: 0

thank you @DBDuncan . I reproduced the fail. I had tested this PR on CUDA windows, but somehow I missed the Test failed message.
The failure is caused by value mismatch between input

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

Line 275 in df49dcc

CType == sycl::image_channel_type::unorm_int8 ? 255 : i));

and output

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

Line 183 in df49dcc

CType == sycl::image_channel_type::unorm_int8 ? 0.5f : (i / 2.f)));

. The output first divide and then cast. This results in wrong value if globalSize.size() is larger than the maximum value of the type.
The test was passing before this PR is because the output value after multiplying 10.1f is always larger than the maximum value of the type.

I've submitted a fix: #15086

[DBDuncan]

No problem. Thanks for the fix!