Skip to content
New issue

Have a question about this project? Sign up for a free GitHub account to open an issue and contact its maintainers and the community.

By clicking “Sign up for GitHub”, you agree to our terms of service and privacy statement. We’ll occasionally send you account related emails.

Already on GitHub? Sign in to your account

[SYCL] Queue shortcut functions with placeholder accessors #7266

Merged
merged 12 commits into from
Nov 14, 2022
118 changes: 118 additions & 0 deletions sycl/include/sycl/queue.hpp
Original file line number Diff line number Diff line change
Expand Up @@ -1059,6 +1059,124 @@ class __SYCL_EXPORT queue {
CodeLoc);
}

/// Copies data from a memory region pointed to by a placeholder accessor to
/// another memory region pointed to by a shared_ptr.
///
/// \param Src is a placeholder accessor to the source memory.
Copy link
Contributor

Choose a reason for hiding this comment

The reason will be displayed to describe this comment to others. Learn more.

Not sure the API works with placeholder accessors only, does it?

Copy link
Contributor

Choose a reason for hiding this comment

The reason will be displayed to describe this comment to others. Learn more.

Technically it should be as this is a function on queue, so at this level it should only be possible to have placeholder and host accessors, the latter being moved to host_accessor. We could have a check to make sure it is not a host accessor in the meantime.

/// \param Dest is a shared_ptr to the destination memory.
/// \return an event representing copy operation.
template <typename SrcT, int SrcDims, access_mode SrcMode, target SrcTgt,
access::placeholder IsPlaceholder, typename DestT>
event copy(accessor<SrcT, SrcDims, SrcMode, SrcTgt, IsPlaceholder> Src,
std::shared_ptr<DestT> Dest _CODELOCPARAM(&CodeLoc)) {
return submit([&](handler &CGH) {
CGH.require(Src);
CGH.copy(Src, Dest);
} _CODELOCFW(CodeLoc));
}

/// Copies data from a memory region pointed to by a shared_ptr to another
/// memory region pointed to by a placeholder accessor.
///
/// \param Src is a shared_ptr to the source memory.
/// \param Dest is a placeholder accessor to the destination memory.
/// \return an event representing copy operation.
template <typename SrcT, typename DestT, int DestDims, access_mode DestMode,
target DestTgt, access::placeholder IsPlaceholder>
event copy(std::shared_ptr<SrcT> Src,
accessor<DestT, DestDims, DestMode, DestTgt, IsPlaceholder> Dest
_CODELOCPARAM(&CodeLoc)) {
return submit([&](handler &CGH) {
CGH.require(Dest);
CGH.copy(Src, Dest);
} _CODELOCFW(CodeLoc));
}

/// Copies data from a memory region pointed to by a placeholder accessor to
/// another memory region pointed to by a raw pointer.
///
/// \param Src is a placeholder accessor to the source memory.
/// \param Dest is a raw pointer to the destination memory.
/// \return an event representing copy operation.
template <typename SrcT, int SrcDims, access_mode SrcMode, target SrcTgt,
access::placeholder IsPlaceholder, typename DestT>
event copy(accessor<SrcT, SrcDims, SrcMode, SrcTgt, IsPlaceholder> Src,
DestT *Dest _CODELOCPARAM(&CodeLoc)) {
return submit([&](handler &CGH) {
CGH.require(Src);
CGH.copy(Src, Dest);
} _CODELOCFW(CodeLoc));
}

/// Copies data from a memory region pointed to by a raw pointer to another
/// memory region pointed to by a placeholder accessor.
///
/// \param Src is a raw pointer to the source memory.
/// \param Dest is a placeholder accessor to the destination memory.
/// \return an event representing copy operation.
template <typename SrcT, typename DestT, int DestDims, access_mode DestMode,
target DestTgt, access::placeholder IsPlaceholder>
event copy(const SrcT *Src,
accessor<DestT, DestDims, DestMode, DestTgt, IsPlaceholder> Dest
_CODELOCPARAM(&CodeLoc)) {
return submit([&](handler &CGH) {
CGH.require(Dest);
CGH.copy(Src, Dest);
} _CODELOCFW(CodeLoc));
}

/// Copies data from one memory region to another, both pointed by placeholder
/// accessors.
///
/// \param Src is a placeholder accessor to the source memory.
/// \param Dest is a placeholder accessor to the destination memory.
/// \return an event representing copy operation.
template <typename SrcT, int SrcDims, access_mode SrcMode, target SrcTgt,
access::placeholder IsSrcPlaceholder, typename DestT, int DestDims,
access_mode DestMode, target DestTgt,
access::placeholder IsDestPlaceholder>
event
copy(accessor<SrcT, SrcDims, SrcMode, SrcTgt, IsSrcPlaceholder> Src,
accessor<DestT, DestDims, DestMode, DestTgt, IsDestPlaceholder> Dest
_CODELOCPARAM(&CodeLoc)) {
return submit([&](handler &CGH) {
CGH.require(Src);
CGH.require(Dest);
CGH.copy(Src, Dest);
} _CODELOCFW(CodeLoc));
}

