6 files changed, 29 insertions, 19 deletions

diff --git a/lib/core/c++/chunk.hpp b/lib/core/c++/chunk.hpp
index 607bec0..ad7de34 100644
--- a/lib/core/c++/chunk.hpp
+++ b/lib/core/c++/chunk.hpp
@@ -76,4 +76,10 @@ public:
 		return data<schema::FixedArray<schema::UInt64, sizeof...(Sides)>,Encode>{{Sides...}};
 	}
 };
+
+template<typename Sch, uint64_t Ghost, uint64_t... Sides>
+struct meta_schema<kel::lbm::sch::Chunk<Sch,Ghost,Sides...>> {
+	using MetaSchema = typename meta_schema<Sch>::MetaSchema;
+};
+
 }
diff --git a/lib/core/c++/geometry.hpp b/lib/core/c++/geometry.hpp
index 6875e90..c8a48a6 100644
--- a/lib/core/c++/geometry.hpp
+++ b/lib/core/c++/geometry.hpp
@@ -21,7 +21,9 @@ class component<Schema, Desc, cmpt::PoiseulleChannel, Encode> final {
 private:
 public:
 	template<typename CellFieldSchema>
-	void apply(saw::data<CellFieldSchema,Encode>& field, const saw::data<sch::FixedArraysch::UInt64,Desc::D>)
+	void apply(saw::data<CellFieldSchema,Encode>& field, const saw::data<sch::FixedArraysch::UInt64,Desc::D>){
+		auto& info_f = field.template get<"info">();
+	}
 };
 
 // Ghost - 0
diff --git a/lib/core/c++/lbm.hpp b/lib/core/c++/lbm.hpp
index aff38e9..00f153a 100644
--- a/lib/core/c++/lbm.hpp
+++ b/lib/core/c++/lbm.hpp
@@ -2,6 +2,7 @@
 
 #include "schema.hpp"
 #include "flatten.hpp"
+#include "chunk.hpp"
 #include "descriptor.hpp"
 #include "boundary.hpp"
 #include "converter.hpp"
diff --git a/lib/sycl/c++/SConscript b/lib/sycl/c++/SConscript
index 85a078f..2ed63ba 100644
--- a/lib/sycl/c++/SConscript
+++ b/lib/sycl/c++/SConscript
@@ -15,18 +15,13 @@ core_env = env.Clone();
 core_env.sources = sorted(glob.glob(dir_path + "/*.cpp"));
 core_env.headers = sorted(glob.glob(dir_path + "/*.hpp"));
 
-core_env.particle_headers = sorted(glob.glob(dir_path + "/particle/*.hpp"));
-core_env.particle_geometry_headers = sorted(glob.glob(dir_path + "/particle/geometry/*.hpp"));
-
 env.sources += core_env.sources;
 env.headers += core_env.headers;
 
 ## Static lib
 objects = []
 core_env.add_source_files(objects, core_env.sources, shared=False);
-env.library_static = core_env.StaticLibrary('#build/kel-lbm', [objects]);
+env.library_static = core_env.StaticLibrary('#build/kel-lbm-sycl', [objects]);
 
 env.Install('$prefix/lib/', env.library_static);
-env.Install('$prefix/include/kel/lbm/', core_env.headers);
-env.Install('$prefix/include/kel/lbm/particle/', core_env.particle_headers);
-env.Install('$prefix/include/kel/lbm/particle/geometry/', core_env.particle_geometry_headers);
+env.Install('$prefix/include/kel/lbm/sycl/', core_env.headers);
diff --git a/lib/sycl/c++/data.hpp b/lib/sycl/c++/data.hpp
index 44bc5dc..bb8b4bf 100644
--- a/lib/sycl/c++/data.hpp
+++ b/lib/sycl/c++/data.hpp
@@ -1,6 +1,7 @@
 #pragma once
 
 #include "common.hpp"
