@@ -601,6 +601,9 @@ class host_buffer {
601
601
}
602
602
};
603
603
604
+ class memcpy_3d_detail ;
605
+ class compat_memcpy_3d_detail_usmnone ;
606
+
604
607
// / copy 3D matrix specified by \p size from 3D matrix specified by \p from_ptr
605
608
// / and \p from_range to another specified by \p to_ptr and \p to_range.
606
609
static inline std::vector<sycl::event>
@@ -703,7 +706,7 @@ memcpy(sycl::queue q, void *to_ptr, const void *from_ptr,
703
706
sycl::access::target::device>
704
707
from_acc (from_alloc.buffer , cgh,
705
708
get_copy_range (size, from_slice, from_range.get (0 )), from_o);
706
- cgh.parallel_for <class compat_memcpy_3d_detail_usmnone >(
709
+ cgh.parallel_for <compat_memcpy_3d_detail_usmnone>(
707
710
size, [=](sycl::id<3 > id) {
708
711
to_acc[get_offset (id, to_slice, to_range.get (0 ))] =
709
712
from_acc[get_offset (id, from_slice, from_range.get (0 ))];
@@ -713,7 +716,7 @@ memcpy(sycl::queue q, void *to_ptr, const void *from_ptr,
713
716
#else
714
717
event_list.push_back (q.submit ([&](sycl::handler &cgh) {
715
718
cgh.depends_on (dep_events);
716
- cgh.parallel_for <class memcpy_3d_detail >(size, [=](sycl::id<3 > id) {
719
+ cgh.parallel_for <memcpy_3d_detail>(size, [=](sycl::id<3 > id) {
717
720
to_surface[get_offset (id, to_slice, to_range.get (0 ))] =
718
721
from_surface[get_offset (id, from_slice, from_range.get (0 ))];
719
722
});
0 commit comments