/// Provides guarantees that the memory object accessed via Acc is updated
/// on the host after operation is complete.
///
/// \param Acc is a SYCL accessor that needs to be updated on host.
/// \return an event representing update_host operation.
template <typename T, int Dims, access_mode Mode, target Tgt,
access::placeholder IsPlaceholder>
event update_host(
accessor<T, Dims, Mode, Tgt, IsPlaceholder> Acc _CODELOCPARAM(&CodeLoc)) {
return submit([&](handler &CGH) {
CGH.require(Acc);
CGH.update_host(Acc);
} _CODELOCFW(CodeLoc));
}

/// Fills the specified memory with the specified data.
///
/// \param Dest is the placeholder accessor to the memory to fill.
/// \param Src is the data to fill the memory with. T should be
/// trivially copyable.
/// \return an event representing fill operation.
template <typename T, int Dims, access_mode Mode, target Tgt,
access::placeholder IsPlaceholder>
event fill(accessor<T, Dims, Mode, Tgt, IsPlaceholder> Dest,
const T &Src _CODELOCPARAM(&CodeLoc)) {
return submit([&](handler &CGH) {
CGH.require(Dest);
CGH.fill<T>(Dest, Src);
} _CODELOCFW(CodeLoc));
}

// Clean KERNELFUNC macros.
#undef _KERNELFUNCPARAM

Expand Down
1 change: 1 addition & 0 deletions sycl/unittests/queue/CMakeLists.txt
Original file line number Diff line number Diff line change
Expand Up @@ -4,4 +4,5 @@ add_sycl_unittest(QueueTests OBJECT
USM.cpp
Wait.cpp
GetProfilingInfo.cpp
ShortcutFunctions.cpp
)
235 changes: 235 additions & 0 deletions sycl/unittests/queue/ShortcutFunctions.cpp
Original file line number Diff line number Diff line change
@@ -0,0 +1,235 @@
//==-------------- ShortcutFunctions.cpp --- queue unit tests --------------==//
//
// Part of the LLVM Project, under the Apache License v2.0 with LLVM Exceptions.
// See https://llvm.org/LICENSE.txt for license information.
// SPDX-License-Identifier: Apache-2.0 WITH LLVM-exception
//
//===----------------------------------------------------------------------===//

#include <detail/context_impl.hpp>
#include <gtest/gtest.h>
#include <helpers/PiMock.hpp>
#include <sycl/ext/oneapi/accessor_property_list.hpp>
#include <sycl/handler.hpp>
#include <sycl/queue.hpp>
#include <sycl/sycl.hpp>

#include <memory>

using namespace sycl;

namespace {
struct TestCtx {
bool BufferFillCalled = false;
bool BufferReadCalled = false;
bool BufferWriteCalled = false;
bool BufferCopyCalled = false;
};
} // namespace

static std::unique_ptr<TestCtx> TestContext;

pi_result redefinedEnqueueMemBufferWrite(pi_queue command_queue, pi_mem buffer,
pi_bool blocking_write, size_t offset,
size_t size, const void *ptr,
pi_uint32 num_events_in_wait_list,
const pi_event *event_wait_list,
pi_event *event) {
TestContext->BufferWriteCalled = true;
return PI_SUCCESS;
}

pi_result redefinedEnqueueMemBufferRead(pi_queue queue, pi_mem buffer,
pi_bool blocking_read, size_t offset,
size_t size, void *ptr,
pi_uint32 num_events_in_wait_list,
const pi_event *event_wait_list,
pi_event *event) {
TestContext->BufferReadCalled = true;
return PI_SUCCESS;
}

pi_result redefinedEnqueueMemBufferCopy(pi_queue command_queue,
pi_mem src_buffer, pi_mem dst_buffer,
size_t src_offset, size_t dst_offset,
size_t size,
pi_uint32 num_events_in_wait_list,
const pi_event *event_wait_list,
pi_event *event) {
TestContext->BufferCopyCalled = true;
return PI_SUCCESS;
}

pi_result redefinedEnqueueMemBufferFill(pi_queue command_queue, pi_mem buffer,
const void *pattern,
size_t pattern_size, size_t offset,
size_t size,
pi_uint32 num_events_in_wait_list,
const pi_event *event_wait_list,
pi_event *event) {
TestContext->BufferFillCalled = true;
return PI_SUCCESS;
}

