[SYCL] Add on device tests moved from intel/llvm

SYCL/Basic/aspects.cpp

@@ -0,0 +1,106 @@
+// RUN: %clangxx -fsycl %s -o %t.out
+// RUN: env SYCL_DEVICE_FILTER=%sycl_be %t.out
+//
+// Hip is missing some of the parameters tested here so it fails with ROCm for
+// NVIDIA
+// XFAIL: rocm_nvidia
+
+//==--------------- aspects.cpp - SYCL device test ------------------------==//
+//
+// Returns the various aspects of a device  and platform.
+//
+// Part of the LLVM Project, under the Apache License v2.0 with LLVM Exceptions.
+// See https://llvm.org/LICENSE.txt for license information.
+// SPDX-License-Identifier: Apache-2.0 WITH LLVM-exception
+//
+//===----------------------------------------------------------------------===//
+
+#include <CL/sycl.hpp>
+#include <iostream>
+
+using namespace cl::sycl;
+
+// platform::has() calls device::has() for each device on the platform.
+
+int main() {
+  bool failed = false;
+  int pltIdx = 0;
+  for (const auto &plt : platform::get_platforms()) {
+    pltIdx++;
+    if (plt.has(aspect::host)) {
+      std::cout << "Platform #" << pltIdx
+                << " type: Host supports:" << std::endl;
+    } else if (plt.has(aspect::cpu)) {
+      std::cout << "Platform #" << pltIdx
+                << " type: CPU supports:" << std::endl;
+    } else if (plt.has(aspect::gpu)) {
+      std::cout << "Platform #" << pltIdx
+                << " type: GPU supports:" << std::endl;
+    } else if (plt.has(aspect::accelerator)) {
+      std::cout << "Platform #" << pltIdx
+                << " type: Accelerator supports:" << std::endl;
+    } else if (plt.has(aspect::custom)) {
+      std::cout << "Platform #" << pltIdx
+                << " type: Custom supports:" << std::endl;
+    } else {
+      failed = true;
+      std::cout << "Failed: platform #" << pltIdx << " type: unknown"
+                << std::endl;
+      return 1;
+    }
+
+    if (plt.has(aspect::fp16)) {
+      std::cout << "  fp16" << std::endl;
+    }
+    if (plt.has(aspect::fp64)) {
+      std::cout << "  fp64" << std::endl;
+    }
+    if (plt.has(aspect::int64_base_atomics)) {
+      std::cout << "  base atomic operations" << std::endl;
+    }
+    if (plt.has(aspect::int64_extended_atomics)) {
+      std::cout << "  extended atomic operations" << std::endl;
+    }
+    if (plt.has(aspect::atomic64)) {
+      std::cout << "  atomic64" << std::endl;
+    }
+    if (plt.has(aspect::image)) {
+      std::cout << "  images" << std::endl;
+    }
+    if (plt.has(aspect::online_compiler)) {
+      std::cout << "  online compiler" << std::endl;
+    }
+    if (plt.has(aspect::online_linker)) {
+      std::cout << "  online linker" << std::endl;
+    }
+    if (plt.has(aspect::queue_profiling)) {
+      std::cout << "  queue profiling" << std::endl;
+    }
+    if (plt.has(aspect::usm_device_allocations)) {
+      std::cout << "  USM allocations" << std::endl;
+    }
+    if (plt.has(aspect::usm_host_allocations)) {
+      std::cout << "  USM host allocations" << std::endl;
+    }
+    if (plt.has(aspect::usm_atomic_host_allocations)) {
+      std::cout << "  USM atomic host allocations" << std::endl;
+    }
+    if (plt.has(aspect::usm_shared_allocations)) {
+      std::cout << "  USM shared allocations" << std::endl;
+    }
+    if (plt.has(aspect::usm_atomic_shared_allocations)) {
+      std::cout << "  USM atomic shared allocations" << std::endl;
+    }
+    if (plt.has(aspect::usm_restricted_shared_allocations)) {
+      std::cout << "  USM restricted shared allocations" << std::endl;
+    }
+    if (plt.has(aspect::usm_system_allocator)) {
+      std::cout << "  USM system allocator" << std::endl;
+    }
+    if (plt.has(aspect::usm_system_allocations)) {
+      std::cout << "  USM system allocations" << std::endl;
+    }
+  }
+  std::cout << "Passed." << std::endl;
+  return 0;
+}

SYCL/Basic/diagnostics/device-check.cpp

