Progress in fixing remote-sycl and ammended minor parts in echo_client

7 files changed, 206 insertions, 23 deletions

diff --git a/modules/remote-sycl/.nix/derivation.nix b/modules/remote-sycl/.nix/derivation.nix
index 4c994a9..440bd86 100644
--- a/modules/remote-sycl/.nix/derivation.nix
+++ b/modules/remote-sycl/.nix/derivation.nix
@@ -4,13 +4,16 @@
 , clang-tools
 , version
 , forstio
+, openmp
 , opensycl
+
+, build_examples ? "false"
 }:
 
 let
 
 in stdenv.mkDerivation {
-  pname = "forstio-device-opencl";
+  pname = "forstio-remote-sycl";
   inherit version;
   src = ./..;
 
@@ -26,9 +29,24 @@ in stdenv.mkDerivation {
 		forstio.codec
 		forstio.async
 		forstio.io
-		forstio.io_codec
+    forstio.io_codec
+    openmp
 		opensycl
   ];
+	
+	buildPhase = ''
+    scons build_examples=${build_examples}
+  '';
+
+  installPhase = ''
+    scons prefix=$out build_examples=${build_examples} install
+  '';
+
+  doCheck = true;
+  checkPhase = ''
+    scons test
+    ./bin/tests
+  '';
 
   outputs = ["out" "dev"];
 }
diff --git a/modules/remote-sycl/SConstruct b/modules/remote-sycl/SConstruct
index 876638a..4bdab03 100644
--- a/modules/remote-sycl/SConstruct
+++ b/modules/remote-sycl/SConstruct
@@ -43,6 +43,13 @@ env_vars.Add('prefix',
 	validator=isAbsolutePath
 )
 
+env_vars.Add(
+    BoolVariable('build_examples',
+        help='Build examples',
+        default=False
+    )
+);
+
 env=Environment(ENV=os.environ, variables=env_vars, CPPPATH=[],
     CPPDEFINES=['SAW_UNIX'],
     CXXFLAGS=['-std=c++20','-g','-Wall','-Wextra'],
@@ -52,6 +59,7 @@ env=Environment(ENV=os.environ, variables=env_vars, CPPPATH=[],
         ,'forstio-io'
         ,'forstio-io_codec'
         ,'hipSYCL-rt'
+        ,'omp'
         #,'OpenCL'
     ]
 );
@@ -67,6 +75,7 @@ env.targets = [];
 Export('env')
 SConscript('c++/SConscript')
 SConscript('tests/SConscript')
+SConscript('examples/SConscript')
 
 env.Alias('cdb', env.cdb);
 env.Alias('all', [env.targets]);
diff --git a/modules/remote-sycl/c++/SConscript b/modules/remote-sycl/c++/SConscript
index e902c22..29b7ca4 100644
--- a/modules/remote-sycl/c++/SConscript
+++ b/modules/remote-sycl/c++/SConscript
@@ -10,29 +10,29 @@ Import('env')
 dir_path = Dir('.').abspath
 
 # Environment for base library
-dev_opencl_env = env.Clone();
+dev_sycl_env = env.Clone();
 
-dev_opencl_env.sources = sorted(glob.glob(dir_path + "/*.cpp"))
-dev_opencl_env.headers = sorted(glob.glob(dir_path + "/*.hpp"))
+dev_sycl_env.sources = sorted(glob.glob(dir_path + "/*.cpp"))
+dev_sycl_env.headers = sorted(glob.glob(dir_path + "/*.hpp"))
 
-env.sources += dev_opencl_env.sources;
-env.headers += dev_opencl_env.headers;
+env.sources += dev_sycl_env.sources;
+env.headers += dev_sycl_env.headers;
 
 ## Shared lib
 objects_shared = []
-dev_opencl_env.add_source_files(objects_shared, dev_opencl_env.sources, shared=True);
-env.library_shared = dev_opencl_env.SharedLibrary('#build/forstio-remote-opencl', [objects_shared]);
+dev_sycl_env.add_source_files(objects_shared, dev_sycl_env.sources, shared=True);
+env.library_shared = dev_sycl_env.SharedLibrary('#build/forstio-remote-sycl', [objects_shared]);
 
 ## Static lib
 objects_static = []
-dev_opencl_env.add_source_files(objects_static, dev_opencl_env.sources, shared=False);
-env.library_static = dev_opencl_env.StaticLibrary('#build/forstio-remote-opencl', [objects_static]);
+dev_sycl_env.add_source_files(objects_static, dev_sycl_env.sources, shared=False);
+env.library_static = dev_sycl_env.StaticLibrary('#build/forstio-remote-sycl', [objects_static]);
 
 # Set Alias
-env.Alias('library_remote_opencl', [env.library_shared, env.library_static]);
+env.Alias('library_remote_sycl', [env.library_shared, env.library_static]);
 
-env.targets += ['library_remote_opencl'];
+env.targets += ['library_remote_sycl'];
 
 # Install
 env.Install('$prefix/lib/', [env.library_shared, env.library_static]);