TEST(ShortcutFunctions, ShortcutsCallCorrectPIFunctions) {
unittest::PiMock Mock;
platform Plt = Mock.getPlatform();

Mock.redefine<detail::PiApiKind::piEnqueueMemBufferWrite>(
Copy link
Contributor

Choose a reason for hiding this comment

The reason will be displayed to describe this comment to others. Learn more.

Minor. It's not critical in this specific case, but it would be better to use redefineAfter or redefineBefore to let the original mock function execute.

redefinedEnqueueMemBufferWrite);
Mock.redefine<detail::PiApiKind::piEnqueueMemBufferRead>(
redefinedEnqueueMemBufferRead);
Mock.redefine<detail::PiApiKind::piEnqueueMemBufferCopy>(
redefinedEnqueueMemBufferCopy);

Mock.redefine<detail::PiApiKind::piEnqueueMemBufferFill>(
redefinedEnqueueMemBufferFill);

context Ctx(Plt);
queue Q{Ctx, default_selector()};

constexpr std::size_t Size = 1;

// Queue.copy(accessor src, shared_ptr dest);
{
TestContext.reset(new TestCtx());

int Data[Size];
buffer<int> Buf(Data, Size);

accessor<int, 1, access::mode::read, access::target::device,
access::placeholder::true_t, ext::oneapi::accessor_property_list<>>
Src(Buf);
ASSERT_TRUE(Src.is_placeholder());

std::shared_ptr<int> Dest = std::make_shared<int>(0);

Q.copy(Src, Dest);
Q.wait();

EXPECT_TRUE(TestContext->BufferReadCalled);
}

// Queue.copy(shared_ptr src, accessor dest);
{
TestContext.reset(new TestCtx());

int Data[Size];
buffer<int> Buf(Data, Size);

std::shared_ptr<int> Src = std::make_shared<int>(42);

accessor<int, 1, access::mode::write, access::target::device,
access::placeholder::true_t, ext::oneapi::accessor_property_list<>>
Comment on lines +122 to +123
Copy link
Contributor

Choose a reason for hiding this comment

The reason will be displayed to describe this comment to others. Learn more.

I think SYCL 2020 makes access::placeholder::true_t unnecessary. I'd suggest to either simplify or at least add as extra simple

  accessor Dest(Buf, sycl::write);

test.

Dest(Buf);
ASSERT_TRUE(Dest.is_placeholder());

Q.copy(Src, Dest);
Q.wait();

EXPECT_TRUE(TestContext->BufferWriteCalled);
}

// Queue.copy(accessor src, ptr* dest);
{
TestContext.reset(new TestCtx());

int Data[Size];
buffer<int> Buf(Data, Size);

accessor<int, 1, access::mode::read, access::target::device,
access::placeholder::true_t, ext::oneapi::accessor_property_list<>>
Src(Buf);
ASSERT_TRUE(Src.is_placeholder());

std::unique_ptr<int> Dest = std::make_unique<int>(0);

Q.copy(Src, Dest.get());
Q.wait();

EXPECT_TRUE(TestContext->BufferReadCalled);
}

// Queue.copy(ptr* src, accessor dest);
{
TestContext.reset(new TestCtx());

int Data[Size];
buffer<int> Buf(Data, Size);

std::unique_ptr<int> Src = std::make_unique<int>(42);

accessor<int, 1, access::mode::write, access::target::device,
access::placeholder::true_t, ext::oneapi::accessor_property_list<>>
Dest(Buf);
ASSERT_TRUE(Dest.is_placeholder());

Q.copy(Src.get(), Dest);
Q.wait();

EXPECT_TRUE(TestContext->BufferWriteCalled);
}

// Queue.copy(accessor src, accessor dest);
{
TestContext.reset(new TestCtx());

int SrcData[Size];
buffer<int> SrcBuf(SrcData, Size);

int DestData[Size];
buffer<int> DestBuf(DestData, Size);

accessor<int, 1, access::mode::read, access::target::device,
access::placeholder::true_t, ext::oneapi::accessor_property_list<>>
Src(SrcBuf);
accessor<int, 1, access::mode::write, access::target::device,
access::placeholder::true_t, ext::oneapi::accessor_property_list<>>
Dest(DestBuf);

ASSERT_TRUE(Src.is_placeholder());
ASSERT_TRUE(Dest.is_placeholder());

Q.copy(Src, Dest);
Q.wait();

EXPECT_TRUE(TestContext->BufferCopyCalled);
}

// Queue.update_host(accessor acc);
{
TestContext.reset(new TestCtx());

int Data[Size];
buffer<int> Buf(Data, Size);

accessor<int, 1, access::mode::read_write, access::target::device,
access::placeholder::true_t, ext::oneapi::accessor_property_list<>>
Acc(Buf);

ASSERT_TRUE(Acc.is_placeholder());

Q.update_host(Acc);
Q.wait();

// No PI functions expected.
}

// Queue.fill<T>(accessor Dest, T src)
{
TestContext.reset(new TestCtx());

int Data[Size];
buffer<int> Buf(Data, Size);

accessor<int, 1, access::mode::read_write, access::target::device,
access::placeholder::true_t, ext::oneapi::accessor_property_list<>>
Acc(Buf);
ASSERT_TRUE(Acc.is_placeholder());

Q.fill(Acc, 42);
Q.wait();

EXPECT_TRUE(TestContext->BufferFillCalled);
}
}