diff --git a/sycl/include/CL/sycl/accessor.hpp b/sycl/include/CL/sycl/accessor.hpp old mode 100644 new mode 100755 index 125944e6c10eb..52594b21eea6f --- a/sycl/include/CL/sycl/accessor.hpp +++ b/sycl/include/CL/sycl/accessor.hpp @@ -944,9 +944,9 @@ class accessor : template () && - IsValidTag() && IsPlaceH && - (IsGlobalBuf || IsConstantBuf)>> + typename = detail::enable_if_t< + IsSameAsBuffer() && IsValidTag() && IsPlaceH && + (IsGlobalBuf || IsConstantBuf || IsHostBuf)>> accessor(buffer &BufferRef, TagT, const property_list &PropertyList = {}) : accessor(BufferRef, PropertyList) {} @@ -980,9 +980,9 @@ class accessor : template () && - IsValidTag() && !IsPlaceH && - (IsGlobalBuf || IsConstantBuf)>> + typename = detail::enable_if_t< + IsSameAsBuffer() && IsValidTag() && !IsPlaceH && + (IsGlobalBuf || IsConstantBuf || IsHostBuf)>> accessor(buffer &BufferRef, handler &CommandGroupHandler, TagT, const property_list &PropertyList = {}) : accessor(BufferRef, CommandGroupHandler, PropertyList) {} @@ -1014,9 +1014,9 @@ class accessor : #endif template () && - (!IsPlaceH && - (IsGlobalBuf || IsConstantBuf))>> + typename = detail::enable_if_t< + IsSameAsBuffer() && + (!IsPlaceH && (IsGlobalBuf || IsConstantBuf || IsHostBuf))>> accessor(buffer &BufferRef, handler &CommandGroupHandler, range AccessRange, const property_list &PropertyList = {}) @@ -1027,9 +1027,9 @@ class accessor : template () && - IsValidTag() && !IsPlaceH && - (IsGlobalBuf || IsConstantBuf)>> + typename = detail::enable_if_t< + IsSameAsBuffer() && IsValidTag() && !IsPlaceH && + (IsGlobalBuf || IsConstantBuf || IsHostBuf)>> accessor(buffer &BufferRef, handler &CommandGroupHandler, range AccessRange, TagT, const property_list &PropertyList = {}) @@ -1078,9 +1078,9 @@ class accessor : #endif template () && - (!IsPlaceH && - (IsGlobalBuf || IsConstantBuf))>> + typename = detail::enable_if_t< + IsSameAsBuffer() && + (!IsPlaceH && (IsGlobalBuf || IsConstantBuf || IsHostBuf))>> accessor(buffer &BufferRef, handler &CommandGroupHandler, range AccessRange, id AccessOffset, const property_list &PropertyList = {}) @@ -1105,9 +1105,9 @@ class accessor : template () && - IsValidTag() && !IsPlaceH && - (IsGlobalBuf || IsConstantBuf)>> + typename = detail::enable_if_t< + IsSameAsBuffer() && IsValidTag() && !IsPlaceH && + (IsGlobalBuf || IsConstantBuf || IsHostBuf)>> accessor(buffer &BufferRef, handler &CommandGroupHandler, range AccessRange, id AccessOffset, TagT, const property_list &PropertyList = {}) @@ -1620,8 +1620,6 @@ class host_accessor // buffer | handler | range | id | | property_list // buffer | handler | range | id | mode_tag | property_list // -------+---------+-------+----+----------+-------------- - // host_accessor with handler argument will be added later - // to facilitate non-blocking accessor use case template , const property_list &PropertyList = {}) : host_accessor(BufferRef, PropertyList) {} +#endif + + template ()>> + host_accessor(buffer &BufferRef, + handler &CommandGroupHandler, + const property_list &PropertyList = {}) + : AccessorT(BufferRef, CommandGroupHandler, PropertyList) {} + +#if __cplusplus > 201402L + + template ()>> + host_accessor(buffer &BufferRef, + handler &CommandGroupHandler, mode_tag_t, + const property_list &PropertyList = {}) + : host_accessor(BufferRef, CommandGroupHandler, PropertyList) {} + #endif template ()>> + host_accessor(buffer &BufferRef, + handler &CommandGroupHandler, range AccessRange, + const property_list &PropertyList = {}) + : AccessorT(BufferRef, CommandGroupHandler, AccessRange, {}, + PropertyList) {} + +#if __cplusplus > 201402L + + template ()>> + host_accessor(buffer &BufferRef, + handler &CommandGroupHandler, range AccessRange, + mode_tag_t, const property_list &PropertyList = {}) + : host_accessor(BufferRef, CommandGroupHandler, AccessRange, {}, + PropertyList) {} + #endif template , const property_list &PropertyList = {}) : host_accessor(BufferRef, AccessRange, AccessOffset, PropertyList) {} +#endif + + template ()>> + host_accessor(buffer &BufferRef, + handler &CommandGroupHandler, range AccessRange, + id AccessOffset, + const property_list &PropertyList = {}) + : AccessorT(BufferRef, CommandGroupHandler, AccessRange, AccessOffset, + PropertyList) {} + +#if __cplusplus > 201402L + + template ()>> + host_accessor(buffer &BufferRef, + handler &CommandGroupHandler, range AccessRange, + id AccessOffset, mode_tag_t, + const property_list &PropertyList = {}) + : host_accessor(BufferRef, CommandGroupHandler, AccessRange, AccessOffset, + PropertyList) {} + #endif }; @@ -1712,6 +1770,13 @@ host_accessor(buffer, Type1, Type2, Type3, Type4) ->host_accessor()>; +template +host_accessor(buffer, Type1, Type2, Type3, Type4, + Type5) + ->host_accessor()>; + #endif } // namespace sycl diff --git a/sycl/include/CL/sycl/buffer.hpp b/sycl/include/CL/sycl/buffer.hpp old mode 100644 new mode 100755 index 98b4c217bac39..2299ac7c3d483 --- a/sycl/include/CL/sycl/buffer.hpp +++ b/sycl/include/CL/sycl/buffer.hpp @@ -292,6 +292,11 @@ class buffer { return host_accessor{*this, args...}; } + template + auto get_host_access(handler &commandGroupHandler, Ts... args) { + return host_accessor{*this, commandGroupHandler, args...}; + } + #endif template diff --git a/sycl/test/basic_tests/accessor/Inputs/host_task_accessor.cpp b/sycl/test/basic_tests/accessor/Inputs/host_task_accessor.cpp new file mode 100644 index 0000000000000..91728cf1831de --- /dev/null +++ b/sycl/test/basic_tests/accessor/Inputs/host_task_accessor.cpp @@ -0,0 +1,125 @@ +//==-------- host_task_accessor.cpp - SYCL accessor basic test -------------==// +// +// 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 +#include + +int main() { + // Non-placeholder accessors. + { + int data[9] = {1, 2, 3, 4, 5, 6, 7, 8, 9}; + + sycl::buffer buf_data(data, sycl::range<1>(9), + {cl::sycl::property::buffer::use_host_ptr()}); + + sycl::queue Queue; + + Queue.submit([&](sycl::handler &cgh) { + +#if defined(accessor_new_api_test) + sycl::host_accessor acc_1(buf_data, cgh); + sycl::host_accessor acc_2(buf_data, cgh, sycl::range<1>(8)); + sycl::host_accessor acc_3(buf_data, cgh, sycl::range<1>(8), sycl::id<1>(1)); + sycl::host_accessor acc_4(buf_data, cgh, sycl::read_only); + sycl::host_accessor acc_5(buf_data, cgh, sycl::range<1>(8), sycl::read_only); + sycl::host_accessor acc_6(buf_data, cgh, sycl::range<1>(8), sycl::id<1>(1), + sycl::read_only); + sycl::host_accessor acc_7(buf_data, cgh, sycl::write_only); + sycl::host_accessor acc_8(buf_data, cgh, sycl::range<1>(8), sycl::write_only); + sycl::host_accessor acc_9(buf_data, cgh, sycl::range<1>(8), sycl::id<1>(1), + sycl::write_only); +#elif defined(buffer_new_api_test) + auto acc_1 = buf_data.get_host_access(cgh); + auto acc_2 = buf_data.get_host_access(cgh, sycl::range<1>(8)); + auto acc_3 = buf_data.get_host_access(cgh, sycl::range<1>(8), sycl::id<1>(1)); + auto acc_4 = buf_data.get_host_access(cgh, sycl::read_only); + auto acc_5 = buf_data.get_host_access(cgh, sycl::range<1>(8), sycl::read_only); + auto acc_6 = buf_data.get_host_access(cgh, sycl::range<1>(8), sycl::id<1>(1), + sycl::read_only); + auto acc_7 = buf_data.get_host_access(cgh, sycl::write_only); + auto acc_8 = buf_data.get_host_access(cgh, sycl::range<1>(8), sycl::write_only); + auto acc_9 = buf_data.get_host_access(cgh, sycl::range<1>(8), sycl::id<1>(1), + sycl::write_only); +#endif + + cgh.codeplay_host_task( + [=]() { + acc_7[6] = acc_1[0]; + acc_8[7] = acc_2[1]; + acc_9[7] = acc_3[1]; + acc_1[0] = acc_4[3]; + acc_2[1] = acc_5[4]; + acc_3[1] = acc_6[4]; + }); + }); + Queue.wait(); + +#if defined(accessor_new_api_test) + sycl::host_accessor host_acc(buf_data, sycl::read_only); +#elif defined(buffer_new_api_test) + auto host_acc = buf_data.get_host_access(sycl::read_only); +#endif + assert(host_acc[0] == 4 && host_acc[1] == 5 && host_acc[2] == 6); + assert(host_acc[3] == 4 && host_acc[4] == 5 && host_acc[5] == 6); + assert(host_acc[6] == 1 && host_acc[7] == 2 && host_acc[8] == 3); + } + + // noinit accessors. + { + int data[9] = {1, 2, 3, 4, 5, 6, 7, 8, 9}; + + sycl::buffer buf_data(data, sycl::range<1>(9), + {cl::sycl::property::buffer::use_host_ptr()}); + + sycl::queue Queue; + + Queue.submit([&](sycl::handler &cgh) { + +#if defined(accessor_new_api_test) + sycl::host_accessor acc_1(buf_data, cgh, sycl::noinit); + sycl::host_accessor acc_2(buf_data, cgh, sycl::range<1>(8), sycl::noinit); + sycl::host_accessor acc_3(buf_data, cgh, sycl::range<1>(8), sycl::id<1>(1), + sycl::noinit); + sycl::host_accessor acc_7(buf_data, cgh, sycl::write_only, sycl::noinit); + sycl::host_accessor acc_8(buf_data, cgh, sycl::range<1>(8), sycl::write_only, + sycl::noinit); + sycl::host_accessor acc_9(buf_data, cgh, sycl::range<1>(8), sycl::id<1>(1), + sycl::write_only, sycl::noinit); +#elif defined(buffer_new_api_test) + auto acc_1 = buf_data.get_host_access(cgh, sycl::noinit); + auto acc_2 = buf_data.get_host_access(cgh, sycl::range<1>(8), sycl::noinit); + auto acc_3 = buf_data.get_host_access(cgh, sycl::range<1>(8), sycl::id<1>(1), + sycl::noinit); + auto acc_7 = buf_data.get_host_access(cgh, sycl::write_only, sycl::noinit); + auto acc_8 = buf_data.get_host_access(cgh, sycl::range<1>(8), sycl::write_only, + sycl::noinit); + auto acc_9 = buf_data.get_host_access(cgh, sycl::range<1>(8), sycl::id<1>(1), + sycl::write_only, sycl::noinit); +#endif + + cgh.codeplay_host_task( + [=]() { + acc_7[6] = acc_1[0]; + acc_8[7] = acc_2[1]; + acc_9[7] = acc_3[1]; + acc_1[0] = 4; + acc_2[1] = 5; + acc_3[1] = 6; + }); + }); + Queue.wait(); + +#if defined(accessor_new_api_test) + sycl::host_accessor host_acc(buf_data, sycl::read_only); +#elif defined(buffer_new_api_test) + auto host_acc = buf_data.get_host_access(sycl::read_only); +#endif + assert(host_acc[0] == 4 && host_acc[1] == 5 && host_acc[2] == 6); + assert(host_acc[3] == 4 && host_acc[4] == 5 && host_acc[5] == 6); + assert(host_acc[6] == 1 && host_acc[7] == 2 && host_acc[8] == 3); + } +} diff --git a/sycl/test/basic_tests/accessor/get_host_task_access_deduction.cpp b/sycl/test/basic_tests/accessor/get_host_task_access_deduction.cpp new file mode 100755 index 0000000000000..97aee77f9c497 --- /dev/null +++ b/sycl/test/basic_tests/accessor/get_host_task_access_deduction.cpp @@ -0,0 +1,5 @@ +// RUN: %clangxx -fsycl -fsycl-targets=%sycl_triple -Dbuffer_new_api_test -std=c++17 %S/Inputs/host_task_accessor.cpp -o %t.out +// RUN: env SYCL_DEVICE_TYPE=HOST %t.out +// RUN: %CPU_RUN_PLACEHOLDER %t.out +// RUN: %GPU_RUN_PLACEHOLDER %t.out +// RUN: %ACC_RUN_PLACEHOLDER %t.out diff --git a/sycl/test/basic_tests/accessor/host_task_accessor_deduction.cpp b/sycl/test/basic_tests/accessor/host_task_accessor_deduction.cpp new file mode 100755 index 0000000000000..760bee675f4cf --- /dev/null +++ b/sycl/test/basic_tests/accessor/host_task_accessor_deduction.cpp @@ -0,0 +1,5 @@ +// RUN: %clangxx -fsycl -fsycl-targets=%sycl_triple -Daccessor_new_api_test -std=c++17 %S/Inputs/host_task_accessor.cpp -o %t.out +// RUN: env SYCL_DEVICE_TYPE=HOST %t.out +// RUN: %CPU_RUN_PLACEHOLDER %t.out +// RUN: %GPU_RUN_PLACEHOLDER %t.out +// RUN: %ACC_RUN_PLACEHOLDER %t.out