[SYCL] Fix failures on tests that use double datatype (#1246)

Author: aelovikov-intel

SYCL/ESIMD/regression/Inputs/dgetrf.hpp

@@ -298,11 +298,10 @@ ESIMD_INLINE void dgetrfnp_esimd(int64_t m, int64_t n, double *a, int64_t lda,
 #endif // defined(USE_REF)
 }
 
-void dgetrfnp_batch_strided_c(int64_t m, int64_t n, double *a, int64_t lda,
-                              int64_t stride_a, int64_t *ipiv,
+void dgetrfnp_batch_strided_c(queue &queue, int64_t m, int64_t n, double *a,
+                              int64_t lda, int64_t stride_a, int64_t *ipiv,
                               int64_t stride_ipiv, int64_t batch,
                               int64_t *info) {
-  queue queue((gpu_selector()));
   auto device = queue.get_device();
   auto context = queue.get_context();
   int status;
@@ -446,6 +445,11 @@ void dgetrfnp_batch_strided_c(int64_t m, int64_t n, double *a, int64_t lda,
                               int64_t *info);
 
 int main(int argc, char *argv[]) {
+  queue queue((gpu_selector()));
+
+  if (!queue.get_device().has(aspect::fp64))
+    return 0;
+
   int exit_status = 0;
   int64_t m = 64, n = 64, lda = 64;
   int64_t stride_a = lda * n, stride_ipiv = n;
@@ -472,8 +476,8 @@ int main(int argc, char *argv[]) {
     }
 
     /* Run the tested function */
-    dgetrfnp_batch_strided_c(m, n, a, lda, stride_a, ipiv, stride_ipiv, batch,
-                             info);
+    dgetrfnp_batch_strided_c(queue, m, n, a, lda, stride_a, ipiv, stride_ipiv,
+                             batch, info);
 
     /* Check that the computation completed successfully */
     exit_status += dgetrfnp_batch_strided_check(m, n, a_copy, a, lda, stride_a,

SYCL/ESIMD/regression/dgetrf_8x8.cpp

@@ -125,11 +125,10 @@ ESIMD_INLINE void dgetrfnp_esimd_8x8(double *a, int64_t lda, int64_t *ipiv,
   dgetrfnp_left_step<8, 8, 0>(a, lda, info);
 }
 
-void dgetrfnp_batch_strided_c(int64_t m, int64_t n, double *a, int64_t lda,
-                              int64_t stride_a, int64_t *ipiv,
+void dgetrfnp_batch_strided_c(queue &queue, int64_t m, int64_t n, double *a,
+                              int64_t lda, int64_t stride_a, int64_t *ipiv,
                               int64_t stride_ipiv, int64_t batch,
                               int64_t *info) {
-  queue queue((gpu_selector()));
   auto device = queue.get_device();
   auto context = queue.get_context();
   int status;
@@ -267,12 +266,12 @@ static int dgetrfnp_batch_strided_check(int64_t m, int64_t n, double *a_in,
   return fail;
 }
 
-void dgetrfnp_batch_strided_c(int64_t m, int64_t n, double *a, int64_t lda,
-                              int64_t stride_a, int64_t *ipiv,
-                              int64_t stride_ipiv, int64_t batch,
-                              int64_t *info);
-
 int main(int argc, char *argv[]) {
+  queue queue((gpu_selector()));
+
+  if (!queue.get_device().has(aspect::fp64))
+    return 0;
+
   int exit_status = 0;
   constexpr int64_t m = 8, n = 8, lda = 8;
   int64_t stride_a = lda * n, stride_ipiv = n;
@@ -299,8 +298,8 @@ int main(int argc, char *argv[]) {
     }
 
     /* Run the tested function */
-    dgetrfnp_batch_strided_c(m, n, a, lda, stride_a, ipiv, stride_ipiv, batch,
-                             info);
+    dgetrfnp_batch_strided_c(queue, m, n, a, lda, stride_a, ipiv, stride_ipiv,
+                             batch, info);
 
     /* Check that the computation completed successfully */
     exit_status += dgetrfnp_batch_strided_check(m, n, a_copy, a, lda, stride_a,

SYCL/ESIMD/spec_const/Inputs/spec-const-2020-common.hpp

@@ -39,6 +39,8 @@ int main(int argc, char **argv) {
   queue q(esimd_test::ESIMDSelector{}, esimd_test::createExceptionHandler());
 
   auto dev = q.get_device();
+  if (std::is_same_v<spec_const_t, double> && !dev.has(aspect::fp64))
+    return 0;
   std::cout << "Running on " << dev.get_info<info::device::name>() << "\n";
 
   std::vector<container_t> etalon = {DEF_VAL, REDEF_VAL};

SYCL/KernelParams/union_kernel_param.cpp

@@ -12,27 +12,27 @@ union TestUnion {
 public:
   int myint;
   char mychar;
-  double mydouble;
+  float myfloat;
 
-  TestUnion() { mydouble = 0.0; };
+  TestUnion() { myfloat = 0.0f; };
 };
 
 int main(int argc, char **argv) {
   TestUnion x;
-  x.mydouble = 5.0;
-  double mydouble = 0.0;
+  x.myfloat = 5.0f;
+  float myfloat = 0.0f;
 
   sycl::queue queue;
   {
-    sycl::buffer<double, 1> buf(&mydouble, 1);
+    sycl::buffer<float, 1> buf(&myfloat, 1);
     queue.submit([&](sycl::handler &cgh) {
       auto acc = buf.get_access<sycl::access::mode::read_write>(cgh);
-      cgh.single_task<class test>([=]() { acc[0] = x.mydouble; });
+      cgh.single_task<class test>([=]() { acc[0] = x.myfloat; });
     });
   }
 
-  if (mydouble != 5.0) {
-    printf("FAILED\nmydouble = %d\n", mydouble);
+  if (myfloat != 5.0f) {
+    printf("FAILED\nmyfloat = %d\n", myfloat);
     return 1;
   }
   return 0;