-env.Install('$prefix/include/forstio/remote/opencl/', [dev_opencl_env.headers]);
+env.Install('$prefix/include/forstio/remote/sycl/', [dev_sycl_env.headers]);
diff --git a/modules/remote-sycl/c++/remote.hpp b/modules/remote-sycl/c++/remote.hpp
index 86799b6..d4b114a 100644
--- a/modules/remote-sycl/c++/remote.hpp
+++ b/modules/remote-sycl/c++/remote.hpp
@@ -36,6 +36,16 @@ public:
 	 */
 	conveyor<data<T,Encoding>> on_receive(); /// Stopped here
 };
+
+template<typename T, uint64_t D>
+class data<schema::Array<T,D>, encode::Native<rmt::Sycl>> {
+private:
+	cl::sycl::buffer<typename native_data_type<T>::type, D> device_data_;
+
+	static_assert(D==1u, "For now we only support 1D Arrays");
+public:
+};
+
 namespace impl {
 
 template<typename Iface, typename Encoding>
@@ -53,8 +63,10 @@ struct rpc_id_map_helper<schema::Interface<Members...>, Encoding> {
  */
 template<typename Iface, typename Encoding>
 class rpc_server<Iface, Encoding, rmt::Sycl> {
+public:
+	using InterfaceCtxT = cl::sycl::queue*;
+	using InterfaceT = interface<Iface, Encoding, InterfaceCtxT>;
 private:
-	using IfaceCtx = cl::sycl::queue*;
 	/**
 	 * Command queue for the sycl backend
 	 */
@@ -63,14 +75,14 @@ private:
 	/**
 	 * The interface including the relevant context class.
 	 */
-	interface<Iface, Encoding, IfaceCtx> cl_interface_;
+	interface<Iface, Encoding, InterfaceCtxT> cl_interface_;
 
 	/**
 	 *
 	 */
 	impl::rpc_id_map_helper<Iface, Encoding> storage_;
 public:
-	rpc_server(interface<Iface, Encoding, IfaceCtx> cl_iface):
+	rpc_server(interface<Iface, Encoding, InterfaceCtxT> cl_iface):
 		cmd_queue_{},
 		cl_interface_{std::move(cl_iface)},
 		storage_{}
@@ -87,9 +99,9 @@ public:
 	template<string_literal Name>
 	error_or<
 		id<
-			typename schema_member_type<Name, Iface>::type::ValueType::ResponseT
+			typename schema_member_type<Name, Iface>::type::ResponseT
 		>
-	> call(data_or_id<typename schema_member_type<Name, Iface>::type::ValueType::RequestT, Encoding> input){
+	> call(data_or_id<typename schema_member_type<Name, Iface>::type::RequestT, Encoding> input){
 
 		auto eod = cl_interface_.template call<Name>(std::move(input), &cmd_queue_);
 
@@ -97,7 +109,9 @@ public:
 			return std::move(eod.get_error());
 		}
 
-		return id<typename schema_member_type<Name, Iface>::type::ValueType::ResponseT>{};
+		// using ResponseTMap = id_map<data<>>
+
+		return id<typename schema_member_type<Name, Iface>::type::ResponseT>{};
 	}
 };
 
@@ -139,8 +153,10 @@ public:
 	 * Spin up a rpc server
 	 */
 	template<typename Iface, typename Encoding>
-	conveyor<rpc_server<Iface, Encoding, rmt::Sycl>> listen(const remote_address<rmt::Sycl>&){
-		return {};
+	rpc_server<Iface, Encoding, rmt::Sycl> listen(const remote_address<rmt::Sycl>&, typename rpc_server<Iface, Encoding, rmt::Sycl>::InterfaceT iface){
+		using RpcServerT = rpc_server<Iface, Encoding, rmt::Sycl>;
+		using InterfaceT = typename RpcServerT::InterfaceT;
+		return {std::move(iface)};
 	}
 };
 
diff --git a/modules/remote-sycl/examples/SConscript b/modules/remote-sycl/examples/SConscript
new file mode 100644
index 0000000..7fa9429
--- /dev/null
+++ b/modules/remote-sycl/examples/SConscript
@@ -0,0 +1,31 @@
+#!/bin/false
+
+import os
+import os.path
+import glob
+
+
+Import('env')
+
+dir_path = Dir('.').abspath
+
+# Environment for base library
+examples_env = env.Clone();
+
+examples_env.sources = sorted(glob.glob(dir_path + "/*.cpp"))
+examples_env.headers = sorted(glob.glob(dir_path + "/*.hpp"))
+
+env.sources += examples_env.sources;
+env.headers += examples_env.headers;
+
+objects_static = []
+examples_env.sycl_basic = examples_env.Program('#bin/sycl_basic', ['sycl_basic.cpp', env.library_static]);
+
+# Set Alias
+env.examples = [examples_env.sycl_basic];
+env.Alias('examples', env.examples);
+
+if env["build_examples"]:
+	env.targets += ['examples'];
+	env.Install('$prefix/bin/', env.examples);
+#endif
diff --git a/modules/remote-sycl/examples/sycl_basic.cpp b/modules/remote-sycl/examples/sycl_basic.cpp
new file mode 100644
index 0000000..a180d23
--- /dev/null
+++ b/modules/remote-sycl/examples/sycl_basic.cpp
@@ -0,0 +1,47 @@
+#include "../c++/remote.hpp"
+
+namespace schema {
+using namespace saw::schema;
+
+using BasicInterface = Interface<
+	Member<Function<UInt64, UInt64>, "increment">
+>;
+}
+
+int main(){
+	saw::remote<saw::rmt::Sycl> remote_ctx;
+
+	saw::own<saw::remote_address<saw::rmt::Sycl>> rmt_addr{};
+	
+	saw::event_loop loop;
+	saw::wait_scope wait{loop};
+
+	remote_ctx.resolve_address().then([&](auto addr){
+		rmt_addr = std::move(addr);
+	}).detach();
+
+	wait.poll();
+
+	if(!rmt_addr){
+		return -1;
+	}
+
+	saw::interface<schema::BasicInterface, saw::encode::Native<saw::storage::Default>, cl::sycl::queue*> iface{
+		[](saw::data<saw::schema::UInt64> in, cl::sycl::queue* q) -> saw::data<saw::schema::UInt64> {
+			return {in.get() + 1u};
+		}
+	};
+	auto rpc_server = remote_ctx.template listen<schema::BasicInterface, saw::encode::Native<saw::storage::Default>>(*rmt_addr, std::move(iface));
+
+	{
+		auto eov = rpc_server.template call<"increment">({1u});
+		if(eov.is_error()){
+			return -2;
+		}
+		auto& val = eov.get_value();
+		std::cout<<"Value: "<<val<<std::endl;
+	}
+
+
+	return 0;
+}
diff --git a/modules/remote-sycl/tests/calculator.cpp b/modules/remote-sycl/tests/calculator.cpp
index 9ee1583..730838d 100644
--- a/modules/remote-sycl/tests/calculator.cpp
+++ b/modules/remote-sycl/tests/calculator.cpp
@@ -2,6 +2,68 @@
 
 #include "../c++/remote.hpp"
 
-SAW_TEST("OpenCl Calculator"){
+namespace {
+namespace schema {
+using namespace saw::schema;
+using Calculator = Interface<
+	Member<
+		Function<Tuple<Int64, Int64>, Int64>, "add"
+	>
+,	Member<
+		Function<Tuple<Int64, Int64>, Int64>, "multiply"
+	>
+>;
+}
+
+SAW_TEST("Sycl Interface Calculator"){
+	using namespace saw;
+
+	cl::sycl::queue cmd_queue;
+
+	interface<schema::Calculator, encode::Native<storage::Default>, cl::sycl::queue*> cl_iface {
+[](data<schema::Tuple<schema::Int64, schema::Int64>> in, cl::sycl::queue* cmd) -> data<schema::Int64> {
+			std::array<int64_t,2> h_xy{in.get<0>().get(), in.get<1>().get()};
+			int64_t res{};
+			cl::sycl::buffer<int64_t,1> d_xy { h_xy.data(), h_xy.size() };
+			cl::sycl::buffer<int64_t,1> d_z { &res, 1u };
+			cmd->submit([&](cl::sycl::handler& h){
+				auto a_xy = d_xy.get_access<cl::sycl::access::mode::read>(h);
+				auto a_z = d_z.get_access<cl::sycl::access::mode::write>(h);
+
+				h.parallel_for(cl::sycl::range<1>(1u), [=] (cl::sycl::id<1> it){
+					a_z[0] =  a_xy[0] + a_xy[1];
+				});
+			});
+			cmd->wait();
+			return data<schema::Int64>{res};
+		},
+		[](data<schema::Tuple<schema::Int64, schema::Int64>> in, cl::sycl::queue* cmd) -> data<schema::Int64> {
+			return data<schema::Int64>{in.get<0>().get() * in.get<1>().get()};
+		}
+	};
+
+	int64_t x = 1;
+	int64_t y = -2;
 
+	int64_t sum = x + y;
+	int64_t mult = x * y;
+
+
+	data<schema::Tuple<schema::Int64, schema::Int64>> input;
+	input.template get<0>().set(x);
+	input.template get<1>().set(y);
+
+	{
+		auto eov = cl_iface.template call<"add">(input, &cmd_queue);
+		SAW_EXPECT(eov.is_value(), "Returned error on add");
+
+		SAW_EXPECT(eov.get_value().get() == sum, "Addition was incorrect");
+	}
+	{
+		auto eov = cl_iface.template call<"multiply">(input, &cmd_queue);
+		SAW_EXPECT(eov.is_value(), "Returned error on add");
+
+		SAW_EXPECT(eov.get_value().get() == mult, "Addition was incorrect");
+	}
+}
 }