@@ -0,0 +1,51 @@
+// RUN: %clangxx -fsycl -fsycl-targets=%sycl_triple %s -o %t.out
+// RUN: env SYCL_DEVICE_FILTER=%sycl_be SYCL_DEVICE_TYPE=cpu %t.out
+// RUN: env SYCL_DEVICE_FILTER=%sycl_be SYCL_DEVICE_TYPE=gpu %t.out
+// RUN: env SYCL_DEVICE_FILTER=%sycl_be SYCL_DEVICE_TYPE=acc %t.out
+// RUN: env SYCL_DEVICE_FILTER=%sycl_be SYCL_DEVICE_TYPE=host %t.out
+// RUN: env SYCL_DEVICE_FILTER=%sycl_be SYCL_DEVICE_TYPE=CPU %t.out
+// RUN: env SYCL_DEVICE_FILTER=%sycl_be SYCL_DEVICE_TYPE=GPU %t.out
+// RUN: env SYCL_DEVICE_FILTER=%sycl_be SYCL_DEVICE_TYPE=ACC %t.out
+// RUN: env SYCL_DEVICE_FILTER=%sycl_be SYCL_DEVICE_TYPE=HOST %t.out
+// RUN: env SYCL_DEVICE_FILTER=%sycl_be SYCL_DEVICE_TYPE=Cpu %t.out
+// RUN: env SYCL_DEVICE_FILTER=%sycl_be SYCL_DEVICE_TYPE=Gpu %t.out
+// RUN: env SYCL_DEVICE_FILTER=%sycl_be SYCL_DEVICE_TYPE=Acc %t.out
+// RUN: env SYCL_DEVICE_FILTER=%sycl_be SYCL_DEVICE_TYPE=Host %t.out
+// RUN: env SYCL_DEVICE_FILTER=%sycl_be SYCL_DEVICE_TYPE=XPU %t.out
+
+//==------------------- device-check.cpp --------------------------==//
+// This is a diagnostic test which ensures that
+// device types are case-insensitive.
+// It also checks for SYCL_DEVICE being set incorrectly.
+//==---------------------------------------------------------------==//
+
+#include <CL/sycl.hpp>
+#include <iostream>
+
+using namespace cl::sycl;
+
+int main() {
+  try {
+    queue q = queue();
+    auto device = q.get_device();
+    auto deviceName = device.get_info<cl::sycl::info::device::name>();
+    std::cout << " Device Name: " << deviceName << std::endl;
+  }
+
+  catch (runtime_error &E) {
+    if (std::string(E.what()).find("SYCL_DEVICE_TYPE is not recognized.  Must "
+                                   "be GPU, CPU, ACC or HOST.") ==
+            std::string::npos &&
+        std::string(E.what()).find("No device of requested type available.") ==
+            std::string::npos) {
+      std::cout << "Test failed: received error is incorrect." << std::endl;
+      return 1;
+    } else {
+      std::cout << "Test passed: caught the expected error." << std::endl;
+      return 0;
+    }
+  }
+
+  std::cout << "Test passed: results are correct." << std::endl;
+  return 0;
+}

SYCL/Basic/image/srgba-read.cpp

