From 0af971309a4ef31ab0767391b2a9f56770371190 Mon Sep 17 00:00:00 2001 From: Ivan Karachun Date: Thu, 2 Apr 2020 23:15:53 +0300 Subject: [PATCH 1/2] [SYCL] Fixed sub-buffer memory allocation update In some cases parent`s memory allocation might change (e.g., after map/unmap operations). If parent`s memory allocation changes, sub-buffer memory allocation should be changed as well. Signed-off-by: Ivan Karachun --- sycl/source/detail/scheduler/commands.cpp | 5 +++-- sycl/source/detail/scheduler/commands.hpp | 15 ++++++++++++++- sycl/source/detail/scheduler/graph_builder.cpp | 1 + sycl/test/basic_tests/buffer/subbuffer.cpp | 2 +- 4 files changed, 19 insertions(+), 4 deletions(-) diff --git a/sycl/source/detail/scheduler/commands.cpp b/sycl/source/detail/scheduler/commands.cpp index b0e00e505010d..41575fada6617 100644 --- a/sycl/source/detail/scheduler/commands.cpp +++ b/sycl/source/detail/scheduler/commands.cpp @@ -694,7 +694,9 @@ void AllocaSubBufCommand::emitInstrumentationData() { #endif } -cl_int AllocaSubBufCommand::enqueueImp() { +cl_int AllocaSubBufCommand::enqueueImp() { return CL_SUCCESS; } + +void AllocaSubBufCommand::updateMemAllocation() { std::vector EventImpls = Command::prepareEvents(detail::getSyclObjImpl(MQueue->get_context())); RT::PiEvent &Event = MEvent->getHandleRef(); @@ -704,7 +706,6 @@ cl_int AllocaSubBufCommand::enqueueImp() { MParentAlloca->getMemAllocation(), MRequirement.MElemSize, MRequirement.MOffsetInBytes, MRequirement.MAccessRange, std::move(EventImpls), Event); - return CL_SUCCESS; } void AllocaSubBufCommand::printDot(std::ostream &Stream) const { diff --git a/sycl/source/detail/scheduler/commands.hpp b/sycl/source/detail/scheduler/commands.hpp index 004045db3e8e2..89dbd286a7e34 100644 --- a/sycl/source/detail/scheduler/commands.hpp +++ b/sycl/source/detail/scheduler/commands.hpp @@ -264,7 +264,7 @@ class AllocaCommandBase : public Command { SYCLMemObjI *getSYCLMemObj() const { return MRequirement.MSYCLMemObj; } - void *getMemAllocation() const { return MMemAllocation; } + virtual void *getMemAllocation() = 0; const Requirement *getRequirement() const final { return &MRequirement; } @@ -298,6 +298,7 @@ class AllocaCommand : public AllocaCommandBase { bool InitFromUserData = true, AllocaCommandBase *LinkedAllocaCmd = nullptr); + void *getMemAllocation() final { return MMemAllocation; } void printDot(std::ostream &Stream) const final; void emitInstrumentationData(); @@ -314,14 +315,26 @@ class AllocaSubBufCommand : public AllocaCommandBase { AllocaSubBufCommand(QueueImplPtr Queue, Requirement Req, AllocaCommandBase *ParentAlloca); + void *getMemAllocation() final { + // In some cases parent`s memory allocation might change (e.g., after + // map/unmap operations). If parent`s memory allocation changes, sub-buffer + // memory allocation should be changed as well. + if (MParentMemCache != MParentAlloca->getMemAllocation()) { + MParentMemCache = MParentAlloca->getMemAllocation(); + updateMemAllocation(); + } + return MMemAllocation; + } void printDot(std::ostream &Stream) const final; AllocaCommandBase *getParentAlloca() { return MParentAlloca; } void emitInstrumentationData(); private: cl_int enqueueImp() final; + void updateMemAllocation(); AllocaCommandBase *MParentAlloca = nullptr; + void *MParentMemCache = nullptr; }; class MapMemObject : public Command { diff --git a/sycl/source/detail/scheduler/graph_builder.cpp b/sycl/source/detail/scheduler/graph_builder.cpp index 2105c22cf6af7..c6934cebf9e00 100644 --- a/sycl/source/detail/scheduler/graph_builder.cpp +++ b/sycl/source/detail/scheduler/graph_builder.cpp @@ -533,6 +533,7 @@ AllocaCommandBase *Scheduler::GraphBuilder::findAllocaForReq( bool Res = sameCtx(AllocaCmd->getQueue()->getContextImplPtr(), Context); if (IsSuitableSubReq(Req)) { const Requirement *TmpReq = AllocaCmd->getRequirement(); + Res &= AllocaCmd->getType() == Command::CommandType::ALLOCA_SUB_BUF; Res &= TmpReq->MOffsetInBytes == Req->MOffsetInBytes; Res &= TmpReq->MSYCLMemObj->getSize() == Req->MSYCLMemObj->getSize(); } diff --git a/sycl/test/basic_tests/buffer/subbuffer.cpp b/sycl/test/basic_tests/buffer/subbuffer.cpp index 8842fd2b73464..c0274a0ddb7ab 100644 --- a/sycl/test/basic_tests/buffer/subbuffer.cpp +++ b/sycl/test/basic_tests/buffer/subbuffer.cpp @@ -279,7 +279,7 @@ void checkMultipleContexts() { { sycl::queue queue1; sycl::buffer buf(a, sycl::range<1>(N)); - sycl::buffer subbuf1(buf, sycl::id<1>(0), sycl::range<1>(N / 2)); + sycl::buffer subbuf1(buf, sycl::id<1>(N / 2), sycl::range<1>(N / 2)); queue1.submit([&](sycl::handler &cgh) { auto bufacc = subbuf1.get_access(cgh); cgh.parallel_for( From 12a18fbd3d8a002c0af82afa8cfdf26c31a34c3a Mon Sep 17 00:00:00 2001 From: Ivan Karachun Date: Tue, 7 Apr 2020 17:32:36 +0300 Subject: [PATCH 2/2] Simplified solution Signed-off-by: Ivan Karachun --- sycl/source/detail/scheduler/commands.cpp | 15 +++++++++++++-- sycl/source/detail/scheduler/commands.hpp | 17 +++-------------- 2 files changed, 16 insertions(+), 16 deletions(-) diff --git a/sycl/source/detail/scheduler/commands.cpp b/sycl/source/detail/scheduler/commands.cpp index 41575fada6617..bb783468ad604 100644 --- a/sycl/source/detail/scheduler/commands.cpp +++ b/sycl/source/detail/scheduler/commands.cpp @@ -694,9 +694,19 @@ void AllocaSubBufCommand::emitInstrumentationData() { #endif } -cl_int AllocaSubBufCommand::enqueueImp() { return CL_SUCCESS; } +void *AllocaSubBufCommand::getMemAllocation() const { + // In some cases parent`s memory allocation might change (e.g., after + // map/unmap operations). If parent`s memory allocation changes, sub-buffer + // memory allocation should be changed as well. + if (MQueue->is_host()) { + return static_cast( + static_cast(MParentAlloca->getMemAllocation()) + + MRequirement.MOffsetInBytes); + } + return MMemAllocation; +} -void AllocaSubBufCommand::updateMemAllocation() { +cl_int AllocaSubBufCommand::enqueueImp() { std::vector EventImpls = Command::prepareEvents(detail::getSyclObjImpl(MQueue->get_context())); RT::PiEvent &Event = MEvent->getHandleRef(); @@ -706,6 +716,7 @@ void AllocaSubBufCommand::updateMemAllocation() { MParentAlloca->getMemAllocation(), MRequirement.MElemSize, MRequirement.MOffsetInBytes, MRequirement.MAccessRange, std::move(EventImpls), Event); + return CL_SUCCESS; } void AllocaSubBufCommand::printDot(std::ostream &Stream) const { diff --git a/sycl/source/detail/scheduler/commands.hpp b/sycl/source/detail/scheduler/commands.hpp index 89dbd286a7e34..8cbe2ccf8ff98 100644 --- a/sycl/source/detail/scheduler/commands.hpp +++ b/sycl/source/detail/scheduler/commands.hpp @@ -264,7 +264,7 @@ class AllocaCommandBase : public Command { SYCLMemObjI *getSYCLMemObj() const { return MRequirement.MSYCLMemObj; } - virtual void *getMemAllocation() = 0; + virtual void *getMemAllocation() const = 0; const Requirement *getRequirement() const final { return &MRequirement; } @@ -298,7 +298,7 @@ class AllocaCommand : public AllocaCommandBase { bool InitFromUserData = true, AllocaCommandBase *LinkedAllocaCmd = nullptr); - void *getMemAllocation() final { return MMemAllocation; } + void *getMemAllocation() const final { return MMemAllocation; } void printDot(std::ostream &Stream) const final; void emitInstrumentationData(); @@ -315,26 +315,15 @@ class AllocaSubBufCommand : public AllocaCommandBase { AllocaSubBufCommand(QueueImplPtr Queue, Requirement Req, AllocaCommandBase *ParentAlloca); - void *getMemAllocation() final { - // In some cases parent`s memory allocation might change (e.g., after - // map/unmap operations). If parent`s memory allocation changes, sub-buffer - // memory allocation should be changed as well. - if (MParentMemCache != MParentAlloca->getMemAllocation()) { - MParentMemCache = MParentAlloca->getMemAllocation(); - updateMemAllocation(); - } - return MMemAllocation; - } + void *getMemAllocation() const final; void printDot(std::ostream &Stream) const final; AllocaCommandBase *getParentAlloca() { return MParentAlloca; } void emitInstrumentationData(); private: cl_int enqueueImp() final; - void updateMemAllocation(); AllocaCommandBase *MParentAlloca = nullptr; - void *MParentMemCache = nullptr; }; class MapMemObject : public Command {