From a3b64625e3226b90ee9b4d8a66acf1f3fabbb7dc Mon Sep 17 00:00:00 2001
From: Duncan McBain <gitlab@mail.duncanmcba.in>
Date: Tue, 8 Mar 2022 22:44:18 +0000
Subject: [PATCH] Remove ComputeCpp-specific code from SYCL Vptr

---
 Eigen/src/Core/arch/SYCL/SyclMemoryModel.h | 20 ++++++++++----------
 1 file changed, 10 insertions(+), 10 deletions(-)

diff --git a/Eigen/src/Core/arch/SYCL/SyclMemoryModel.h b/Eigen/src/Core/arch/SYCL/SyclMemoryModel.h
index 2b96587f4..54eedfaa0 100644
--- a/Eigen/src/Core/arch/SYCL/SyclMemoryModel.h
+++ b/Eigen/src/Core/arch/SYCL/SyclMemoryModel.h
@@ -141,7 +141,7 @@ class PointerMapper {
 
   /* basic type for all buffers
    */
-  using buffer_t = cl::sycl::buffer_mem;
+  using buffer_t = cl::sycl::buffer<buffer_data_type_t>;
 
   /**
    * Node that stores information about a device allocation.
@@ -237,17 +237,14 @@ class PointerMapper {
   template <typename buffer_data_type = buffer_data_type_t>
   cl::sycl::buffer<buffer_data_type, 1> get_buffer(
       const virtual_pointer_t ptr) {
-    using sycl_buffer_t = cl::sycl::buffer<buffer_data_type, 1>;
 
-    // get_node() returns a `buffer_mem`, so we need to cast it to a `buffer<>`.
-    // We can do this without the `buffer_mem` being a pointer, as we
-    // only declare member variables in the base class (`buffer_mem`) and not in
-    // the child class (`buffer<>).
     auto node = get_node(ptr);
+    auto& map_node = node->second;
     eigen_assert(node->first == ptr || node->first < ptr);
-    eigen_assert(ptr < static_cast<virtual_pointer_t>(node->second.m_size +
+    eigen_assert(ptr < static_cast<virtual_pointer_t>(map_node.m_size +
                                                       node->first));
-    return *(static_cast<sycl_buffer_t *>(&node->second.m_buffer));
+    return map_node.m_buffer.reinterpret<buffer_data_type>(
+        cl::sycl::range<1>{map_node.m_size / sizeof(buffer_data_type)});
   }
 
   /**
@@ -429,8 +426,11 @@ class PointerMapper {
   template <class BufferT>
   virtual_pointer_t add_pointer_impl(BufferT b) {
     virtual_pointer_t retVal = nullptr;
-    size_t bufSize = b.get_count();
-    pMapNode_t p{b, bufSize, false};
+    size_t bufSize = b.get_count() * sizeof(buffer_data_type_t);
+    auto byte_buffer =
+        b.template reinterpret<buffer_data_type_t>(cl::sycl::range<1>{bufSize});
+    pMapNode_t p{byte_buffer, bufSize, false};
+
     // If this is the first pointer:
     if (m_pointerMap.empty()) {
       virtual_pointer_t initialVal{m_baseAddress};