[SYCL][Bindless][E2E] Improve bindless images E2E test times (#15490)

A significant number of bindless images tests use image sizes that are
far larger then necessary. This commit reduces them to decrease times
that bindless images tests take to complete and adds deterministic
varying of local and global sizes in bindless tests.

Author: DBDuncan

sycl/test-e2e/bindless_images/array/fetch_sampled_array.cpp

@@ -616,9 +616,9 @@ int main() {
 
   unsigned int seed = 0;
   std::cout << "Running 1D Sampled Image Array Tests!\n";
-  bool result1D = runAll<2>({64, 32}, {16, 16}, 20, seed);
+  bool result1D = runAll<2>({16, 16}, {8, 8}, 20, seed);
   std::cout << "Running 2D Sampled Image Array Tests!\n";
-  bool result2D = runAll<3>({64, 32, 16}, {16, 16, 4}, 20, seed);
+  bool result2D = runAll<3>({16, 8, 4}, {8, 8, 2}, 20, seed);
 
   if (result1D && result2D) {
     std::cout << "All tests passed!\n";

sycl/test-e2e/bindless_images/array/read_sampled_array.cpp

@@ -171,99 +171,148 @@ void printTestName(std::string name) {
 #endif
 }
 
+sycl::range<2> getCubeGlobal(size_t index) {
+  const std::vector<sycl::range<2>> globalSizes = {{6, 6}, {10, 10}, {8, 8}};
+
+  return globalSizes[index % 3];
+}
+
+sycl::range<3> getCubeLocal(size_t index) {
+
+  const std::vector<sycl::range<3>> localSizes = {
+      {3, 3, 1}, {5, 5, 3}, {4, 4, 2}};
+
+  return localSizes[index % 3];
+}
+
 int main() {
 
   unsigned int seed = 0;
   bool failed = false;
+  int sizeIndex = 0;
 
   printTestName("Running cube int\n");
   failed |= run_test<int32_t, 1, sycl::image_channel_type::signed_int32,
-                     class int_cube>({32, 32}, {16, 16, 2}, seed);
+                     class int_cube>(getCubeGlobal(sizeIndex),
+                                     getCubeLocal(sizeIndex), seed);
+  sizeIndex++;
   printTestName("Running cube int2\n");
   failed |= run_test<int32_t, 2, sycl::image_channel_type::signed_int32,
-                     class int2_cube>({128, 128}, {16, 16, 3}, seed);
+                     class int2_cube>(getCubeGlobal(sizeIndex),
+                                      getCubeLocal(sizeIndex), seed);
+  sizeIndex++;
   printTestName("Running cube int4\n");
   failed |= run_test<int32_t, 4, sycl::image_channel_type::signed_int32,
-                     class int4_cube>({64, 64}, {32, 16, 1}, seed);
-
+                     class int4_cube>(getCubeGlobal(sizeIndex),
+                                      getCubeLocal(sizeIndex), seed);
+  sizeIndex++;
   printTestName("Running cube unsigned int\n");
   failed |= run_test<uint32_t, 1, sycl::image_channel_type::unsigned_int32,
-                     class uint_cube>({15, 15}, {5, 3, 1}, seed);
+                     class uint_cube>(getCubeGlobal(sizeIndex),
+                                      getCubeLocal(sizeIndex), seed);
+  sizeIndex++;
   printTestName("Running cube unsigned int2\n");
   failed |= run_test<uint32_t, 2, sycl::image_channel_type::unsigned_int32,
-                     class uint2_cube>({90, 90}, {10, 9, 3}, seed);
+                     class uint2_cube>(getCubeGlobal(sizeIndex),
+                                       getCubeLocal(sizeIndex), seed);
+  sizeIndex++;
   printTestName("Running cube unsigned int4\n");
   failed |= run_test<uint32_t, 4, sycl::image_channel_type::unsigned_int32,
-                     class uint4_cube>({1024, 1024}, {16, 16, 2}, seed);
-
+                     class uint4_cube>(getCubeGlobal(sizeIndex),
+                                       getCubeLocal(sizeIndex), seed);
+  sizeIndex++;
   printTestName("Running cube short\n");
   failed |= run_test<short, 1, sycl::image_channel_type::signed_int16,
-                     class short_cube>({8, 8}, {2, 2, 1}, seed);
+                     class short_cube>(getCubeGlobal(sizeIndex),
+                                       getCubeLocal(sizeIndex), seed);
+  sizeIndex++;
   printTestName("Running cube short2\n");
   failed |= run_test<short, 2, sycl::image_channel_type::signed_int16,
-                     class short2_cube>({8, 8}, {4, 4, 2}, seed);
+                     class short2_cube>(getCubeGlobal(sizeIndex),
+                                        getCubeLocal(sizeIndex), seed);
+  sizeIndex++;
   printTestName("Running cube short4\n");
   failed |= run_test<short, 4, sycl::image_channel_type::signed_int16,
-                     class short4_cube>({8, 8}, {8, 8, 3}, seed);
-
+                     class short4_cube>(getCubeGlobal(sizeIndex),
+                                        getCubeLocal(sizeIndex), seed);
+  sizeIndex++;
   printTestName("Running cube unsigned short\n");
   failed |=
       run_test<unsigned short, 1, sycl::image_channel_type::unsigned_int16,
-               class ushort_cube>({75, 75}, {25, 5, 1}, seed);
+               class ushort_cube>(getCubeGlobal(sizeIndex),
+                                  getCubeLocal(sizeIndex), seed);
+  sizeIndex++;
   printTestName("Running cube unsigned short2\n");
   failed |=
       run_test<unsigned short, 2, sycl::image_channel_type::unsigned_int16,
-               class ushort2_cube>({75, 75}, {15, 3, 2}, seed);
+               class ushort2_cube>(getCubeGlobal(sizeIndex),
+                                   getCubeLocal(sizeIndex), seed);
+  sizeIndex++;
   printTestName("Running cube unsigned short4\n");
   failed |=
       run_test<unsigned short, 4, sycl::image_channel_type::unsigned_int16,
-               class ushort4_cube>({75, 75}, {5, 25, 3}, seed);
-
+               class ushort4_cube>(getCubeGlobal(sizeIndex),
+                                   getCubeLocal(sizeIndex), seed);
+  sizeIndex++;
   printTestName("Running cube char\n");
   failed |= run_test<signed char, 1, sycl::image_channel_type::signed_int8,
-                     class char_cube>({60, 60}, {10, 6, 1}, seed);
+                     class char_cube>(getCubeGlobal(sizeIndex),
+                                      getCubeLocal(sizeIndex), seed);
+  sizeIndex++;
   printTestName("Running cube char2\n");
   failed |= run_test<signed char, 2, sycl::image_channel_type::signed_int8,
-                     class char2_cube>({60, 60}, {5, 3, 2}, seed);
+                     class char2_cube>(getCubeGlobal(sizeIndex),
+                                       getCubeLocal(sizeIndex), seed);
+  sizeIndex++;
   printTestName("Running cube char4\n");
   failed |= run_test<signed char, 4, sycl::image_channel_type::signed_int8,
-                     class char4_cube>({60, 60}, {6, 10, 3}, seed);
-
+                     class char4_cube>(getCubeGlobal(sizeIndex),
+                                       getCubeLocal(sizeIndex), seed);
+  sizeIndex++;
   printTestName("Running cube unsigned char\n");
   failed |= run_test<unsigned char, 1, sycl::image_channel_type::unsigned_int8,
-                     class uchar_cube>({128, 128}, {16, 16, 3}, seed);
+                     class uchar_cube>(getCubeGlobal(sizeIndex),
+                                       getCubeLocal(sizeIndex), seed);
+  sizeIndex++;
   printTestName("Running cube unsigned char2\n");
   failed |= run_test<unsigned char, 2, sycl::image_channel_type::unsigned_int8,
-                     class uchar2_cube>({128, 128}, {16, 16, 3}, seed);
+                     class uchar2_cube>(getCubeGlobal(sizeIndex),
+                                        getCubeLocal(sizeIndex), seed);
+  sizeIndex++;
   printTestName("Running cube unsigned char4\n");
   failed |= run_test<unsigned char, 4, sycl::image_channel_type::unsigned_int8,
-                     class uchar4_cube>({128, 128}, {16, 16, 3}, seed);
-
+                     class uchar4_cube>(getCubeGlobal(sizeIndex),
+                                        getCubeLocal(sizeIndex), seed);
+  sizeIndex++;
   printTestName("Running cube float\n");
   failed |=
       run_test<float, 1, sycl::image_channel_type::fp32, class float_cube>(
-          {1024, 1024}, {16, 16, 1}, seed);
+          getCubeGlobal(sizeIndex), getCubeLocal(sizeIndex), seed);
+  sizeIndex++;
   printTestName("Running cube float2\n");
   failed |=
       run_test<float, 2, sycl::image_channel_type::fp32, class float2_cube>(
-          {1024, 1024}, {16, 16, 3}, seed);
+          getCubeGlobal(sizeIndex), getCubeLocal(sizeIndex), seed);
+  sizeIndex++;
   printTestName("Running cube float4\n");
   failed |=
       run_test<float, 4, sycl::image_channel_type::fp32, class float4_cube>(
-          {1024, 1024}, {16, 16, 2}, seed);
-
+          getCubeGlobal(sizeIndex), getCubeLocal(sizeIndex), seed);
+  sizeIndex++;
   printTestName("Running cube half\n");
   failed |=
       run_test<sycl::half, 1, sycl::image_channel_type::fp16, class half_cube>(
-          {48, 48}, {8, 8, 1}, seed);
+          getCubeGlobal(sizeIndex), getCubeLocal(sizeIndex), seed);
+  sizeIndex++;
   printTestName("Running cube half2\n");
   failed |=
       run_test<sycl::half, 2, sycl::image_channel_type::fp16, class half2_cube>(
-          {48, 48}, {8, 8, 3}, seed);
+          getCubeGlobal(sizeIndex), getCubeLocal(sizeIndex), seed);
+  sizeIndex++;
   printTestName("Running cube half4\n");
   failed |=
       run_test<sycl::half, 4, sycl::image_channel_type::fp16, class half4_cube>(
-          {48, 48}, {8, 8, 2}, seed);
+          getCubeGlobal(sizeIndex), getCubeLocal(sizeIndex), seed);
 
   if (failed) {
     std::cerr << "An error has occured!\n";

sycl/test-e2e/bindless_images/array/read_write_unsampled_array.cpp

@@ -184,4 +184,52 @@ template <int NDims> struct ImageArrayDims {
   unsigned int array_count;
 };
 
+template <int NDims> static sycl::range<NDims> getGlobalSize(size_t index) {
+
+  const std::vector<sycl::range<1>> globalSizes1D = {{32}, {16}, {20},
+                                                     {9},  {14}, {2}};
+  const std::vector<sycl::range<2>> globalSizes2D = {{32, 16}, {8, 32}, {20, 5},
+                                                     {3, 9},   {14, 7}, {2, 2}};
+  const std::vector<sycl::range<3>> globalSizes3D = {
+      {16, 8, 4}, {2, 6, 12}, {10, 15, 5}, {9, 6, 3}, {15, 7, 3}, {2, 2, 2}};
+
+  const size_t globalIndex = index % 6;
+
+  if constexpr (NDims == 1) {
+    return {globalSizes1D[globalIndex]};
+  }
+
+  if constexpr (NDims == 2) {
+    return {globalSizes2D[globalIndex]};
+  }
+
+  if constexpr (NDims == 3) {
+    return {globalSizes3D[globalIndex]};
+  }
+}
+
+template <int NDims> static sycl::range<NDims> getLocalSize(size_t index) {
+
+  const std::vector<sycl::range<1>> localSizes1D = {{2}, {16}, {5},
+                                                    {3}, {7},  {1}};
+  const std::vector<sycl::range<2>> localSizes2D = {{16, 4}, {2, 32}, {5, 5},
+                                                    {3, 3},  {7, 7},  {1, 1}};
+  const std::vector<sycl::range<3>> localSizes3D = {
+      {8, 4, 2}, {1, 3, 12}, {5, 5, 5}, {3, 3, 3}, {5, 7, 3}, {1, 1, 1}};
+
+  const size_t localIndex = index % 6;
+
+  if constexpr (NDims == 1) {
+    return localSizes1D[localIndex];
+  }
+
+  if constexpr (NDims == 2) {
+    return localSizes2D[localIndex];
+  }
+
+  if constexpr (NDims == 3) {
+    return localSizes3D[localIndex];
+  }
+}
+
 }; // namespace bindless_helpers

sycl/test-e2e/bindless_images/cubemap/cubemap_unsampled.cpp

@@ -636,11 +636,13 @@ bool runAll(sycl::range<NDims> dims, sycl::range<NDims> localSize, float offset,
 
 int main() {
 
-  unsigned int seed = 0;
+  const unsigned int seed = 0;
+  const float offset = 20.0;
+
   std::cout << "Running 1D Sampled Image Tests!\n";
-  bool result1D = runAll<1>({256}, {32}, 20, seed);
+  bool result1D = runAll<1>({128}, {32}, offset, seed);
   std::cout << "Running 2D Sampled Image Tests!\n";
-  bool result2D = runAll<2>({256, 256}, {32, 32}, 20, seed);
+  bool result2D = runAll<2>({16, 16}, {8, 8}, offset, seed);
 
   if (result1D && result2D) {
     std::cout << "All tests passed!\n";