@@ -0,0 +1,136 @@
+// RUN: %clangxx -fsycl  -fsycl-targets=%sycl_triple %s -o %t.out
+// RUN: %CPU_RUN_PLACEHOLDER %t.out %CPU_CHECK_PLACEHOLDER
+// RUN: %GPU_RUN_PLACEHOLDER %t.out %GPU_CHECK_PLACEHOLDER
+
+// XFAIL: level_zero
+// UNSUPPORTED: cuda
+// UNSUPPORTED: rocm_nvidia
+// UNSUPPORTED: rocm_amd
+
+#include <CL/sycl.hpp>
+
+using namespace cl::sycl;
+
+using accessorPixelT = sycl::float4;
+using dataPixelT = uint32_t;
+
+// will output a pixel as {r,g,b,a}.  provide override if a different pixelT is
+// defined.
+void outputPixel(sycl::float4 somePixel) {
+  std::cout << "{" << somePixel[0] << "," << somePixel[1] << "," << somePixel[2]
+            << "," << somePixel[3] << "} ";
+}
+
+constexpr long width = 4;
+constexpr long height = 3;
+
+void test_rd(image_channel_order ChanOrder, image_channel_type ChanType) {
+
+  int numTests = 4; // drives the size of the testResults buffer, and the number
+                    // of report iterations. Kludge.
+
+  // this should yield a read of approximate 0.5 for each channel
+  // when read directly with a normal non-linearized image (e.g.
+  // image_channel_order::rgba). For sRGB
+  // (image_channel_order::ext_oneapi_srgba), this is the value with maximum
+  // conversion. So we should read values of approximately 0.2
+  dataPixelT basicPixel{127 << 24 | 127 << 16 | 127 << 8 | 127};
+
+  queue Q;
+  const sycl::range<2> ImgRange_2D(width, height);
+
+  // IMPORTANT: const data is *required* for sRGBA images.
+  // OpenCL support is limited for 2D/3D images that are read only.
+  const std::vector<dataPixelT> ImgData(ImgRange_2D.size(), basicPixel);
+  try { // closure
+
+    image<2> image_2D(ImgData.data(), ChanOrder, ChanType, ImgRange_2D);
+    // use a buffer to report back test results.
+    buffer<accessorPixelT, 1> testResults((range<1>(numTests)));
+
+    Q.submit([&](handler &cgh) {
+      auto image_acc =
+          image_2D.get_access<accessorPixelT, access::mode::read>(cgh);
+      auto test_acc = testResults.get_access<access::mode::write>(cgh);
+
+      cgh.single_task<class im2D_rw>([=]() {
+        int i = 0; // the index for writing into the testResult buffer.
+
+        // verify our four pixels were set up correctly.
+        // 0-3 read four pixels. no sampler
+        test_acc[i++] = image_acc.read(sycl::int2{0, 0});
+        test_acc[i++] = image_acc.read(sycl::int2{1, 0});
+        test_acc[i++] = image_acc.read(sycl::int2{0, 1});
+        test_acc[i++] = image_acc.read(sycl::int2{2, 2});
+      });
+    });
+    Q.wait_and_throw();
+
+    // REPORT RESULTS
+    auto test_acc = testResults.get_access<access::mode::read>();
+    for (int i = 0, idx = 0; i < numTests; i++, idx++) {
+      if (i == 0) {
+        idx = 0;
+        std::cout << "read four pixels, no sampler" << std::endl;
+      }
+
+      accessorPixelT testPixel = test_acc[i];
+      std::cout << i << /* " -- " << idx << */ ": ";
+      outputPixel(testPixel);
+      std::cout << std::endl;
+    }
+  } catch (sycl::exception e) {
+    std::cout << "exception caught: " << e.what() << std::endl;
+  } // ~image / ~buffer
+}
+
+int main() {
+
+#ifdef SYCL_EXT_ONEAPI_SRGB
+  std::cout << "SYCL_EXT_ONEAPI_SRGB defined" << std::endl;
+#endif
+
+  queue Q;
+  device D = Q.get_device();
+
+  // test aspect
+  if (D.has(aspect::ext_oneapi_srgb))
+    std::cout << "aspect::ext_oneapi_srgb detected" << std::endl;
+
+  if (D.has(aspect::image)) {
+    // RGBA -- (normal, non-linearized)
+    std::cout << "rgba -------" << std::endl;
+    test_rd(image_channel_order::rgba, image_channel_type::unorm_int8);
+
+    // sRGBA -- (linearized reads)
+    std::cout << "srgba -------" << std::endl;
+    test_rd(image_channel_order::ext_oneapi_srgba,
+            image_channel_type::unorm_int8);
+  } else {
+    std::cout << "device does not support image operations" << std::endl;
+  }
+
+  return 0;
+}
+
+// clang-format off
+// CHECK: SYCL_EXT_ONEAPI_SRGB defined
+// CHECK: aspect::ext_oneapi_srgb detected
+
+// CHECK: rgba -------
+// CHECK-NEXT: read four pixels, no sampler
+//   these next four reads should all be close to 0.5
+// CHECK-NEXT: 0: {0.498039,0.498039,0.498039,0.498039} 
+// CHECK-NEXT: 1: {0.498039,0.498039,0.498039,0.498039} 
+// CHECK-NEXT: 2: {0.498039,0.498039,0.498039,0.498039} 
+// CHECK-NEXT: 3: {0.498039,0.498039,0.498039,0.498039} 
+// CHECK: srgba -------
+// CHECK-NEXT: read four pixels, no sampler
+//   these next four reads should have R, G, B values close to 0.2 
+//   presently the values differ slightly between OpenCL GPU and CPU
+// (e.g. GPU: 0.21231, CPU: 0.211795 )
+// CHECK-NEXT: 0: {0.21 
+// CHECK-NEXT: 1: {0.21 
+// CHECK-NEXT: 2: {0.21 
+// CHECK-NEXT: 3: {0.21
+// clang-format on