1 files changed, 42 insertions, 0 deletions
diff --git a/modules/remote-sycl/benchmarks/kernel_mixed_precision.cpp b/modules/remote-sycl/benchmarks/kernel_mixed_precision.cpp
new file mode 100644
index 0000000..0ac9756
--- /dev/null
+++ b/modules/remote-sycl/benchmarks/kernel_mixed_precision.cpp
@@ -0,0 +1,42 @@
+#include "mixed_precision.hpp"
+
+saw::interface<sch::MixedPrecisionBenchmarkInterface, saw::encode::Native, saw::rmt::Sycl, cl::sycl::queue*> listen_mixed_precision(cl::sycl::event& mixed_ev, cl::sycl::event& float64_ev, cl::sycl::event& float32_ev){
+	return {
+		/**
+		 * Mixed
+		 */
+		[&](saw::data<sch::MixedArray, saw::encode::Native, saw::rmt::Sycl>& in, cl::sycl::queue* cmd) -> saw::error_or<void> {
+	
+			mixed_ev = cmd->submit([&](cl::sycl::handler& h){
+				auto acc_buff = in.template access<cl::sycl::access::mode::read_write>(h);
+
+				h.parallel_for(cl::sycl::range<1>(in.size()), [=] (cl::sycl::id<1> it){
+					acc_buff[0u].at(it[0u]) = acc_buff[0u].at(it[0u]) * saw::data<sch::Float64>{2.0};
+				});
+			});
+			return saw::void_t{};
+		},
+		[&](saw::data<sch::Float64Array, saw::encode::Native, saw::rmt::Sycl>& in, cl::sycl::queue* cmd) -> saw::error_or<void> {
+	
+			float64_ev = cmd->submit([&](cl::sycl::handler& h){
+				auto acc_buff = in.template access<cl::sycl::access::mode::read_write>(h);
+
+				h.parallel_for(cl::sycl::range<1>(in.size()), [=] (cl::sycl::id<1> it){
+					acc_buff[0u].at(it[0u]) = acc_buff[0u].at(it[0u]) * saw::data<sch::Float64>{2.0};
+				});
+			});
+			return saw::void_t{};
+		},
+		[&](saw::data<sch::Float32Array, saw::encode::Native, saw::rmt::Sycl>& in, cl::sycl::queue* cmd) -> saw::error_or<void> {
+	
+			float32_ev = cmd->submit([&](cl::sycl::handler& h){
+				auto acc_buff = in.template access<cl::sycl::access::mode::read_write>(h);
+
+				h.parallel_for(cl::sycl::range<1>(in.size()), [=] (cl::sycl::id<1> it){
+					acc_buff[0u].at(it[0u]) = acc_buff[0u].at(it[0u]) * saw::data<sch::Float32>{2.0f};
+				});
+			});
+			return saw::void_t{};
+		}
+	};
+}