Skip to content

[SYCL] Add unittest for recent handler::set_arg fixes #7333

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

Merged
Merged
Show file tree
Hide file tree
Changes from all commits
Commits
File filter

Filter by extension

Filter by extension

Conversations
Failed to load comments.
Loading
Jump to
Jump to file
Failed to load files.
Loading
Diff view
Diff view
2 changes: 1 addition & 1 deletion sycl/source/detail/kernel_info.hpp
Original file line number Diff line number Diff line change
Expand Up @@ -28,7 +28,7 @@ typename std::enable_if<
get_kernel_info(RT::PiKernel Kernel, const plugin &Plugin) {
static_assert(detail::is_kernel_info_desc<Param>::value,
"Invalid kernel information descriptor");
size_t ResultSize;
size_t ResultSize = 0;

// TODO catch an exception and put it to list of asynchronous exceptions
Plugin.call<PiApiKind::piKernelGetInfo>(Kernel, PiInfoCode<Param>::value, 0,
Expand Down
1 change: 1 addition & 0 deletions sycl/unittests/CMakeLists.txt
Original file line number Diff line number Diff line change
Expand Up @@ -48,3 +48,4 @@ add_subdirectory(event)
add_subdirectory(buffer)
add_subdirectory(context)
add_subdirectory(accessor)
add_subdirectory(handler)
3 changes: 3 additions & 0 deletions sycl/unittests/handler/CMakeLists.txt
Original file line number Diff line number Diff line change
@@ -0,0 +1,3 @@
add_sycl_unittest(HandlerTests OBJECT
SetArgForLocalAccessor.cpp
)
157 changes: 157 additions & 0 deletions sycl/unittests/handler/SetArgForLocalAccessor.cpp
Original file line number Diff line number Diff line change
@@ -0,0 +1,157 @@
#include <gtest/gtest.h>
#include <helpers/PiMock.hpp>

#include <sycl/sycl.hpp>

// This test checks that we pass the correct buffer size value when setting
// local_accessor as an argument through handler::set_arg to a kernel created
// using OpenCL interoperability methods.

namespace {

struct TestContext {
size_t localBufferArgSize = 0;

// SYCL RT has number of checks that all devices and contexts are consistent
// between kernel, kernel_bundle and other objects.
//
// To ensure that those checks pass, we intercept some PI calls to extract
// the exact PI handles of device and context used in queue creation to later
// return them when program/context/kernel info is requested.
pi_device deviceHandle;
pi_context contextHandle;

pi_program programHandle = createDummyHandle<pi_program>();

~TestContext() { releaseDummyHandle<pi_program>(programHandle); }
};

TestContext GlobalContext;

} // namespace

pi_result redefined_piKernelSetArg(pi_kernel kernel, pi_uint32 arg_index,
size_t arg_size, const void *arg_value) {
GlobalContext.localBufferArgSize = arg_size;

return PI_SUCCESS;
}

pi_result after_piContextGetInfo(pi_context context, pi_context_info param_name,
size_t param_value_size, void *param_value,
size_t *param_value_size_ret) {
switch (param_name) {
case PI_CONTEXT_INFO_DEVICES:
if (param_value)
*static_cast<pi_device *>(param_value) = GlobalContext.deviceHandle;
if (param_value_size_ret)
*param_value_size_ret = sizeof(GlobalContext.deviceHandle);
default:;
}

return PI_SUCCESS;
}

pi_result after_piProgramGetInfo(pi_program program, pi_program_info param_name,
size_t param_value_size, void *param_value,
size_t *param_value_size_ret) {

switch (param_name) {
case PI_PROGRAM_INFO_DEVICES:
if (param_value_size_ret)
*param_value_size_ret = sizeof(GlobalContext.deviceHandle);
if (param_value)
*static_cast<pi_device *>(param_value) = GlobalContext.deviceHandle;
break;
default:;
}

return PI_SUCCESS;
}

pi_result redefined_piProgramGetBuildInfo(pi_program program, pi_device device,
_pi_program_build_info param_name,
size_t param_value_size,
void *param_value,
size_t *param_value_size_ret) {
switch (param_name) {
case PI_PROGRAM_BUILD_INFO_BINARY_TYPE:
if (param_value_size_ret)
*param_value_size_ret = sizeof(pi_program_binary_type);
if (param_value)
*static_cast<pi_program_binary_type *>(param_value) =
PI_PROGRAM_BINARY_TYPE_EXECUTABLE;
break;
default:;
}

return PI_SUCCESS;
}

pi_result after_piContextCreate(const pi_context_properties *properties,
pi_uint32 num_devices, const pi_device *devices,
void (*pfn_notify)(const char *errinfo,
const void *private_info,
size_t cb, void *user_data),
void *user_data, pi_context *ret_context) {
if (ret_context)
GlobalContext.contextHandle = *ret_context;
GlobalContext.deviceHandle = *devices;
return PI_SUCCESS;
}

pi_result after_piKernelGetInfo(pi_kernel kernel, pi_kernel_info param_name,
size_t param_value_size, void *param_value,
size_t *param_value_size_ret) {
switch (param_name) {
case PI_KERNEL_INFO_CONTEXT:
if (param_value_size_ret)
*param_value_size_ret = sizeof(GlobalContext.contextHandle);
if (param_value)
*static_cast<pi_context *>(param_value) = GlobalContext.contextHandle;
break;
case PI_KERNEL_INFO_PROGRAM:
if (param_value_size_ret)
*param_value_size_ret = sizeof(GlobalContext.programHandle);
if (param_value)
*(pi_program *)param_value = GlobalContext.programHandle;
break;
default:;
}

return PI_SUCCESS;
}

TEST(HandlerSetArg, LocalAccessor) {
sycl::unittest::PiMock Mock;

Mock.redefine<sycl::detail::PiApiKind::piKernelSetArg>(
redefined_piKernelSetArg);
Mock.redefineAfter<sycl::detail::PiApiKind::piContextCreate>(
after_piContextCreate);
Mock.redefineAfter<sycl::detail::PiApiKind::piProgramGetInfo>(
after_piProgramGetInfo);
Mock.redefineAfter<sycl::detail::PiApiKind::piContextGetInfo>(
after_piContextGetInfo);
Mock.redefineAfter<sycl::detail::PiApiKind::piKernelGetInfo>(
after_piKernelGetInfo);
Mock.redefine<sycl::detail::PiApiKind::piProgramGetBuildInfo>(
redefined_piProgramGetBuildInfo);

constexpr size_t Size = 128;
sycl::queue Q;

DummyHandleT handle;
auto KernelCL = reinterpret_cast<typename sycl::backend_traits<
sycl::backend::opencl>::template input_type<sycl::kernel>>(&handle);
auto Kernel =
sycl::make_kernel<sycl::backend::opencl>(KernelCL, Q.get_context());

Q.submit([&](sycl::handler &CGH) {
sycl::local_accessor<float, 1> Acc(Size, CGH);
CGH.set_arg(0, Acc);
CGH.single_task(Kernel);
}).wait();

ASSERT_EQ(GlobalContext.localBufferArgSize, Size * sizeof(float));
}