+#include <kel/lbm/lbm.hpp>
 
 namespace kel {
 namespace lbm {
@@ -19,9 +20,9 @@ struct struct_has_only_equal_dimension_array{};
 
 namespace saw {
 template<uint64_t Ghost, uint64_t... Meta, typename... Sch, string_literal... Keys, typename Encode>
-class data<schema::Struct<schema::Member<schema::Chunk<Sch,Ghost,Meta...>, Keys>...>, kel::lbm::encode::Sycl<Encode>> final {
+class data<schema::Struct<schema::Member<kel::lbm::sch::Chunk<Sch,Ghost,Meta...>, Keys>...>, kel::lbm::encode::Sycl<Encode> > final {
 public:
-	static constexpr data<schema::FixedArray meta = {{Meta...}};
+	static constexpr data<schema::FixedArray<schema::UInt64,sizeof...(Meta)>> meta = {{Meta...}};
 	using StorageT = std::tuple<data<Sch,Encode>*...>;
 private:
 
@@ -50,14 +51,14 @@ public:
 			if(not arg){
 				return;
 			}
-			sycl::free(arg,*q_);
+			acpp::sycl::free(arg,*q_);
 			arg = nullptr;
 		},members_);
 	}
 
 	template<saw::string_literal K>
 	auto* get_ptr(){
-		return std::get<parameter_key_pack_index<K, Keys...>::value(members_);
+		return std::get<parameter_key_pack_index<K, Keys...>::value>(members_);
 	}
 
 	template<saw::string_literal K>
@@ -67,6 +68,7 @@ public:
 		return *ptr;
 	}
 };
+}
 
 namespace kel {
 namespace lbm {
@@ -76,14 +78,15 @@ struct sycl_malloc_struct_helper;
 	
 template<typename... Members, typename Encode>
 struct sycl_malloc_struct_helper<sch::Struct<Members...>, Encode> final {
+	using Schema = sch::Struct<Members...>;
+
 	template<uint64_t i>
-	static saw::error_or<void> allocate_on_device_member(typename data<Sch,encode::Sycl<Encode>>::StorageT& storage, sycl::queue& q){
+	static saw::error_or<void> allocate_on_device_member(typename saw::data<typename saw::parameter_pack_type<i,Members...>::type::ValueType,encode::Sycl<Encode>>::StorageT& storage, sycl::queue& q){
 		if constexpr (i < sizeof...(Members)){
 			using M = typename saw::parameter_pack_type<i,Members...>::type;
 			auto& ptr = std::get<i>(storage);
 
-			
-			ptr = sycl::malloc_device<M::ValueType::InnerSchema>(,q);
+			ptr = sycl::malloc_device<M::ValueType::InnerSchema>(1u,q);
 
 			return allocate_on_device_member<i+1u>(storage,q);
 		}
@@ -91,8 +94,8 @@ struct sycl_malloc_struct_helper<sch::Struct<Members...>, Encode> final {
 		return saw::make_void();
 	}
 	
-	static saw::error_or<void> allocate_on_device(data<Sch,encode::Sycl<Encode>>& sycl_data, sycl::queue& q){
-		typename data<Sch,encode::Sycl<Encode>>::StorageT storage;
+	static saw::error_or<void> allocate_on_device(saw::data<Schema,encode::Sycl<Encode>>& sycl_data, sycl::queue& q){
+		typename saw::data<Schema,encode::Sycl<Encode>>::StorageT storage;
  
 		auto eov = allocate_on_device_member<0u>(storage,q);
 		sycl_data = {storage, q};
@@ -112,8 +115,8 @@ public:
 	~device() = default;
 
 	template<typename Sch, typename Encode>
-	saw::error_or<void> allocate_on_device(data<Sch,encode::Sycl<Encode>>& sycl_data){
-		auto eov = sycl_malloc_struct_helper<Sch,Encode>::allocate_on_device(sycl_data, q_);
+	saw::error_or<void> allocate_on_device(saw::data<Sch,encode::Sycl<Encode>>& sycl_data){
+		auto eov = impl::sycl_malloc_struct_helper<Sch,Encode>::allocate_on_device(sycl_data, q_);
 		if(eov.is_error()){
 			return eov;
 		}
diff --git a/lib/sycl/c++/lbm.hpp b/lib/sycl/c++/lbm.hpp
new file mode 100644
index 0000000..3e67ae6
--- /dev/null
+++ b/lib/sycl/c++/lbm.hpp
@@ -0,0 +1,3 @@
+#pragma once
+
+#include "data.hpp"