diff --git a/sycl/test-e2e/WorkGroupMemory/basic_usage.cpp b/sycl/test-e2e/WorkGroupMemory/basic_usage_common.hpp similarity index 95% rename from sycl/test-e2e/WorkGroupMemory/basic_usage.cpp rename to sycl/test-e2e/WorkGroupMemory/basic_usage_common.hpp index c63f16733b289..49e8428b6ed92 100644 --- a/sycl/test-e2e/WorkGroupMemory/basic_usage.cpp +++ b/sycl/test-e2e/WorkGroupMemory/basic_usage_common.hpp @@ -1,22 +1,20 @@ -// UNSUPPORTED: hip -// UNSUPPORTED-TRACKER: https://github.com/intel/llvm/issues/17339 -// RUN: %{build} -o %t.out -// RUN: %{run} %t.out -// XFAIL: spirv-backend -// XFAIL-TRACKER: https://github.com/intel/llvm/issues/18230 +// Shared implementation for the work-group memory basic usage tests. + +#pragma once #include #include #include #include #include +#include #include #include namespace syclexp = sycl::ext::oneapi::experimental; -sycl::queue q; +extern sycl::queue q; // This test performs a swap of two scalars/arrays inside a kernel using a // work_group_memory object as a temporary buffer. The test is done for scalar @@ -407,19 +405,3 @@ template void test_ptr() { } swap_array_2d(arr1, arr2, 8); } - -int main() { - test(); - test(); - test(); - if (q.get_device().has(sycl::aspect::fp16)) - test(); - test_ptr(); - test_ptr(); - test_ptr(); - test_ptr(); - if (q.get_device().has(sycl::aspect::fp16)) - test_ptr(); - test_ptr(); - return 0; -} diff --git a/sycl/test-e2e/WorkGroupMemory/basic_usage_test.cpp b/sycl/test-e2e/WorkGroupMemory/basic_usage_test.cpp new file mode 100644 index 0000000000000..858093da9da05 --- /dev/null +++ b/sycl/test-e2e/WorkGroupMemory/basic_usage_test.cpp @@ -0,0 +1,26 @@ +// Non-pointer types version of the basic usage test. + +// UNSUPPORTED: hip +// UNSUPPORTED-TRACKER: https://github.com/intel/llvm/issues/17339 + +// XFAIL: spirv-backend +// XFAIL-TRACKER: https://github.com/intel/llvm/issues/18230 + +// RUN: %{build} -o %t.out +// RUN: %{run} %t.out + +#include + +#include "./basic_usage_common.hpp" + +sycl::queue q; + +int main() { + test(); + test(); + test(); + test(); + if (q.get_device().has(sycl::aspect::fp16)) + test(); + return 0; +} diff --git a/sycl/test-e2e/WorkGroupMemory/basic_usage_test_ptr.cpp b/sycl/test-e2e/WorkGroupMemory/basic_usage_test_ptr.cpp new file mode 100644 index 0000000000000..29bb8147a1dd5 --- /dev/null +++ b/sycl/test-e2e/WorkGroupMemory/basic_usage_test_ptr.cpp @@ -0,0 +1,26 @@ +// Pointer-types version of the basic usage test. + +// UNSUPPORTED: hip +// UNSUPPORTED-TRACKER: https://github.com/intel/llvm/issues/17339 + +// XFAIL: spirv-backend +// XFAIL-TRACKER: https://github.com/intel/llvm/issues/18230 + +// RUN: %{build} -o %t.out +// RUN: %{run} %t.out + +#include + +#include "./basic_usage_common.hpp" + +sycl::queue q; + +int main() { + test_ptr(); + test_ptr(); + test_ptr(); + test_ptr(); + if (q.get_device().has(sycl::aspect::fp16)) + test_ptr(); + return 0; +} diff --git a/sycl/test-e2e/bindless_images/read_sampled_1d.cpp b/sycl/test-e2e/bindless_images/read_sampled_1d.cpp new file mode 100644 index 0000000000000..d179fcb2401e6 --- /dev/null +++ b/sycl/test-e2e/bindless_images/read_sampled_1d.cpp @@ -0,0 +1,31 @@ +// 1D version of sampled image read test. + +// REQUIRES: aspect-ext_oneapi_bindless_images + +// UNSUPPORTED: hip +// UNSUPPORTED-INTENDED: Returning non-FP values from sampling fails on HIP. + +// UNSUPPORTED: linux && arch-intel_gpu_bmg_g21 && level_zero_v2_adapter +// UNSUPPORTED-TRACKER: https://github.com/intel/llvm/issues/20223 + +// RUN: %{build} -o %t.out +// RUN: %{run} %t.out + +#include "./read_sampled_common.hpp" + +int main() { + const int int seed = 0; + const float offset = 20.0f; + + std::cout << "Running 1D Sampled Image Tests!\n"; + bool result1D = runAll1D(offset, seed); + + if (result1D) { + std::cout << "All tests passed!\n"; + } else { + std::cerr << "An error has occurred!\n"; + return 1; + } + + return 0; +} diff --git a/sycl/test-e2e/bindless_images/read_sampled_2d.cpp b/sycl/test-e2e/bindless_images/read_sampled_2d.cpp new file mode 100644 index 0000000000000..4170a53592d0a --- /dev/null +++ b/sycl/test-e2e/bindless_images/read_sampled_2d.cpp @@ -0,0 +1,34 @@ +// 2D version of sampled image read test. + +// REQUIRES: aspect-ext_oneapi_bindless_images + +// UNSUPPORTED: hip +// UNSUPPORTED-INTENDED: Returning non-FP values from sampling fails on HIP. + +// UNSUPPORTED: linux && arch-intel_gpu_bmg_g21 && level_zero_v2_adapter +// UNSUPPORTED-TRACKER: https://github.com/intel/llvm/issues/20223 + +// XFAIL: cuda && cuda-major-ge-13 +// XFAIL-TRACKER: https://github.com/intel/llvm/issues/21807 + +// RUN: %{build} -o %t.out +// RUN: %{run} %t.out + +#include "./read_sampled_common.hpp" + +int main() { + const int seed = 0; + const float offset = 20.0f; + + std::cout << "Running 2D Sampled Image Tests!\n"; + bool result2D = runAll2D(offset, seed); + + if (result2D) { + std::cout << "All tests passed!\n"; + } else { + std::cerr << "An error has occurred!\n"; + return 1; + } + + return 0; +} diff --git a/sycl/test-e2e/bindless_images/read_sampled.cpp b/sycl/test-e2e/bindless_images/read_sampled_common.hpp similarity index 97% rename from sycl/test-e2e/bindless_images/read_sampled.cpp rename to sycl/test-e2e/bindless_images/read_sampled_common.hpp index c30f0a05b2c65..3166a2c9b275f 100644 --- a/sycl/test-e2e/bindless_images/read_sampled.cpp +++ b/sycl/test-e2e/bindless_images/read_sampled_common.hpp @@ -1,13 +1,6 @@ -// REQUIRES: aspect-ext_oneapi_bindless_images +// Shared implementation for the sampled image read tests. -// UNSUPPORTED: hip -// UNSUPPORTED-INTENDED: Returning non-FP values from sampling fails on HIP. - -// UNSUPPORTED: linux && arch-intel_gpu_bmg_g21 && level_zero_v2_adapter -// UNSUPPORTED-TRACKER: https://github.com/intel/llvm/issues/20223 - -// RUN: %{build} -o %t.out -// RUN: %{run} %t.out +#pragma once // Print test names and pass status // #define VERBOSE_LV1 @@ -26,6 +19,8 @@ #include #include #include +#include +#include #include @@ -488,7 +483,7 @@ bool runTests(sycl::range<2> dims, sycl::range<2> localSize, float offset, int seed, sycl::coordinate_normalization_mode normMode) { // addressing_mode::none currently removed due to - // inconsistent behavour when switching between + // inconsistent behaviour when switching between // normalized and unnormalized coords. sycl::addressing_mode addrModes[4] = { sycl::addressing_mode::repeat, sycl::addressing_mode::mirrored_repeat, @@ -718,22 +713,10 @@ bool runAll(sycl::range dims, sycl::range localSize, float offset, return offsetPassed && noOffsetPassed; } -int main() { - - const unsigned int seed = 0; - const float offset = 20.0; - - std::cout << "Running 1D Sampled Image Tests!\n"; - bool result1D = runAll<1>({128}, {32}, offset, seed); - std::cout << "Running 2D Sampled Image Tests!\n"; - bool result2D = runAll<2>({16, 16}, {8, 8}, offset, seed); - - if (result1D && result2D) { - std::cout << "All tests passed!\n"; - } else { - std::cerr << "An error has occurred!\n"; - return 1; - } +inline bool runAll1D(float offset, int seed) { + return runAll<1>(sycl::range<1>{128}, sycl::range<1>{32}, offset, seed); +} - return 0; +inline bool runAll2D(float offset, int seed) { + return runAll<2>(sycl::range<2>{16, 16}, sycl::range<2>{8, 8}, offset, seed); } diff --git a/sycl/test-e2e/bindless_images/vulkan_interop/vulkan_sycl_image_interop_read_1d_channels.cpp b/sycl/test-e2e/bindless_images/vulkan_interop/vulkan_sycl_image_interop_read_1d_channels.cpp new file mode 100644 index 0000000000000..1f5a63e5abb90 --- /dev/null +++ b/sycl/test-e2e/bindless_images/vulkan_interop/vulkan_sycl_image_interop_read_1d_channels.cpp @@ -0,0 +1,39 @@ +// Unsampled version of the Vulkan/SYCL 1D image read interop test. + +// REQUIRES: aspect-ext_oneapi_bindless_images +// REQUIRES: aspect-ext_oneapi_external_memory_import || (windows && level_zero && aspect-ext_oneapi_bindless_images) +// REQUIRES: vulkan + +// RUN: %{build} %link-vulkan -o %t.out %if target-spir %{ -Wno-ignored-attributes %} + +// clang-format off +// RUN: %{run} %t.out --type float --channels 1 32 +// RUN: %{run} %t.out --type float --channels 2 32 +// RUN: %{run} %t.out --type float --channels 4 32 +// RUN: %{run} %t.out --type half --channels 1 32 +// RUN: %{run} %t.out --type half --channels 2 32 +// RUN: %{run} %t.out --type half --channels 4 32 +// RUN: %{run} %t.out --type int32 --channels 1 32 +// RUN: %{run} %t.out --type int32 --channels 2 32 +// RUN: %{run} %t.out --type int32 --channels 4 32 +// RUN: %{run} %t.out --type uint32 --channels 1 32 +// RUN: %{run} %t.out --type uint32 --channels 2 32 +// RUN: %{run} %t.out --type uint32 --channels 4 32 +// RUN: %{run} %t.out --type int16 --channels 1 32 +// RUN: %{run} %t.out --type int16 --channels 2 32 +// RUN: %{run} %t.out --type int16 --channels 4 32 +// RUN: %{run} %t.out --type uint16 --channels 1 32 +// RUN: %{run} %t.out --type uint16 --channels 2 32 +// RUN: %{run} %t.out --type uint16 --channels 4 32 +// RUN: %{run} %t.out --type uint8 --channels 1 32 +// RUN: %{run} %t.out --type uint8 --channels 2 32 +// RUN: %{run} %t.out --type uint8 --channels 4 32 +// RUN: %{run} %t.out --type int8 --channels 1 32 +// RUN: %{run} %t.out --type int8 --channels 2 32 +// RUN: %{run} %t.out --type int8 --channels 4 32 +// RUN-IF: !cuda, %{run} %t.out --type unorm8 --channels 1 32 +// RUN-IF: !cuda, %{run} %t.out --type unorm8 --channels 2 32 +// RUN-IF: !cuda, %{run} %t.out --type unorm8 --channels 4 32 +// clang-format on + +#include "./vulkan_sycl_image_interop_read_1d_common.hpp" diff --git a/sycl/test-e2e/bindless_images/vulkan_interop/vulkan_sycl_image_interop_read_1d.cpp b/sycl/test-e2e/bindless_images/vulkan_interop/vulkan_sycl_image_interop_read_1d_common.hpp similarity index 71% rename from sycl/test-e2e/bindless_images/vulkan_interop/vulkan_sycl_image_interop_read_1d.cpp rename to sycl/test-e2e/bindless_images/vulkan_interop/vulkan_sycl_image_interop_read_1d_common.hpp index 1e5a333d2e26d..fb710b001f4b6 100644 --- a/sycl/test-e2e/bindless_images/vulkan_interop/vulkan_sycl_image_interop_read_1d.cpp +++ b/sycl/test-e2e/bindless_images/vulkan_interop/vulkan_sycl_image_interop_read_1d_common.hpp @@ -1,136 +1,61 @@ -// REQUIRES: aspect-ext_oneapi_bindless_images -// REQUIRES: aspect-ext_oneapi_external_memory_import || (windows && level_zero && aspect-ext_oneapi_bindless_images) -// REQUIRES: vulkan - -// RUN: %{build} %link-vulkan -o %t.out %if target-spir %{ -Wno-ignored-attributes %} +// Shared implementation for the Vulkan/SYCL 1D image read interop tests. /* - Run ALL the vulkan formats through the gauntlet. sampled and unsampled. - This entire test takes less than 30 seconds on a slow machine. MUCH faster - (and more complete coveraage) than SFINAE based approach. - - IF a particular variant is having problems on some platform, please do NOT - just disable the whole test, instead use RUN~IF (SOMETHING) yadda-yadda - to enable/disable that variant. + Run ALL the vulkan formats through the gauntlet. sampled and unsampled. + This entire test takes less than 30 seconds on a slow machine. MUCH faster + (and more complete coveraage) than SFINAE based approach. - For semaphore testing, we run just a sampling. Note, that on Linux if there - is a failure in the first section, then likely ALL semaphore tests afterwards - will fail. This is being tracked as a separate issue. + IF a particular variant is having problems on some platform, please do NOT + just disable the whole test, instead use RUN~IF (SOMETHING) yadda-yadda + to enable/disable that variant. + For semaphore testing, we run just a sampling. Note, that on Linux if there + is a failure in the first section, then likely ALL semaphore tests afterwards + will fail. This is being tracked as a separate issue. */ -// clang-format off -// RUN: %{run} %t.out --type float --channels 1 32 -// RUN: %{run} %t.out --type float --channels 2 32 -// RUN: %{run} %t.out --type float --channels 4 32 -// RUN: %{run} %t.out --type half --channels 1 32 -// RUN: %{run} %t.out --type half --channels 2 32 -// RUN: %{run} %t.out --type half --channels 4 32 -// RUN: %{run} %t.out --type int32 --channels 1 32 -// RUN: %{run} %t.out --type int32 --channels 2 32 -// RUN: %{run} %t.out --type int32 --channels 4 32 -// RUN: %{run} %t.out --type uint32 --channels 1 32 -// RUN: %{run} %t.out --type uint32 --channels 2 32 -// RUN: %{run} %t.out --type uint32 --channels 4 32 -// RUN: %{run} %t.out --type int16 --channels 1 32 -// RUN: %{run} %t.out --type int16 --channels 2 32 -// RUN: %{run} %t.out --type int16 --channels 4 32 -// RUN: %{run} %t.out --type uint16 --channels 1 32 -// RUN: %{run} %t.out --type uint16 --channels 2 32 -// RUN: %{run} %t.out --type uint16 --channels 4 32 -// RUN: %{run} %t.out --type uint8 --channels 1 32 -// RUN: %{run} %t.out --type uint8 --channels 2 32 -// RUN: %{run} %t.out --type uint8 --channels 4 32 -// RUN: %{run} %t.out --type int8 --channels 1 32 -// RUN: %{run} %t.out --type int8 --channels 2 32 -// RUN: %{run} %t.out --type int8 --channels 4 32 -// RUN-IF: !cuda, %{run} %t.out --type unorm8 --channels 1 32 -// RUN-IF: !cuda, %{run} %t.out --type unorm8 --channels 2 32 -// RUN-IF: !cuda, %{run} %t.out --type unorm8 --channels 4 32 -// RUN: %{run} %t.out --type float --channels 1 --sampled 32 -// RUN: %{run} %t.out --type float --channels 2 --sampled 32 -// RUN: %{run} %t.out --type float --channels 4 --sampled 32 -// RUN: %{run} %t.out --type half --channels 1 --sampled 32 -// RUN: %{run} %t.out --type half --channels 2 --sampled 32 -// RUN: %{run} %t.out --type half --channels 4 --sampled 32 -// RUN: %{run} %t.out --type int32 --channels 1 --sampled 32 -// RUN: %{run} %t.out --type int32 --channels 2 --sampled 32 -// RUN: %{run} %t.out --type int32 --channels 4 --sampled 32 -// RUN: %{run} %t.out --type uint32 --channels 1 --sampled 32 -// RUN: %{run} %t.out --type uint32 --channels 2 --sampled 32 -// RUN: %{run} %t.out --type uint32 --channels 4 --sampled 32 -// RUN: %{run} %t.out --type int16 --channels 1 --sampled 32 -// RUN: %{run} %t.out --type int16 --channels 2 --sampled 32 -// RUN: %{run} %t.out --type int16 --channels 4 --sampled 32 -// RUN: %{run} %t.out --type uint16 --channels 1 --sampled 32 -// RUN: %{run} %t.out --type uint16 --channels 2 --sampled 32 -// RUN: %{run} %t.out --type uint16 --channels 4 --sampled 32 -// RUN: %{run} %t.out --type uint8 --channels 1 --sampled 32 -// RUN: %{run} %t.out --type uint8 --channels 2 --sampled 32 -// RUN: %{run} %t.out --type uint8 --channels 4 --sampled 32 -// RUN: %{run} %t.out --type int8 --channels 1 --sampled 32 -// RUN: %{run} %t.out --type int8 --channels 2 --sampled 32 -// RUN: %{run} %t.out --type int8 --channels 4 --sampled 32 -// RUN-IF: !cuda, %{run} %t.out --type unorm8 --channels 1 --sampled 32 -// RUN-IF: !cuda, %{run} %t.out --type unorm8 --channels 2 --sampled 32 -// RUN-IF: !cuda, %{run} %t.out --type unorm8 --channels 4 --sampled 32 - -// RUN: %{run} %t.out --type float --channels 1 32 --semaphores -// RUN: %{run} %t.out --type float --channels 2 32 --semaphores -// RUN: %{run} %t.out --type float --channels 4 32 --semaphores -// RUN: %{run} %t.out --type half --channels 1 32 --semaphores -// RUN: %{run} %t.out --type int32 --channels 2 32 --semaphores -// RUN: %{run} %t.out --type uint32 --channels 4 32 --semaphores -// RUN: %{run} %t.out --type int16 --channels 1 32 --semaphores -// RUN: %{run} %t.out --type uint16 --channels 2 32 --semaphores -// RUN: %{run} %t.out --type uint8 --channels 4 32 --semaphores -// RUN: %{run} %t.out --type int8 --channels 1 32 --semaphores -// RUN-IF: !cuda, %{run} %t.out --type unorm8 --channels 2 32 --semaphores -// RUN: %{run} %t.out --type float --channels 4 --sampled 32 --semaphores -// RUN: %{run} %t.out --type int16 --channels 4 --sampled 32 --semaphores -// RUN: %{run} %t.out --type int8 --channels 4 --sampled 32 --semaphores -// RUN-IF: !cuda, %{run} %t.out --type unorm8 --channels 4 --sampled 32 --semaphores /* - -The block above tests these formats, sampled and unsampled, with and without -semaphores - -VK_FORMAT_R32_SFLOAT -VK_FORMAT_R32G32_SFLOAT -VK_FORMAT_R32G32B32A32_SFLOAT -VK_FORMAT_R16_SFLOAT -VK_FORMAT_R16G16_SFLOAT -VK_FORMAT_R16G16B16A16_SFLOAT -VK_FORMAT_R32_SINT -VK_FORMAT_R32G32_SINT -VK_FORMAT_R32G32B32A32_SINT -VK_FORMAT_R32_UINT -VK_FORMAT_R32G32_UINT -VK_FORMAT_R32G32B32A32_UINT -VK_FORMAT_R16_SINT -VK_FORMAT_R16G16_SINT -VK_FORMAT_R16G16B16A16_SINT -VK_FORMAT_R16_UINT -VK_FORMAT_R16G16_UINT -VK_FORMAT_R16G16B16A16_UINT -VK_FORMAT_R8_UINT -VK_FORMAT_R8G8_UINT -VK_FORMAT_R8G8B8A8_UINT -VK_FORMAT_R8_SINT -VK_FORMAT_R8G8_SINT -VK_FORMAT_R8G8B8A8_SINT -VK_FORMAT_R8_UNORM -VK_FORMAT_R8G8_UNORM -VK_FORMAT_R8G8B8A8_UNORM - + The block above tests these formats, sampled and unsampled, with and without + semaphores + + VK_FORMAT_R32_SFLOAT + VK_FORMAT_R32G32_SFLOAT + VK_FORMAT_R32G32B32A32_SFLOAT + VK_FORMAT_R16_SFLOAT + VK_FORMAT_R16G16_SFLOAT + VK_FORMAT_R16G16B16A16_SFLOAT + VK_FORMAT_R32_SINT + VK_FORMAT_R32G32_SINT + VK_FORMAT_R32G32B32A32_SINT + VK_FORMAT_R32_UINT + VK_FORMAT_R32G32_UINT + VK_FORMAT_R32G32B32A32_UINT + VK_FORMAT_R16_SINT + VK_FORMAT_R16G16_SINT + VK_FORMAT_R16G16B16A16_SINT + VK_FORMAT_R16_UINT + VK_FORMAT_R16G16_UINT + VK_FORMAT_R16G16B16A16_UINT + VK_FORMAT_R8_UINT + VK_FORMAT_R8G8_UINT + VK_FORMAT_R8G8B8A8_UINT + VK_FORMAT_R8_SINT + VK_FORMAT_R8G8_SINT + VK_FORMAT_R8G8B8A8_SINT + VK_FORMAT_R8_UNORM + VK_FORMAT_R8G8_UNORM + VK_FORMAT_R8G8B8A8_UNORM */ +// clang-format off /* Vulkan/SYCL 1D Image Read Test (Sampled + Unsampled) - clang++ -fsycl -o vsr_1d_test.bin vulkan_sycl_image_interop_read_1d.cpp -lvulkan -I$VULKAN_SDK/include -L$VULKAN_SDK/lib + clang++ -fsycl -o vsr_1d_test.bin vulkan_sycl_image_interop_read_1d.cpp + -lvulkan -I$VULKAN_SDK/include -L$VULKAN_SDK/lib - clang++ -fsycl -o vsr_1d_test.exe vulkan_sycl_image_interop_read_1d.cpp -Wno-ignored-attributes -lvulkan-1 -I$VULKAN_SDK/Include -L$VULKAN_SDK/Lib + clang++ -fsycl -o vsr_1d_test.exe vulkan_sycl_image_interop_read_1d.cpp + -Wno-ignored-attributes -lvulkan-1 -I$VULKAN_SDK/Include -L$VULKAN_SDK/Lib USAGE: ./vsr_1d_test.bin [FLAGS] [Wx] @@ -140,8 +65,9 @@ VK_FORMAT_R8G8B8A8_UNORM --semaphores Use Vulkan Semaphores for SYCL Interop Sync --linear Use LINEAR tiling for the Vulkan Image (default is OPTIMAL) --channels X Set number of channels (1, 2, or 4). Default is 4 (RGBA) - --type XXX Set data type (float, half, uint32, int32, uint16, int16, uint8, int8, unorm8). Default is float - Wx Set custom Width (e.g. 64x) + --type XXX Set data type (float, half, uint32, int32, uint16, int16, + uint8, int8, unorm8). Default is float Wx Set custom Width (e.g. + 64x) EXAMPLES: ./vsr_1d_test.bin @@ -152,6 +78,8 @@ VK_FORMAT_R8G8B8A8_UNORM */ // clang-format on +#pragma once + #include "vulkan_setup.hpp" #include @@ -548,4 +476,4 @@ int main(int argc, char **argv) { std::cerr << "Unknown type: " << type << std::endl; return 1; -} \ No newline at end of file +} diff --git a/sycl/test-e2e/bindless_images/vulkan_interop/vulkan_sycl_image_interop_read_1d_sampled.cpp b/sycl/test-e2e/bindless_images/vulkan_interop/vulkan_sycl_image_interop_read_1d_sampled.cpp new file mode 100644 index 0000000000000..9cee6db306dbf --- /dev/null +++ b/sycl/test-e2e/bindless_images/vulkan_interop/vulkan_sycl_image_interop_read_1d_sampled.cpp @@ -0,0 +1,39 @@ +// Sampled-only version of the Vulkan/SYCL 1D image read interop test. + +// REQUIRES: aspect-ext_oneapi_bindless_images +// REQUIRES: aspect-ext_oneapi_external_memory_import || (windows && level_zero && aspect-ext_oneapi_bindless_images) +// REQUIRES: vulkan + +// RUN: %{build} %link-vulkan -o %t.out %if target-spir %{ -Wno-ignored-attributes %} + +// clang-format off +// RUN: %{run} %t.out --type float --channels 1 --sampled 32 +// RUN: %{run} %t.out --type float --channels 2 --sampled 32 +// RUN: %{run} %t.out --type float --channels 4 --sampled 32 +// RUN: %{run} %t.out --type half --channels 1 --sampled 32 +// RUN: %{run} %t.out --type half --channels 2 --sampled 32 +// RUN: %{run} %t.out --type half --channels 4 --sampled 32 +// RUN: %{run} %t.out --type int32 --channels 1 --sampled 32 +// RUN: %{run} %t.out --type int32 --channels 2 --sampled 32 +// RUN: %{run} %t.out --type int32 --channels 4 --sampled 32 +// RUN: %{run} %t.out --type uint32 --channels 1 --sampled 32 +// RUN: %{run} %t.out --type uint32 --channels 2 --sampled 32 +// RUN: %{run} %t.out --type uint32 --channels 4 --sampled 32 +// RUN: %{run} %t.out --type int16 --channels 1 --sampled 32 +// RUN: %{run} %t.out --type int16 --channels 2 --sampled 32 +// RUN: %{run} %t.out --type int16 --channels 4 --sampled 32 +// RUN: %{run} %t.out --type uint16 --channels 1 --sampled 32 +// RUN: %{run} %t.out --type uint16 --channels 2 --sampled 32 +// RUN: %{run} %t.out --type uint16 --channels 4 --sampled 32 +// RUN: %{run} %t.out --type uint8 --channels 1 --sampled 32 +// RUN: %{run} %t.out --type uint8 --channels 2 --sampled 32 +// RUN: %{run} %t.out --type uint8 --channels 4 --sampled 32 +// RUN: %{run} %t.out --type int8 --channels 1 --sampled 32 +// RUN: %{run} %t.out --type int8 --channels 2 --sampled 32 +// RUN: %{run} %t.out --type int8 --channels 4 --sampled 32 +// RUN-IF: !cuda, %{run} %t.out --type unorm8 --channels 1 --sampled 32 +// RUN-IF: !cuda, %{run} %t.out --type unorm8 --channels 2 --sampled 32 +// RUN-IF: !cuda, %{run} %t.out --type unorm8 --channels 4 --sampled 32 +// clang-format on + +#include "./vulkan_sycl_image_interop_read_1d_common.hpp" diff --git a/sycl/test-e2e/bindless_images/vulkan_interop/vulkan_sycl_image_interop_read_1d_semaphores.cpp b/sycl/test-e2e/bindless_images/vulkan_interop/vulkan_sycl_image_interop_read_1d_semaphores.cpp new file mode 100644 index 0000000000000..75811685f7c57 --- /dev/null +++ b/sycl/test-e2e/bindless_images/vulkan_interop/vulkan_sycl_image_interop_read_1d_semaphores.cpp @@ -0,0 +1,27 @@ +// Semaphore version of the Vulkan/SYCL 1D image read interop test. + +// REQUIRES: aspect-ext_oneapi_bindless_images +// REQUIRES: aspect-ext_oneapi_external_memory_import || (windows && level_zero && aspect-ext_oneapi_bindless_images) +// REQUIRES: vulkan + +// RUN: %{build} %link-vulkan -o %t.out %if target-spir %{ -Wno-ignored-attributes %} + +// clang-format off +// RUN: %{run} %t.out --type float --channels 1 32 --semaphores +// RUN: %{run} %t.out --type float --channels 2 32 --semaphores +// RUN: %{run} %t.out --type float --channels 4 32 --semaphores +// RUN: %{run} %t.out --type half --channels 1 32 --semaphores +// RUN: %{run} %t.out --type int32 --channels 2 32 --semaphores +// RUN: %{run} %t.out --type uint32 --channels 4 32 --semaphores +// RUN: %{run} %t.out --type int16 --channels 1 32 --semaphores +// RUN: %{run} %t.out --type uint16 --channels 2 32 --semaphores +// RUN: %{run} %t.out --type uint8 --channels 4 32 --semaphores +// RUN: %{run} %t.out --type int8 --channels 1 32 --semaphores +// RUN-IF: !cuda, %{run} %t.out --type unorm8 --channels 2 32 --semaphores +// RUN: %{run} %t.out --type float --channels 4 --sampled 32 --semaphores +// RUN: %{run} %t.out --type int16 --channels 4 --sampled 32 --semaphores +// RUN: %{run} %t.out --type int8 --channels 4 --sampled 32 --semaphores +// RUN-IF: !cuda, %{run} %t.out --type unorm8 --channels 4 --sampled 32 --semaphores +// clang-format on + +#include "./vulkan_sycl_image_interop_read_1d_common.hpp" diff --git a/sycl/test-e2e/bindless_images/vulkan_interop/vulkan_sycl_image_interop_read_2d_channels.cpp b/sycl/test-e2e/bindless_images/vulkan_interop/vulkan_sycl_image_interop_read_2d_channels.cpp new file mode 100644 index 0000000000000..fa2017aa25683 --- /dev/null +++ b/sycl/test-e2e/bindless_images/vulkan_interop/vulkan_sycl_image_interop_read_2d_channels.cpp @@ -0,0 +1,43 @@ +// Unsampled channel-coverage version of the Vulkan/SYCL 2D image read interop +// test + +// REQUIRES: aspect-ext_oneapi_bindless_images +// REQUIRES: aspect-ext_oneapi_external_memory_import || (windows && level_zero && aspect-ext_oneapi_bindless_images) +// REQUIRES: vulkan + +// RUN: %{build} %link-vulkan -o %t.out %if target-spir %{ -Wno-ignored-attributes %} + +// UNSUPPORTED: linux +// UNSUPPORTED-TRACKER: GSD-12357 + +// clang-format off +// RUN: %{run} %t.out --type float --channels 1 32x33 +// RUN: %{run} %t.out --type float --channels 2 32x33 +// RUN: %{run} %t.out --type float --channels 4 32x33 +// RUN: %{run} %t.out --type half --channels 1 32x33 +// RUN: %{run} %t.out --type half --channels 2 32x33 +// RUN: %{run} %t.out --type half --channels 4 32x33 +// RUN: %{run} %t.out --type int32 --channels 1 32x33 +// RUN: %{run} %t.out --type int32 --channels 2 32x33 +// RUN: %{run} %t.out --type int32 --channels 4 32x33 +// RUN: %{run} %t.out --type uint32 --channels 1 32x33 +// RUN: %{run} %t.out --type uint32 --channels 2 32x33 +// RUN: %{run} %t.out --type uint32 --channels 4 32x33 +// RUN: %{run} %t.out --type int16 --channels 1 32x33 +// RUN: %{run} %t.out --type int16 --channels 2 32x33 +// RUN: %{run} %t.out --type int16 --channels 4 32x33 +// RUN: %{run} %t.out --type uint16 --channels 1 32x33 +// RUN: %{run} %t.out --type uint16 --channels 2 32x33 +// RUN: %{run} %t.out --type uint16 --channels 4 32x33 +// RUN: %{run} %t.out --type uint8 --channels 1 32x33 +// RUN: %{run} %t.out --type uint8 --channels 2 32x33 +// RUN: %{run} %t.out --type uint8 --channels 4 32x33 +// RUN: %{run} %t.out --type int8 --channels 1 32x33 +// RUN: %{run} %t.out --type int8 --channels 2 32x33 +// RUN: %{run} %t.out --type int8 --channels 4 32x33 +// RUN: %{run} %t.out --type unorm8 --channels 1 32x33 +// RUN: %{run} %t.out --type unorm8 --channels 2 32x33 +// RUN: %{run} %t.out --type unorm8 --channels 4 32x33 +// clang-format on + +#include "./vulkan_sycl_image_interop_read_2d_common.hpp" diff --git a/sycl/test-e2e/bindless_images/vulkan_interop/vulkan_sycl_image_interop_read_2d.cpp b/sycl/test-e2e/bindless_images/vulkan_interop/vulkan_sycl_image_interop_read_2d_common.hpp similarity index 77% rename from sycl/test-e2e/bindless_images/vulkan_interop/vulkan_sycl_image_interop_read_2d.cpp rename to sycl/test-e2e/bindless_images/vulkan_interop/vulkan_sycl_image_interop_read_2d_common.hpp index 50c4147d9fbe1..d73edb886ed14 100644 --- a/sycl/test-e2e/bindless_images/vulkan_interop/vulkan_sycl_image_interop_read_2d.cpp +++ b/sycl/test-e2e/bindless_images/vulkan_interop/vulkan_sycl_image_interop_read_2d_common.hpp @@ -1,94 +1,19 @@ -// REQUIRES: aspect-ext_oneapi_bindless_images -// REQUIRES: aspect-ext_oneapi_external_memory_import || (windows && level_zero && aspect-ext_oneapi_bindless_images) -// REQUIRES: vulkan - -// RUN: %{build} %link-vulkan -o %t.out %if target-spir %{ -Wno-ignored-attributes %} - -// UNSUPPORTED: linux -// UNSUPPORTED-TRACKER: GSD-12357 +// Shared implementation for the Vulkan/SYCL 2D image read interop tests. /* - Run ALL the vulkan formats through the gauntlet. sampled and unsampled. - This entire test takes less than 30 seconds on a slow machine. MUCH faster - (and more complete coveraage) than SFINAE based approach. - - IF a particular variant is having problems on some platform, please do NOT - just disable the whole test, instead use RUN~IF (SOMETHING) yadda-yadda - to enable/disable that variant. + Run ALL the vulkan formats through the gauntlet. sampled and unsampled. + This entire test takes less than 30 seconds on a slow machine. MUCH faster + (and more complete coveraage) than SFINAE based approach. - For semaphore testing, we run just a sampling. Note, that on Linux if there - is a failure in the first section, then likely ALL semaphore tests afterwards - will fail. This is being tracked as a separate issue. + IF a particular variant is having problems on some platform, please do NOT + just disable the whole test, instead use RUN~IF (SOMETHING) yadda-yadda + to enable/disable that variant. + For semaphore testing, we run just a sampling. Note, that on Linux if there + is a failure in the first section, then likely ALL semaphore tests afterwards + will fail. This is being tracked as a separate issue. */ -// RUN: %{run} %t.out --type float --channels 1 32x33 -// RUN: %{run} %t.out --type float --channels 2 32x33 -// RUN: %{run} %t.out --type float --channels 4 32x33 -// RUN: %{run} %t.out --type half --channels 1 32x33 -// RUN: %{run} %t.out --type half --channels 2 32x33 -// RUN: %{run} %t.out --type half --channels 4 32x33 -// RUN: %{run} %t.out --type int32 --channels 1 32x33 -// RUN: %{run} %t.out --type int32 --channels 2 32x33 -// RUN: %{run} %t.out --type int32 --channels 4 32x33 -// RUN: %{run} %t.out --type uint32 --channels 1 32x33 -// RUN: %{run} %t.out --type uint32 --channels 2 32x33 -// RUN: %{run} %t.out --type uint32 --channels 4 32x33 -// RUN: %{run} %t.out --type int16 --channels 1 32x33 -// RUN: %{run} %t.out --type int16 --channels 2 32x33 -// RUN: %{run} %t.out --type int16 --channels 4 32x33 -// RUN: %{run} %t.out --type uint16 --channels 1 32x33 -// RUN: %{run} %t.out --type uint16 --channels 2 32x33 -// RUN: %{run} %t.out --type uint16 --channels 4 32x33 -// RUN: %{run} %t.out --type uint8 --channels 1 32x33 -// RUN: %{run} %t.out --type uint8 --channels 2 32x33 -// RUN: %{run} %t.out --type uint8 --channels 4 32x33 -// RUN: %{run} %t.out --type int8 --channels 1 32x33 -// RUN: %{run} %t.out --type int8 --channels 2 32x33 -// RUN: %{run} %t.out --type int8 --channels 4 32x33 -// RUN: %{run} %t.out --type unorm8 --channels 1 32x33 -// RUN: %{run} %t.out --type unorm8 --channels 2 32x33 -// RUN: %{run} %t.out --type unorm8 --channels 4 32x33 -// RUN: %{run} %t.out --type float --channels 1 --sampled 32x33 -// RUN: %{run} %t.out --type float --channels 2 --sampled 32x33 -// RUN: %{run} %t.out --type float --channels 4 --sampled 32x33 -// RUN: %{run} %t.out --type half --channels 1 --sampled 32x33 -// RUN: %{run} %t.out --type half --channels 2 --sampled 32x33 -// RUN: %{run} %t.out --type half --channels 4 --sampled 32x33 -// RUN: %{run} %t.out --type int32 --channels 1 --sampled 32x33 -// RUN: %{run} %t.out --type int32 --channels 2 --sampled 32x33 -// RUN: %{run} %t.out --type int32 --channels 4 --sampled 32x33 -// RUN: %{run} %t.out --type uint32 --channels 1 --sampled 32x33 -// RUN: %{run} %t.out --type uint32 --channels 2 --sampled 32x33 -// RUN: %{run} %t.out --type uint32 --channels 4 --sampled 32x33 -// RUN: %{run} %t.out --type int16 --channels 1 --sampled 32x33 -// RUN: %{run} %t.out --type int16 --channels 2 --sampled 32x33 -// RUN: %{run} %t.out --type int16 --channels 4 --sampled 32x33 -// RUN: %{run} %t.out --type uint16 --channels 1 --sampled 32x33 -// RUN: %{run} %t.out --type uint16 --channels 2 --sampled 32x33 -// RUN: %{run} %t.out --type uint16 --channels 4 --sampled 32x33 -// RUN: %{run} %t.out --type uint8 --channels 1 --sampled 32x33 -// RUN: %{run} %t.out --type uint8 --channels 2 --sampled 32x33 -// RUN: %{run} %t.out --type uint8 --channels 4 --sampled 32x33 -// RUN: %{run} %t.out --type int8 --channels 1 --sampled 32x33 -// RUN: %{run} %t.out --type int8 --channels 2 --sampled 32x33 -// RUN: %{run} %t.out --type int8 --channels 4 --sampled 32x33 -// RUN: %{run} %t.out --type unorm8 --channels 1 --sampled 32x33 -// RUN: %{run} %t.out --type unorm8 --channels 2 --sampled 32x33 -// RUN: %{run} %t.out --type unorm8 --channels 4 --sampled 32x33 - -// RUN: %{run} %t.out --type float --channels 1 32x33 --semaphores -// RUN: %{run} %t.out --type float --channels 4 32x33 --semaphores -// RUN: %{run} %t.out --type half --channels 1 32x33 --semaphores -// RUN: %{run} %t.out --type uint32 --channels 4 32x33 --semaphores -// RUN: %{run} %t.out --type uint16 --channels 2 32x33 --semaphores -// RUN: %{run} %t.out --type int8 --channels 1 32x33 --semaphores -// RUN: %{run} %t.out --type float --channels 4 --sampled 32x33 --semaphores -// RUN: %{run} %t.out --type int32 --channels 4 --sampled 32x33 --semaphores -// RUN: %{run} %t.out --type int16 --channels 4 --sampled 32x33 --semaphores -// RUN: %{run} %t.out --type uint8 --channels 2 --sampled 32x33 --semaphores -// RUN: %{run} %t.out --type unorm8 --channels 4 --sampled 32x33 --semaphores - // clang-format off /* Vulkan/SYCL 2D Image Read Test (Sampled + Unsampled) @@ -120,6 +45,8 @@ */ // clang-format on +#pragma once + #include "vulkan_setup.hpp" #include @@ -548,4 +475,4 @@ int main(int argc, char **argv) { std::cerr << "Unknown type: " << type << std::endl; return 1; -} \ No newline at end of file +} diff --git a/sycl/test-e2e/bindless_images/vulkan_interop/vulkan_sycl_image_interop_read_2d_sampled.cpp b/sycl/test-e2e/bindless_images/vulkan_interop/vulkan_sycl_image_interop_read_2d_sampled.cpp new file mode 100644 index 0000000000000..868e79fa228d1 --- /dev/null +++ b/sycl/test-e2e/bindless_images/vulkan_interop/vulkan_sycl_image_interop_read_2d_sampled.cpp @@ -0,0 +1,42 @@ +// Sampled-only version of the Vulkan/SYCL 2D image read interop test. + +// REQUIRES: aspect-ext_oneapi_bindless_images +// REQUIRES: aspect-ext_oneapi_external_memory_import || (windows && level_zero && aspect-ext_oneapi_bindless_images) +// REQUIRES: vulkan + +// RUN: %{build} %link-vulkan -o %t.out %if target-spir %{ -Wno-ignored-attributes %} + +// UNSUPPORTED: linux +// UNSUPPORTED-TRACKER: GSD-12357 + +// clang-format off +// RUN: %{run} %t.out --type float --channels 1 --sampled 32x33 +// RUN: %{run} %t.out --type float --channels 2 --sampled 32x33 +// RUN: %{run} %t.out --type float --channels 4 --sampled 32x33 +// RUN: %{run} %t.out --type half --channels 1 --sampled 32x33 +// RUN: %{run} %t.out --type half --channels 2 --sampled 32x33 +// RUN: %{run} %t.out --type half --channels 4 --sampled 32x33 +// RUN: %{run} %t.out --type int32 --channels 1 --sampled 32x33 +// RUN: %{run} %t.out --type int32 --channels 2 --sampled 32x33 +// RUN: %{run} %t.out --type int32 --channels 4 --sampled 32x33 +// RUN: %{run} %t.out --type uint32 --channels 1 --sampled 32x33 +// RUN: %{run} %t.out --type uint32 --channels 2 --sampled 32x33 +// RUN: %{run} %t.out --type uint32 --channels 4 --sampled 32x33 +// RUN: %{run} %t.out --type int16 --channels 1 --sampled 32x33 +// RUN: %{run} %t.out --type int16 --channels 2 --sampled 32x33 +// RUN: %{run} %t.out --type int16 --channels 4 --sampled 32x33 +// RUN: %{run} %t.out --type uint16 --channels 1 --sampled 32x33 +// RUN: %{run} %t.out --type uint16 --channels 2 --sampled 32x33 +// RUN: %{run} %t.out --type uint16 --channels 4 --sampled 32x33 +// RUN: %{run} %t.out --type uint8 --channels 1 --sampled 32x33 +// RUN: %{run} %t.out --type uint8 --channels 2 --sampled 32x33 +// RUN: %{run} %t.out --type uint8 --channels 4 --sampled 32x33 +// RUN: %{run} %t.out --type int8 --channels 1 --sampled 32x33 +// RUN: %{run} %t.out --type int8 --channels 2 --sampled 32x33 +// RUN: %{run} %t.out --type int8 --channels 4 --sampled 32x33 +// RUN: %{run} %t.out --type unorm8 --channels 1 --sampled 32x33 +// RUN: %{run} %t.out --type unorm8 --channels 2 --sampled 32x33 +// RUN: %{run} %t.out --type unorm8 --channels 4 --sampled 32x33 +// clang-format on + +#include "./vulkan_sycl_image_interop_read_2d_common.hpp" diff --git a/sycl/test-e2e/bindless_images/vulkan_interop/vulkan_sycl_image_interop_read_2d_semaphores.cpp b/sycl/test-e2e/bindless_images/vulkan_interop/vulkan_sycl_image_interop_read_2d_semaphores.cpp new file mode 100644 index 0000000000000..6068fa1fc8c94 --- /dev/null +++ b/sycl/test-e2e/bindless_images/vulkan_interop/vulkan_sycl_image_interop_read_2d_semaphores.cpp @@ -0,0 +1,26 @@ +// Semaphore version of the Vulkan/SYCL 2D image read interop test. + +// REQUIRES: aspect-ext_oneapi_bindless_images +// REQUIRES: aspect-ext_oneapi_external_memory_import || (windows && level_zero && aspect-ext_oneapi_bindless_images) +// REQUIRES: vulkan + +// RUN: %{build} %link-vulkan -o %t.out %if target-spir %{ -Wno-ignored-attributes %} + +// UNSUPPORTED: linux +// UNSUPPORTED-TRACKER: GSD-12357 + +// clang-format off +// RUN: %{run} %t.out --type float --channels 1 32x33 --semaphores +// RUN: %{run} %t.out --type float --channels 4 32x33 --semaphores +// RUN: %{run} %t.out --type half --channels 1 32x33 --semaphores +// RUN: %{run} %t.out --type uint32 --channels 4 32x33 --semaphores +// RUN: %{run} %t.out --type uint16 --channels 2 32x33 --semaphores +// RUN: %{run} %t.out --type int8 --channels 1 32x33 --semaphores +// RUN: %{run} %t.out --type float --channels 4 --sampled 32x33 --semaphores +// RUN: %{run} %t.out --type int32 --channels 4 --sampled 32x33 --semaphores +// RUN: %{run} %t.out --type int16 --channels 4 --sampled 32x33 --semaphores +// RUN: %{run} %t.out --type uint8 --channels 2 --sampled 32x33 --semaphores +// RUN: %{run} %t.out --type unorm8 --channels 4 --sampled 32x33 --semaphores +// clang-format on + +#include "./vulkan_sycl_image_interop_read_2d_common.hpp" diff --git a/sycl/test-e2e/bindless_images/vulkan_interop/vulkan_sycl_image_interop_write_1d_unsampled_channels.cpp b/sycl/test-e2e/bindless_images/vulkan_interop/vulkan_sycl_image_interop_write_1d_unsampled_channels.cpp new file mode 100644 index 0000000000000..dccfa78eb2dc2 --- /dev/null +++ b/sycl/test-e2e/bindless_images/vulkan_interop/vulkan_sycl_image_interop_write_1d_unsampled_channels.cpp @@ -0,0 +1,42 @@ +// Channel coverage version of the Vulkan/SYCL 1D unsampled write interop test. + +// REQUIRES: aspect-ext_oneapi_bindless_images +// REQUIRES: aspect-ext_oneapi_external_memory_import || (windows && level_zero && aspect-ext_oneapi_bindless_images) +// REQUIRES: vulkan + +// UNSUPPORTED: windows +// UNSUPPORTED-TRACKER: CMPLRLLVM-73525 + +// RUN: %{build} %link-vulkan -o %t.out %if target-spir %{ -Wno-ignored-attributes %} + +// clang-format off +// RUN: %{run} %t.out --type float --channels 1 32 +// RUN: %{run} %t.out --type float --channels 2 32 +// RUN: %{run} %t.out --type float --channels 4 32 +// RUN: %{run} %t.out --type half --channels 1 32 +// RUN: %{run} %t.out --type half --channels 2 32 +// RUN: %{run} %t.out --type half --channels 4 32 +// RUN: %{run} %t.out --type int32 --channels 1 32 +// RUN: %{run} %t.out --type int32 --channels 2 32 +// RUN: %{run} %t.out --type int32 --channels 4 32 +// RUN: %{run} %t.out --type uint32 --channels 1 32 +// RUN: %{run} %t.out --type uint32 --channels 2 32 +// RUN: %{run} %t.out --type uint32 --channels 4 32 +// RUN: %{run} %t.out --type int16 --channels 1 32 +// RUN: %{run} %t.out --type int16 --channels 2 32 +// RUN: %{run} %t.out --type int16 --channels 4 32 +// RUN: %{run} %t.out --type uint16 --channels 1 32 +// RUN: %{run} %t.out --type uint16 --channels 2 32 +// RUN: %{run} %t.out --type uint16 --channels 4 32 +// RUN: %{run} %t.out --type uint8 --channels 1 32 +// RUN: %{run} %t.out --type uint8 --channels 2 32 +// RUN: %{run} %t.out --type uint8 --channels 4 32 +// RUN: %{run} %t.out --type int8 --channels 1 32 +// RUN: %{run} %t.out --type int8 --channels 2 32 +// RUN: %{run} %t.out --type int8 --channels 4 32 +// RUN-IF: !cuda, %{run} %t.out --type unorm8 --channels 1 32 +// RUN-IF: !cuda, %{run} %t.out --type unorm8 --channels 2 32 +// RUN-IF: !cuda, %{run} %t.out --type unorm8 --channels 4 32 +// clang-format on + +#include "./vulkan_sycl_image_interop_write_1d_unsampled_common.hpp" diff --git a/sycl/test-e2e/bindless_images/vulkan_interop/vulkan_sycl_image_interop_write_1d_unsampled.cpp b/sycl/test-e2e/bindless_images/vulkan_interop/vulkan_sycl_image_interop_write_1d_unsampled_common.hpp similarity index 85% rename from sycl/test-e2e/bindless_images/vulkan_interop/vulkan_sycl_image_interop_write_1d_unsampled.cpp rename to sycl/test-e2e/bindless_images/vulkan_interop/vulkan_sycl_image_interop_write_1d_unsampled_common.hpp index aa8532eaf61a5..2b790b9a91bed 100644 --- a/sycl/test-e2e/bindless_images/vulkan_interop/vulkan_sycl_image_interop_write_1d_unsampled.cpp +++ b/sycl/test-e2e/bindless_images/vulkan_interop/vulkan_sycl_image_interop_write_1d_unsampled_common.hpp @@ -1,12 +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 - -// UNSUPPORTED: windows -// UNSUPPORTED-TRACKER: CMPLRLLVM-73525 - -// RUN: %{build} %link-vulkan -o %t.out %if target-spir %{ -Wno-ignored-attributes %} +// Shared implementation for the Vulkan/SYCL 1D unsampled write interop tests. /* Run all the vulkan formats through a write test. Note this is unsampled only, @@ -14,82 +6,41 @@ IF a particular variant is having problems on some platform, please do NOT just disable the whole test, instead use RUN~IF: (SOMETHING) yadda-yadda - to enable/disable that variant. + to enable/disable that variant. - For semaphore testing, we run just a sampling. Note, that on Linux if there + For semaphore testing, we run just a sampling. Note, that on Linux if there is a failure in the first section, then likely ALL semaphore tests afterwards will fail. This is being tracked as a separate issue. - */ -// clang-format off - -// RUN: %{run} %t.out --type float --channels 1 32 -// RUN: %{run} %t.out --type float --channels 2 32 -// RUN: %{run} %t.out --type float --channels 4 32 -// RUN: %{run} %t.out --type half --channels 1 32 -// RUN: %{run} %t.out --type half --channels 2 32 -// RUN: %{run} %t.out --type half --channels 4 32 -// RUN: %{run} %t.out --type int32 --channels 1 32 -// RUN: %{run} %t.out --type int32 --channels 2 32 -// RUN: %{run} %t.out --type int32 --channels 4 32 -// RUN: %{run} %t.out --type uint32 --channels 1 32 -// RUN: %{run} %t.out --type uint32 --channels 2 32 -// RUN: %{run} %t.out --type uint32 --channels 4 32 -// RUN: %{run} %t.out --type int16 --channels 1 32 -// RUN: %{run} %t.out --type int16 --channels 2 32 -// RUN: %{run} %t.out --type int16 --channels 4 32 -// RUN: %{run} %t.out --type uint16 --channels 1 32 -// RUN: %{run} %t.out --type uint16 --channels 2 32 -// RUN: %{run} %t.out --type uint16 --channels 4 32 -// RUN: %{run} %t.out --type uint8 --channels 1 32 -// RUN: %{run} %t.out --type uint8 --channels 2 32 -// RUN: %{run} %t.out --type uint8 --channels 4 32 -// RUN: %{run} %t.out --type int8 --channels 1 32 -// RUN: %{run} %t.out --type int8 --channels 2 32 -// RUN: %{run} %t.out --type int8 --channels 4 32 -// RUN-IF: !cuda, %{run} %t.out --type unorm8 --channels 1 32 -// RUN-IF: !cuda, %{run} %t.out --type unorm8 --channels 2 32 -// RUN-IF: !cuda, %{run} %t.out --type unorm8 --channels 4 32 - -// On Linux L0, there are problem with semaphores and latest drivers. -// GSD-12371 GSD-12339 - -// RUN-IF: !level_zero, %{run} %t.out --type float --channels 1 32 --semaphores -// RUN-IF: !level_zero, %{run} %t.out --type half --channels 2 32 --semaphores -// RUN-IF: !level_zero, %{run} %t.out --type int32 --channels 4 32 --semaphores -// RUN-IF: !level_zero, %{run} %t.out --type uint32 --channels 1 32 --semaphores -// RUN-IF: !level_zero, %{run} %t.out --type int16 --channels 2 32 --semaphores -// RUN-IF: !level_zero, %{run} %t.out --type uint16 --channels 4 32 --semaphores -// RUN-IF: !level_zero, %{run} %t.out --type uint8 --channels 1 32 --semaphores -// RUN-IF: !level_zero, %{run} %t.out --type int8 --channels 4 32 --semaphores -// CUDA doesn't support unorm8, level_zero has issues with semaphores -// XXX-IF: !cuda, %{run} %t.out --type unorm8 --channels 2 32 --semaphores - +// clang-format off /* - Vulkan/SYCL 1D Unsampled Write Image - + Vulkan/SYCL 1D Unsampled Write Image clang++ -fsycl -o vsw_1d_test.bin vulkan_sycl_image_interop_write_1d_unsampled.cpp -lvulkan -I$VULKAN_SDK/include -L$VULKAN_SDK/lib clang++ -fsycl -o vsw_1d_test.exe vulkan_sycl_image_interop_write_1d_unsampled.cpp -Wno-ignored-attributes -lvulkan-1 -I$VULKAN_SDK/Include -L$VULKAN_SDK/Lib - + USAGE: ./vsw_1d_test.bin ./vsw_1d_test.bin --semaphores - FLAGS + + FLAGS: --sampled ERROR: Sampled image writes are not supported --semaphores Use Vulkan Semaphores for SYCL Interop Sync --linear Use LINEAR tiling for the Vulkan Image (default is OPTIMAL) --channels X Set number of channels (1, 2, or 4). Default is 4 (RGBA) --type XXX Set data type (float, half, uint32, int32, uint16, int16, uint8, int8, unorm8). Default is float Wx Set custom Width . Put "x" after - + + EXAMPLES: ./vsw_1d_test.bin --semaphores --channels 2 --linear 64x - */ +*/ // clang-format on +#pragma once + #include "vulkan_setup.hpp" #include @@ -464,4 +415,4 @@ int main(int argc, char **argv) { getUnorm8Format(channels), sycl::image_channel_type::unorm_int8); return 1; -} \ No newline at end of file +} diff --git a/sycl/test-e2e/bindless_images/vulkan_interop/vulkan_sycl_image_interop_write_1d_unsampled_semaphores.cpp b/sycl/test-e2e/bindless_images/vulkan_interop/vulkan_sycl_image_interop_write_1d_unsampled_semaphores.cpp new file mode 100644 index 0000000000000..35a1384c1aeca --- /dev/null +++ b/sycl/test-e2e/bindless_images/vulkan_interop/vulkan_sycl_image_interop_write_1d_unsampled_semaphores.cpp @@ -0,0 +1,26 @@ +// Semaphore coverage version of the Vulkan/SYCL 1D unsampled write interop +// test. + +// REQUIRES: aspect-ext_oneapi_bindless_images +// REQUIRES: aspect-ext_oneapi_external_memory_import || (windows && level_zero && aspect-ext_oneapi_bindless_images) +// REQUIRES: vulkan + +// UNSUPPORTED: windows +// UNSUPPORTED-TRACKER: CMPLRLLVM-73525 + +// RUN: %{build} %link-vulkan -o %t.out %if target-spir %{ -Wno-ignored-attributes %} + +// clang-format off +// RUN-IF: !level_zero, %{run} %t.out --type float --channels 1 32 --semaphores +// RUN-IF: !level_zero, %{run} %t.out --type half --channels 2 32 --semaphores +// RUN-IF: !level_zero, %{run} %t.out --type int32 --channels 4 32 --semaphores +// RUN-IF: !level_zero, %{run} %t.out --type uint32 --channels 1 32 --semaphores +// RUN-IF: !level_zero, %{run} %t.out --type int16 --channels 2 32 --semaphores +// RUN-IF: !level_zero, %{run} %t.out --type uint16 --channels 4 32 --semaphores +// RUN-IF: !level_zero, %{run} %t.out --type uint8 --channels 1 32 --semaphores +// RUN-IF: !level_zero, %{run} %t.out --type int8 --channels 4 32 --semaphores +// CUDA doesn't support unorm8, level_zero has issues with semaphores +// XXX-IF: !cuda, %{run} %t.out --type unorm8 --channels 2 32 --semaphores +// clang-format on + +#include "./vulkan_sycl_image_interop_write_1d_unsampled_common.hpp" diff --git a/sycl/test-e2e/bindless_images/vulkan_interop/vulkan_sycl_image_interop_write_2d_unsampled_channels.cpp b/sycl/test-e2e/bindless_images/vulkan_interop/vulkan_sycl_image_interop_write_2d_unsampled_channels.cpp new file mode 100644 index 0000000000000..21c8e9a52260d --- /dev/null +++ b/sycl/test-e2e/bindless_images/vulkan_interop/vulkan_sycl_image_interop_write_2d_unsampled_channels.cpp @@ -0,0 +1,42 @@ +// Channel coverage version of the Vulkan/SYCL 2D unsampled write interop test. + +// REQUIRES: aspect-ext_oneapi_bindless_images +// REQUIRES: aspect-ext_oneapi_external_memory_import || (windows && level_zero && aspect-ext_oneapi_bindless_images) +// REQUIRES: vulkan + +// RUN: %{build} %link-vulkan -o %t.out %if target-spir %{ -Wno-ignored-attributes %} + +// UNSUPPORTED: linux +// UNSUPPORTED-TRACKER: GSD-12357 + +// clang-format off +// RUN: %{run} %t.out --type float --channels 1 32x33 +// RUN: %{run} %t.out --type float --channels 2 32x33 +// RUN: %{run} %t.out --type float --channels 4 32x33 +// RUN: %{run} %t.out --type half --channels 1 32x33 +// RUN: %{run} %t.out --type half --channels 2 32x33 +// RUN: %{run} %t.out --type half --channels 4 32x33 +// RUN: %{run} %t.out --type int32 --channels 1 32x33 +// RUN: %{run} %t.out --type int32 --channels 2 32x33 +// RUN: %{run} %t.out --type int32 --channels 4 32x33 +// RUN: %{run} %t.out --type uint32 --channels 1 32x33 +// RUN: %{run} %t.out --type uint32 --channels 2 32x33 +// RUN: %{run} %t.out --type uint32 --channels 4 32x33 +// RUN: %{run} %t.out --type int16 --channels 1 32x33 +// RUN: %{run} %t.out --type int16 --channels 2 32x33 +// RUN: %{run} %t.out --type int16 --channels 4 32x33 +// RUN: %{run} %t.out --type uint16 --channels 1 32x33 +// RUN: %{run} %t.out --type uint16 --channels 2 32x33 +// RUN: %{run} %t.out --type uint16 --channels 4 32x33 +// RUN: %{run} %t.out --type uint8 --channels 1 32x33 +// RUN: %{run} %t.out --type uint8 --channels 2 32x33 +// RUN: %{run} %t.out --type uint8 --channels 4 32x33 +// RUN: %{run} %t.out --type int8 --channels 1 32x33 +// RUN: %{run} %t.out --type int8 --channels 2 32x33 +// RUN: %{run} %t.out --type int8 --channels 4 32x33 +// RUN: %{run} %t.out --type unorm8 --channels 1 32x33 +// RUN: %{run} %t.out --type unorm8 --channels 2 32x33 +// RUN: %{run} %t.out --type unorm8 --channels 4 32x33 +// clang-format on + +#include "./vulkan_sycl_image_interop_write_2d_unsampled_common.hpp" diff --git a/sycl/test-e2e/bindless_images/vulkan_interop/vulkan_sycl_image_interop_write_2d_unsampled.cpp b/sycl/test-e2e/bindless_images/vulkan_interop/vulkan_sycl_image_interop_write_2d_unsampled_common.hpp similarity index 85% rename from sycl/test-e2e/bindless_images/vulkan_interop/vulkan_sycl_image_interop_write_2d_unsampled.cpp rename to sycl/test-e2e/bindless_images/vulkan_interop/vulkan_sycl_image_interop_write_2d_unsampled_common.hpp index 7bda101fdd4b6..fab1d97eefc66 100644 --- a/sycl/test-e2e/bindless_images/vulkan_interop/vulkan_sycl_image_interop_write_2d_unsampled.cpp +++ b/sycl/test-e2e/bindless_images/vulkan_interop/vulkan_sycl_image_interop_write_2d_unsampled_common.hpp @@ -1,63 +1,19 @@ -// REQUIRES: aspect-ext_oneapi_bindless_images -// REQUIRES: aspect-ext_oneapi_external_memory_import || (windows && level_zero && aspect-ext_oneapi_bindless_images) -// REQUIRES: vulkan - -// RUN: %{build} %link-vulkan -o %t.out %if target-spir %{ -Wno-ignored-attributes %} - -// UNSUPPORTED: linux -// UNSUPPORTED-TRACKER: GSD-12357 +// Shared implementation for the Vulkan/SYCL 2D unsampled write interop tests. /* - Run all the vulkan formats through a write test. Note this is unsampled - only, you can't "write" with the image sampler. - IF a particular variant is having problems on some platform, please do NOT - just disable the whole test, instead use RUN~IF: (SOMETHING) yadda-yadda - to enable/disable that variant. + Run all the vulkan formats through a write test. Note this is unsampled + only, you can't "write" with the image sampler. - For semaphore testing, we run just a sampling. Note, that on Linux if there - is a failure in the first section, then likely ALL semaphore tests afterwards - will fail. This is being tracked as a separate issue. + IF a particular variant is having problems on some platform, please do NOT + just disable the whole test, instead use RUN~IF: (SOMETHING) yadda-yadda + to enable/disable that variant. -*/ + For semaphore testing, we run just a sampling. Note, that on Linux if there + is a failure in the first section, then likely ALL semaphore tests afterwards + will fail. This is being tracked as a separate issue. -// RUN: %{run} %t.out --type float --channels 1 32x33 -// RUN: %{run} %t.out --type float --channels 2 32x33 -// RUN: %{run} %t.out --type float --channels 4 32x33 -// RUN: %{run} %t.out --type half --channels 1 32x33 -// RUN: %{run} %t.out --type half --channels 2 32x33 -// RUN: %{run} %t.out --type half --channels 4 32x33 -// RUN: %{run} %t.out --type int32 --channels 1 32x33 -// RUN: %{run} %t.out --type int32 --channels 2 32x33 -// RUN: %{run} %t.out --type int32 --channels 4 32x33 -// RUN: %{run} %t.out --type uint32 --channels 1 32x33 -// RUN: %{run} %t.out --type uint32 --channels 2 32x33 -// RUN: %{run} %t.out --type uint32 --channels 4 32x33 -// RUN: %{run} %t.out --type int16 --channels 1 32x33 -// RUN: %{run} %t.out --type int16 --channels 2 32x33 -// RUN: %{run} %t.out --type int16 --channels 4 32x33 -// RUN: %{run} %t.out --type uint16 --channels 1 32x33 -// RUN: %{run} %t.out --type uint16 --channels 2 32x33 -// RUN: %{run} %t.out --type uint16 --channels 4 32x33 -// RUN: %{run} %t.out --type uint8 --channels 1 32x33 -// RUN: %{run} %t.out --type uint8 --channels 2 32x33 -// RUN: %{run} %t.out --type uint8 --channels 4 32x33 -// RUN: %{run} %t.out --type int8 --channels 1 32x33 -// RUN: %{run} %t.out --type int8 --channels 2 32x33 -// RUN: %{run} %t.out --type int8 --channels 4 32x33 -// RUN: %{run} %t.out --type unorm8 --channels 1 32x33 -// RUN: %{run} %t.out --type unorm8 --channels 2 32x33 -// RUN: %{run} %t.out --type unorm8 --channels 4 32x33 - -// RUN: %{run} %t.out --type float --channels 1 32x33 --semaphores -// RUN: %{run} %t.out --type half --channels 2 32x33 --semaphores -// RUN: %{run} %t.out --type int32 --channels 4 32x33 --semaphores -// RUN: %{run} %t.out --type uint32 --channels 1 32x33 --semaphores -// RUN: %{run} %t.out --type int16 --channels 2 32x33 --semaphores -// RUN: %{run} %t.out --type uint16 --channels 4 32x33 --semaphores -// RUN: %{run} %t.out --type uint8 --channels 1 32x33 --semaphores -// RUN: %{run} %t.out --type int8 --channels 2 32x33 --semaphores -// RUN: %{run} %t.out --type unorm8 --channels 4 32x33 --semaphores +*/ // clang-format off /* @@ -65,9 +21,10 @@ clang++ -fsycl -o vsw_2d_test.exe vulkan_sycl_image_interop_write_2d_unsampled.cpp -Wno-ignored-attributes -lvulkan-1 -I$VULKAN_SDK/Include -L$VULKAN_SDK/Lib + USAGE: ./vsw_2d_test.bin - FLAGS + FLAGS: --sampled ERROR: Sampled image writes are not supported --semaphores Use Vulkan Semaphores for SYCL Interop Sync --linear Use LINEAR tiling for the Vulkan Image (default is OPTIMAL) @@ -76,11 +33,14 @@ Default is float WxH Set custom Width x Height (e.g. 8x4) + EXAMPLES: ./vsw_2d_test.bin --semaphores --channels 2 --linear 8x4 */ // clang-format on +#pragma once + #include "vulkan_setup.hpp" #include @@ -475,4 +435,4 @@ int main(int argc, char **argv) { getUnorm8Format(channels), sycl::image_channel_type::unorm_int8); return 1; -} \ No newline at end of file +} diff --git a/sycl/test-e2e/bindless_images/vulkan_interop/vulkan_sycl_image_interop_write_2d_unsampled_semaphores.cpp b/sycl/test-e2e/bindless_images/vulkan_interop/vulkan_sycl_image_interop_write_2d_unsampled_semaphores.cpp new file mode 100644 index 0000000000000..b587871577c98 --- /dev/null +++ b/sycl/test-e2e/bindless_images/vulkan_interop/vulkan_sycl_image_interop_write_2d_unsampled_semaphores.cpp @@ -0,0 +1,25 @@ +// Semaphore coverage version of the Vulkan/SYCL 2D unsampled write interop +// test. + +// REQUIRES: aspect-ext_oneapi_bindless_images +// REQUIRES: aspect-ext_oneapi_external_memory_import || (windows && level_zero && aspect-ext_oneapi_bindless_images) +// REQUIRES: vulkan + +// RUN: %{build} %link-vulkan -o %t.out %if target-spir %{ -Wno-ignored-attributes %} + +// UNSUPPORTED: linux +// UNSUPPORTED-TRACKER: GSD-12357 + +// clang-format off +// RUN: %{run} %t.out --type float --channels 1 32x33 --semaphores +// RUN: %{run} %t.out --type half --channels 2 32x33 --semaphores +// RUN: %{run} %t.out --type int32 --channels 4 32x33 --semaphores +// RUN: %{run} %t.out --type uint32 --channels 1 32x33 --semaphores +// RUN: %{run} %t.out --type int16 --channels 2 32x33 --semaphores +// RUN: %{run} %t.out --type uint16 --channels 4 32x33 --semaphores +// RUN: %{run} %t.out --type uint8 --channels 1 32x33 --semaphores +// RUN: %{run} %t.out --type int8 --channels 2 32x33 --semaphores +// RUN: %{run} %t.out --type unorm8 --channels 4 32x33 --semaphores +// clang-format on + +#include "./vulkan_sycl_image_interop_write_2d_unsampled_common.hpp"