[SYCL][CUDA] Enable printf test for CUDA (#1096)

This path allows to enable printf testing for CUDA by avoiding the vector format specifiers (%v), which are actually not supported by CUDA printf.

Author: pgorlani

SYCL/DeviceLib/built-ins/printf.cpp

@@ -1,5 +1,6 @@
-// UNSUPPORTED: cuda || hip
-// CUDA and HIP don't support printf.
+// UNSUPPORTED: hip
+// HIP doesn't support printf.
+// CUDA doesn't support vector format specifiers ("%v").
 //
 // RUN: %clangxx -fsycl -fsycl-targets=%sycl_triple %s -o %t.out
 // RUN: %HOST_RUN_PLACEHOLDER %t.out %HOST_CHECK_PLACEHOLDER
@@ -59,7 +60,7 @@ int main() {
         {
           // You can declare format string in non-global scope, but in this case
           // static keyword is required
-          static const CONSTANT char format[] = "%f\n";
+          static const CONSTANT char format[] = "%.1f\n";
           ext::oneapi::experimental::printf(format, 33.4f);
           ext::oneapi::experimental::printf(format, -33.4f);
         }
@@ -68,8 +69,8 @@ int main() {
 
         // Vectors
         sycl::vec<int, 4> v4{5, 6, 7, 8};
-#ifdef __SYCL_DEVICE_ONLY__
-        // On device side, vectors can be printed via native OpenCL types:
+#if defined(__SYCL_DEVICE_ONLY__) && defined(__SPIR__)
+        // On SPIRV devices, vectors can be printed via native OpenCL types:
         using ocl_int4 = sycl::vec<int, 4>::vector_t;
         {
           static const CONSTANT char format[] = "%v4d\n";
@@ -83,7 +84,7 @@ int main() {
                                             (int32_t)v4.x());
         }
 #else
-        // On host side you always have to print them by-element:
+        // Otherwise you always have to print them by-element:
         ext::oneapi::experimental::printf(format_vec, (int32_t)v4.x(),
                                           (int32_t)v4.y(), (int32_t)v4.z(),
                                           (int32_t)v4.w());