Skip to content

Commit

Permalink
[SYCL] Allow host access for interoperability buffers
Browse files Browse the repository at this point in the history 
This change allows using set_final_data and host accessors with
interoperability buffer

Signed-off-by: Mariya Podchishchaeva <mariya.podchishchaeva@intel.com>
  • Loading branch information
@Fznamznon @vladimirlaz
Fznamznon authored and vladimirlaz committed Apr 9, 2019
1 parent c9f21fa commit 39e2c74
Show file tree
Hide file tree
Showing 3 changed files with 80 additions and 46 deletions.
27 changes: 6 additions & 21 deletions sycl/include/CL/sycl/accessor.hpp
View file
Original file line number Diff line number Diff line change
Expand Up @@ -730,13 +730,8 @@ class accessor
#endif
auto BufImpl = detail::getSyclObjImpl(bufferRef);
if (AccessTarget == access::target::host_buffer) {
if (BufImpl->OpenCLInterop) {
throw cl::sycl::runtime_error(
"Host access to interoperability buffer is not allowed");
} else {
simple_scheduler::Scheduler::getInstance()
.copyBack<AccessMode, AccessTarget>(*BufImpl);
}
simple_scheduler::Scheduler::getInstance()
.copyBack<AccessMode, AccessTarget>(*BufImpl);
}
if (BufImpl->OpenCLInterop && !BufImpl->isValidAccessToMem(accessMode)) {
throw cl::sycl::runtime_error(
Expand Down Expand Up @@ -818,13 +813,8 @@ class accessor
#endif
auto BufImpl = detail::getSyclObjImpl(bufferRef);
if (AccessTarget == access::target::host_buffer) {
if (BufImpl->OpenCLInterop) {
throw cl::sycl::runtime_error(
"Host access to interoperability buffer is not allowed");
} else {
simple_scheduler::Scheduler::getInstance()
.copyBack<AccessMode, AccessTarget>(*BufImpl);
}
simple_scheduler::Scheduler::getInstance()
.copyBack<AccessMode, AccessTarget>(*BufImpl);
}
if (BufImpl->OpenCLInterop && !BufImpl->isValidAccessToMem(accessMode)) {
throw cl::sycl::runtime_error(
Expand Down Expand Up @@ -913,13 +903,8 @@ class accessor
bufferRef.MemRange, Offset)) {
auto BufImpl = detail::getSyclObjImpl(bufferRef);
if (AccessTarget == access::target::host_buffer) {
if (BufImpl->OpenCLInterop) {
throw cl::sycl::runtime_error(
"Host access to interoperability buffer is not allowed");
} else {
simple_scheduler::Scheduler::getInstance()
.copyBack<AccessMode, AccessTarget>(*BufImpl);
}
simple_scheduler::Scheduler::getInstance()
.copyBack<AccessMode, AccessTarget>(*BufImpl);
}
if (BufImpl->OpenCLInterop && !BufImpl->isValidAccessToMem(accessMode)) {
throw cl::sycl::runtime_error(
Expand Down
26 changes: 5 additions & 21 deletions sycl/include/CL/sycl/detail/buffer_impl.hpp
View file
Original file line number Diff line number Diff line change
Expand Up @@ -164,16 +164,16 @@ template <typename AllocatorT> class buffer_impl {
"Input context must be the same as the context of cl_mem");
OCLState.Mem = MemObject;
CHECK_OCL_CODE(clRetainMemObject(MemObject));

BufData.resize(get_size());
BufPtr = reinterpret_cast<void *>(BufData.data());
}

size_t get_size() const { return SizeInBytes; }

~buffer_impl() {
if (!OpenCLInterop)
// TODO. Use node instead?
simple_scheduler::Scheduler::getInstance()
.copyBack<access::mode::read_write, access::target::host_buffer>(
*this);
simple_scheduler::Scheduler::getInstance()
.copyBack<access::mode::read_write, access::target::host_buffer>(*this);

if (uploadData != nullptr && NeedWriteBack) {
uploadData();
Expand All @@ -189,9 +189,6 @@ template <typename AllocatorT> class buffer_impl {
void set_final_data(std::nullptr_t) { uploadData = nullptr; }

template <typename T> void set_final_data(weak_ptr_class<T> final_data) {
if (OpenCLInterop)
throw cl::sycl::runtime_error(
"set_final_data could not be used with interoperability buffer");
uploadData = [this, final_data]() {
if (auto finalData = final_data.lock()) {
T *Ptr = reinterpret_cast<T *>(BufPtr);
Expand All @@ -201,9 +198,6 @@ template <typename AllocatorT> class buffer_impl {
}

template <typename Destination> void set_final_data(Destination final_data) {
if (OpenCLInterop)
throw cl::sycl::runtime_error(
"set_final_data could not be used with interoperability buffer");
static_assert(!std::is_const<Destination>::value,
"Can not write in a constant Destination. Destination should "
"not be const.");
Expand Down Expand Up @@ -419,11 +413,6 @@ void buffer_impl<AllocatorT>::moveMemoryTo(

ContextImplPtr Context = detail::getSyclObjImpl(Queue->get_context());

if (OpenCLInterop && (Context->getHandleRef() != OpenCLContext))
throw cl::sycl::runtime_error(
"Interoperability buffer could not be used in a context other than the "
"context associated with the OpenCL memory object.");

// TODO: Move all implementation specific commands to separate file?
// TODO: Make allocation in separate command?

Expand Down Expand Up @@ -552,11 +541,6 @@ void buffer_impl<AllocatorT>::allocate(QueueImplPtr Queue,

ContextImplPtr Context = detail::getSyclObjImpl(Queue->get_context());

if (OpenCLInterop && (Context->getHandleRef() != OpenCLContext))
throw cl::sycl::runtime_error(
"Interoperability buffer could not be used in a context other than the "
"context associated with the OpenCL memory object.");

if (OpenCLInterop) {
// For interoperability instance of the SYCL buffer class being constructed
// must wait for the SYCL event parameter, if one is provided,
Expand Down
73 changes: 69 additions & 4 deletions sycl/test/basic_tests/buffer/buffer_interop.cpp
View file
Original file line number Diff line number Diff line change
Expand Up @@ -18,7 +18,7 @@ using namespace cl::sycl;
int main() {
bool Failed = false;
{
const size_t Size = 32;
constexpr size_t Size = 32;
int Init[Size] = {5};
cl_int Error = CL_SUCCESS;
cl::sycl::range<1> InteropRange;
Expand All @@ -31,7 +31,7 @@ int main() {
MyQueue.get_context().get(), CL_MEM_READ_WRITE | CL_MEM_COPY_HOST_PTR,
Size * sizeof(int), Init, &Error);
CHECK_OCL_CODE(Error);
buffer<int, 1> Buffer(OpenCLBuffer, MyQueue.get_context());
buffer<int, 1> Buffer{OpenCLBuffer, MyQueue.get_context()};

if (Buffer.get_range() != InteropRange) {
assert(false);
Expand All @@ -55,8 +55,8 @@ int main() {
int Data[Size] = {10};
std::vector<int> Result(Size, 0);
{
buffer<int, 1> BufferData(Data, range<1>(Size),
{property::buffer::use_host_ptr()});
buffer<int, 1> BufferData{Data, range<1>(Size),
{property::buffer::use_host_ptr()}};
BufferData.set_final_data(Result.begin());
MyQueue.submit([&](handler &CGH) {
auto Data = BufferData.get_access<access::mode::write>(CGH);
Expand All @@ -79,5 +79,70 @@ int main() {
}
}
}
// Check set_final_data
{
constexpr size_t Size = 32;
int Init[Size] = {5};
int Result[Size] = {5};
cl_int Error = CL_SUCCESS;

queue MyQueue;

cl_mem OpenCLBuffer = clCreateBuffer(
MyQueue.get_context().get(), CL_MEM_READ_WRITE | CL_MEM_COPY_HOST_PTR,
Size * sizeof(int), Init, &Error);
CHECK_OCL_CODE(Error);
{
buffer<int, 1> Buffer{OpenCLBuffer, MyQueue.get_context()};
Buffer.set_final_data(Result);

MyQueue.submit([&](handler &CGH) {
auto B = Buffer.get_access<access::mode::write>(CGH);
CGH.parallel_for<class FinalData>(
range<1>{Size}, [=](id<1> Index) { B[Index] = 10; });
});
}
Error = clReleaseMemObject(OpenCLBuffer);
CHECK_OCL_CODE(Error);
for (size_t i = 0; i < Size; ++i) {
if (Result[i] != 10) {
std::cout << " array[" << i << "] is " << Result[i] << " expected "
<< 10 << std::endl;
assert(false);
Failed = true;
}
}
}
// Check host accessor
{
constexpr size_t Size = 32;
int Init[Size] = {5};
cl_int Error = CL_SUCCESS;

queue MyQueue;

cl_mem OpenCLBuffer = clCreateBuffer(
MyQueue.get_context().get(), CL_MEM_READ_WRITE | CL_MEM_COPY_HOST_PTR,
Size * sizeof(int), Init, &Error);
CHECK_OCL_CODE(Error);
buffer<int, 1> Buffer{OpenCLBuffer, MyQueue.get_context()};

MyQueue.submit([&](handler &CGH) {
auto B = Buffer.get_access<access::mode::write>(CGH);
CGH.parallel_for<class HostAccess>(range<1>{Size},
[=](id<1> Index) { B[Index] = 10; });
});
auto Acc = Buffer.get_access<cl::sycl::access::mode::read>();
for (size_t i = 0; i < Size; ++i) {
if (Acc[i] != 10) {
std::cout << " array[" << i << "] is " << Acc[i] << " expected "
<< 10 << std::endl;
assert(false);
Failed = true;
}
}
Error = clReleaseMemObject(OpenCLBuffer);
CHECK_OCL_CODE(Error);
}
return Failed;
}

0 comments on commit 39e2c74

Please sign in to comment.