diff --git a/unsupported/Eigen/CXX11/src/Tensor/TensorDeviceSycl.h b/unsupported/Eigen/CXX11/src/Tensor/TensorDeviceSycl.h
index f63d79945..844c09396 100644
--- a/unsupported/Eigen/CXX11/src/Tensor/TensorDeviceSycl.h
+++ b/unsupported/Eigen/CXX11/src/Tensor/TensorDeviceSycl.h
@@ -962,6 +962,15 @@ struct SyclDevice : public SyclDeviceBase {
     return queue_stream()->get(data);
   }
 
+  /// attach existing buffer
+  EIGEN_STRONG_INLINE void *attach_buffer(
+      cl::sycl::buffer<buffer_scalar_t, 1> &buf) const {
+    return queue_stream()->attach_buffer(buf);
+  }
+  /// detach buffer
+  EIGEN_STRONG_INLINE void detach_buffer(void *p) const {
+    queue_stream()->detach_buffer(p);
+  }
   EIGEN_STRONG_INLINE ptrdiff_t get_offset(const void *ptr) const {
     return queue_stream()->get_offset(ptr);
   }
diff --git a/unsupported/test/cxx11_tensor_device_sycl.cpp b/unsupported/test/cxx11_tensor_device_sycl.cpp
index a9b542c03..d7ff38d34 100644
--- a/unsupported/test/cxx11_tensor_device_sycl.cpp
+++ b/unsupported/test/cxx11_tensor_device_sycl.cpp
@@ -68,6 +68,31 @@ void test_device_exceptions(const Eigen::SyclDevice &sycl_device) {
   sycl_device.deallocate(gpu_data);
 }
 
+template<typename DataType, int DataLayout, typename IndexType>
+void test_device_attach_buffer(const Eigen::SyclDevice &sycl_device) {
+  IndexType sizeDim1 = 100;
+  
+  array<IndexType, 1> tensorRange = {{sizeDim1}};
+  Tensor<DataType, 1, DataLayout, IndexType> in(tensorRange);
+  
+  cl::sycl::buffer<buffer_scalar_t, 1> buffer(cl::sycl::range<1>(sizeDim1 * sizeof(DataType)));
+  DataType* gpu_in_data = static_cast<DataType*>(sycl_device.attach_buffer(buffer));
+  
+  // fill
+  DataType value = DataType(7);
+  std::fill_n(in.data(), in.size(), value);
+  sycl_device.fill(gpu_in_data, gpu_in_data + in.size(), value);
+  
+  // Check that buffer is filled with the correct value.
+  auto reint = buffer.reinterpret<DataType>(cl::sycl::range<1>(sizeDim1));
+  auto access = reint.template get_access<cl::sycl::access::mode::read>();
+  for (IndexType i=0; i<in.size(); i++) {
+    VERIFY_IS_EQUAL(in(i), access[i]);
+  }
+  
+  sycl_device.detach_buffer(gpu_in_data);
+}
+
 template<typename DataType> void sycl_device_test_per_device(const cl::sycl::device& d){
   std::cout << "Running on " << d.template get_info<cl::sycl::info::device::name>() << std::endl;
   QueueInterface queueInterface(d);
@@ -78,6 +103,7 @@ template<typename DataType> void sycl_device_test_per_device(const cl::sycl::dev
   //test_device_exceptions<DataType, RowMajor>(sycl_device);
   /// this test throw an exception. enable it if you want to see the exception
   //test_device_exceptions<DataType, ColMajor>(sycl_device);
+  test_device_attach_buffer<DataType, ColMajor, int64_t>(sycl_device);
 }
 
 EIGEN_DECLARE_TEST(cxx11_tensor_device_sycl) {