Implement vec::load from / store to raw pointer

celerity · Dec 16, 2024 · df3f368 · df3f368
1 parent b71bad5
commit df3f368
Showing 1 changed file with 18 additions and 0 deletions.
diff --git a/include/simsycl/sycl/vec.hh b/include/simsycl/sycl/vec.hh
@@ -359,11 +359,21 @@ class swizzled_vec {
         for(int i = 0; i < num_elements; ++i) { m_elems[indices[i]] = ptr[offset + i]; }
     }
 
+    void load(size_t offset, const value_type *ptr) const
+        requires(allow_assign)
+    {
+        for(int i = 0; i < num_elements; ++i) { m_elems[indices[i]] = ptr[offset + i]; }
+    }
+
     template<sycl::access::address_space AddressSpace, sycl::access::decorated IsDecorated>
     void store(size_t offset, sycl::multi_ptr<value_type, AddressSpace, IsDecorated> ptr) const {
         for(int i = 0; i < num_elements; ++i) { ptr[offset + i] = m_elems[indices[i]]; }
     }
 
+    void store(size_t offset, value_type *ptr) const {
+        for(int i = 0; i < num_elements; ++i) { ptr[offset + i] = m_elems[indices[i]]; }
+    }
+
     ReferenceDataT &operator[](int index) const {
         SIMSYCL_CHECK(index >= 0 && index < num_elements && "Index out of range");
         return m_elems[indices[index]];
@@ -693,11 +703,19 @@ class alignas(detail::vec_alignment_v<DataT, NumElements>) vec {
         for(int i = 0; i < NumElements; ++i) { m_elems[i] = ptr[offset + i]; }
     }
 
+    void load(size_t offset, const DataT *ptr) {
+        for(int i = 0; i < NumElements; ++i) { m_elems[i] = ptr[offset + i]; }
+    }
+
     template<access::address_space AddressSpace, access::decorated IsDecorated>
     void store(size_t offset, multi_ptr<DataT, AddressSpace, IsDecorated> ptr) const {
         for(int i = 0; i < NumElements; ++i) { ptr[offset + i] = m_elems[i]; }
     }
 
+    void store(size_t offset, DataT *ptr) const {
+        for(int i = 0; i < NumElements; ++i) { ptr[offset + i] = m_elems[i]; }
+    }
+
     DataT &operator[](int index) {
         SIMSYCL_CHECK(index >= 0 && index < NumElements && "Index out of range");
         return m_elems[index];