[SYCLomatic] Add the bf16,fp16, float and double type floating point comparing tolerance.

clang/runtime/dpct-rt/include/dpct/codepin/serialization/basic.hpp

@@ -11,11 +11,13 @@
 #if defined(__linux__)
 #include <cxxabi.h>
 #endif
-#include <iostream>
 #include <fstream>
+#include <iostream>
 #include <sstream>
 #include <string>
 #ifdef __NVCC__
+#include <cuda_bf16.h>
+#include <cuda_fp16.h>
 #include <cuda_runtime.h>
 #else
 #include <dpct/dpct.hpp>
@@ -191,14 +193,25 @@ template <typename T> std::string demangle_name() {
 #endif
   return ret_str;
 }
+
+#ifdef __NVCC__
+template <> std::string demangle_name<__half>() { return "fp16"; }
+template <> std::string demangle_name<__nv_bfloat16>() { return "bf16"; }
+#else
+template <> std::string demangle_name<sycl::half>() { return "fp16"; }
+template <> std::string demangle_name<sycl::ext::oneapi::bfloat16>() {
+  return "bf16";
+}
+#endif
+
 template <class T, class T2 = void> class data_ser {
+
 public:
-  static void dump(json_stringstream &ss, T value,
-                   queue_t stream) {
+  static void dump(json_stringstream &ss, T value, queue_t queue) {
     auto obj = ss.object();
     obj.key("Data");
     obj.value("CODEPIN:ERROR:1: Unable to find the corresponding serialization "
-        "function.");
+              "function.");
   }
   static void print_type_name(json_stringstream::json_obj &obj) {
     obj.key("Type");
@@ -209,16 +222,60 @@ template <class T, class T2 = void> class data_ser {
 template <class T>
 class data_ser<T, typename std::enable_if<std::is_arithmetic<T>::value>::type> {
 public:
-  static void dump(json_stringstream &ss, const T &value,
-                   queue_t stream) {
+  static void dump(json_stringstream &ss, const T &value, queue_t queue) {
     auto arr = ss.array();
     arr.member<T>(value);
   }
-  static void print_type_name(json_stringstream::json_obj &obj){
+  static void print_type_name(json_stringstream::json_obj &obj) {
+    obj.key("Type");
+    obj.value(std::string(demangle_name<T>()));
+  }
+};
+
+#ifdef __NVCC__
+template <> class data_ser<__half> {
+public:
+  static void dump(json_stringstream &ss, const __half &value, queue_t queue) {
+    float f = __half2float(value);
+    auto arr = ss.array();
+    arr.member<float>(value);
+  }
+  static void print_type_name(json_stringstream::json_obj &obj) {
+    obj.key("Type");
+    obj.value(std::string(demangle_name<__half>()));
+  }
+};
+template <> class data_ser<__nv_bfloat16> {
+public:
+  static void dump(json_stringstream &ss, const __nv_bfloat16 &value,
+                   queue_t queue) {
+    float f = __bfloat162float(value);
+    auto arr = ss.array();
+    arr.member<float>(value);
+  }
+  static void print_type_name(json_stringstream::json_obj &obj) {
+    obj.key("Type");
+    obj.value(std::string(demangle_name<__nv_bfloat16>()));
+  }
+};
+#else
+template <typename T>
+class data_ser<T,
+               typename std::enable_if<
+                   std::is_same<T, sycl::half>::value ||
+                   std::is_same<T, sycl::ext::oneapi::bfloat16>::value>::type> {
+public:
+  static void dump(json_stringstream &ss, const T &value, queue_t queue) {
+    auto arr = ss.array();
+    arr.member<T>(value);
+  }
+  static void print_type_name(json_stringstream::json_obj &obj) {
     obj.key("Type");
     obj.value(std::string(demangle_name<T>()));
   }
+
 };
+#endif
 
 #ifdef __NVCC__
 template <> class data_ser<int3> {

clang/test/dpct/codepin/report/change_floating_point_default_tolerances/CodePin_Report_ref.csv

@@ -0,0 +1,7 @@
+// CHECK: CodePin Summary
+// CHECK: Totally APIs count, 1
+// CHECK: Consistently APIs count, 1
+// CHECK: Most Time-consuming Kernel(CUDA), serial123:epilog:0, time:0.0
+// CHECK: Most Time-consuming Kernel(SYCL), serial123:epilog:0, time:0.0
+// CHECK: Peak Device Memory Used(CUDA), 400
+// CHECK: Peak Device Memory Used(SYCL), 400

clang/test/dpct/codepin/report/change_floating_point_default_tolerances/cuda.json

@@ -0,0 +1,22 @@
+[
+    {
+        "ID": "serial123:epilog:0",
+        "Free Device Memory": "100",
+        "Total Device Memory": "500",
+        "Elapse Time(ms)": "0",
+        "CheckPoint": {
+            "var_5": {
+                "Type": "float",
+                "Data": [
+                    1.00001
+                ]
+            },
+            "var_6": {
+                "Type": "float",
+                "Data": [
+                    1.3
+                ]
+            }
+        }
+    }
+]

clang/test/dpct/codepin/report/change_floating_point_default_tolerances/sycl.json

@@ -0,0 +1,22 @@
+[
+    {
+        "ID": "serial123:epilog:0",
+        "Free Device Memory": "100",
+        "Total Device Memory": "500",
+        "Elapse Time(ms)": "0",
+        "CheckPoint": {
+            "var_5": {
+                "Type": "float",
+                "Data": [
+                    1.0
+                ]
+            },
+            "var_6": {
+                "Type": "float",
+                "Data": [
+                    1.3
+                ]
+            }
+        }
+    }
+]

clang/test/dpct/codepin/report/change_floating_point_default_tolerances/test.cpp

@@ -0,0 +1,10 @@
+// RUN: cat %S/cuda.json > %T/cuda.json
+// RUN: cat %S/sycl.json > %T/sycl.json
+// RUN: cat %S/tolerances.json> %T/tolerances.json
+// RUN: cd %T
+// RUN: dpct --codepin-report --instrumented-cuda-log cuda.json --instrumented-sycl-log sycl.json --floating-point-comparison-tolerance=tolerances.json  || true
+
+// RUN: cat %S/CodePin_Report_ref.csv > %T/CodePin_Report_Expected.csv
+// RUN: cat %T/CodePin_Report.csv >> %T/CodePin_Report_Expected.csv
+
+// RUN: FileCheck --match-full-lines --input-file %T/CodePin_Report_Expected.csv %T/CodePin_Report_Expected.csv

clang/test/dpct/codepin/report/change_floating_point_default_tolerances/tolerances.json

@@ -0,0 +1,6 @@
+{
+    "bf16": 9.77e-04,
+    "fp16": 9.77e-04,
+    "float": 1.19e-04,
+    "double": 2.22e-04
+}

clang/test/dpct/codepin/report/floating_point_compare/CodePin_Report_ref.csv

@@ -0,0 +1,5 @@
+// CHECK: CodePin Summary
+// CHECK: Totally APIs count, 1
+// CHECK: Consistently APIs count, 0
+// CHECK: CUDA Meta Data ID, SYCL Meta Data ID, Type, Detail
+// CHECK: serial123:epilog:0,serial123:epilog:0,Data value,The location of failed ID Errors occurred during comparison: var_2->[0]: bf16 values 1.12 and 1.1 are not close enough [Floating Point comparison fail]; var_3->[0]: fp16 values 1.001 and 1 are not close enough [Floating Point comparison fail]; var_4->[0]: fp16 values 1.001 and 1 are not close enough [Floating Point comparison fail]; var_6->[0]: float values 1.2 and 1.3 are not close enough [Floating Point comparison fail]

clang/test/dpct/codepin/report/floating_point_compare/cuda.json

@@ -0,0 +1,46 @@
+[
+    {
+        "ID": "serial123:epilog:0",
+        "Free Device Memory": "100",
+        "Total Device Memory": "500",
+        "Elapse Time(ms)": "0",
+        "CheckPoint": {
+            "var_1": {
+                "Type": "bf16",
+                "Data": [
+                    1.0000001
+                ]
+            },
+            "var_2": {
+                "Type": "bf16",
+                "Data": [
+                    1.12
+                ]
+            },
+            "var_3": {
+                "Type": "fp16",
+                "Data": [
+                    1.001
+                ]
+            },
+            "var_4": {
+                "Type": "fp16",
+                "Data": [
+                    1.001
+                ]
+            },
+            "var_5": {
+                "Type": "float",
+                "Data": [
+                    1.00000001
+                ]
+            },
+            "var_6": {
+                "Type": "float",
+                "Data": [
+                    1.2
+                ]
+            }
+        }
+    }
+]

clang/test/dpct/codepin/report/floating_point_compare/sycl.json

@@ -0,0 +1,46 @@
+[
+    {
+        "ID": "serial123:epilog:0",
+        "Free Device Memory": "100",
+        "Total Device Memory": "500",
+        "Elapse Time(ms)": "0",
+        "CheckPoint": {
+            "var_1": {
+                "Type": "bf16",
+                "Data": [
+                    1.0
+                ]
+            },
+            "var_2": {
+                "Type": "bf16",
+                "Data": [
+                    1.1
+                ]
+            },
+            "var_3": {
+                "Type": "fp16",
+                "Data": [
+                    1
+                ]
+            },
+            "var_4": {
+                "Type": "fp16",
+                "Data": [
+                    1
+                ]
+            },
+            "var_5": {
+                "Type": "float",
+                "Data": [
+                    1.0
+                ]
+            },
+            "var_6": {
+                "Type": "float",
+                "Data": [
+                    1.3
+                ]
+            }
+        }
+    }
+]

clang/test/dpct/codepin/report/floating_point_compare/test.cpp

@@ -0,0 +1,9 @@
+// RUN: cat %S/cuda.json > %T/cuda.json
+// RUN: cat %S/sycl.json > %T/sycl.json
+// RUN: cd %T
+// RUN: dpct --codepin-report --instrumented-cuda-log cuda.json --instrumented-sycl-log sycl.json || true
+
+// RUN: cat %S/CodePin_Report_ref.csv > %T/CodePin_Report_Expected.csv
+// RUN: cat %T/CodePin_Report.csv >> %T/CodePin_Report_Expected.csv
+
+// RUN: FileCheck --match-full-lines --input-file %T/CodePin_Report_Expected.csv %T/CodePin_Report_Expected.csv