diff options
Diffstat (limited to 'modules/remote-hip')
| -rw-r--r-- | modules/remote-hip/c++/device.tmpl.hpp | 27 |
1 files changed, 27 insertions, 0 deletions
diff --git a/modules/remote-hip/c++/device.tmpl.hpp b/modules/remote-hip/c++/device.tmpl.hpp index 4777660..6edf431 100644 --- a/modules/remote-hip/c++/device.tmpl.hpp +++ b/modules/remote-hip/c++/device.tmpl.hpp @@ -11,6 +11,7 @@ struct hip_copy_to_device { template<typename T, uint64_t N, typename Encoding> struct hip_copy_to_device<schema::Primitive<T,N>, Encoding> { using Schema = schema::Primitive<T,N>; + static error_or<void> apply(data<Schema, Encoding>& from, data<Schema,Encoding>** to){ hipError_t malloc_err = hipMalloc(to, sizeof(data<Schema,Encoding>)); // HIP_CHECK(malloc_err); @@ -21,5 +22,31 @@ struct hip_copy_to_device<schema::Primitive<T,N>, Encoding> { return make_void(); } }; + +template<typename T, uint64_t Dim, typename Encoding> +struct hip_copy_to_device<schema::Array<T,Dim>, Encoding> { + static_assert(Dim == 1u, "Only 1D arrays are supported for now."); + static_assert(is_primitive<T>::value, "Arrays can only handle primitives for now."); + + using Schema = schema::Array<T,Dim>; + + static error_or<void> apply(data<Schema, Encoding>& from, data<Schema,Encoding>** to){ + typename native_data_type<T>::type* dat{}; + hipError_t data_malloc_err = hipMalloc(&dat,sizeof(typename native_data_type<T>::type) * from.size()); + hipError_t data_copy_err = hipMemcpy(&dat, &(from.get_raw_data()),sizeof(typename native_data_type<T>::type) * from.size(), hipMemcpyHostToDevice); + + if(from.size() == 0u){ + // Everything is fine. We just don't want to allocate data which doesn't exist. + return make_void(); + } + + // auto from_dat = &from.at(0); + + hipError_t malloc_err = hipMalloc(to, sizeof(data<Schema,Encoding>)); + hipError_t copy_err = hipMemcpy(*to, &from, sizeof(data<Schema,Encoding>), hipMemcpyHostToDevice); + + return make_void(); + } +}; } } |