diff --git a/help_function/help_function.xml b/help_function/help_function.xml index c88c7789c..373ac4ce7 100644 --- a/help_function/help_function.xml +++ b/help_function/help_function.xml @@ -4,8 +4,9 @@ Test the helper function - - + + + diff --git a/help_function/src/assign_device_vector.cpp b/help_function/src/assign_device_vector.cpp index 9deb9d577..8148782fa 100644 --- a/help_function/src/assign_device_vector.cpp +++ b/help_function/src/assign_device_vector.cpp @@ -1,9 +1,130 @@ +// ====------ assign_device_vector.cpp---------- -*- C++ -* ----===//// +// +// 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 #include #include #include +template +bool verify(Vector &D, int N, int V) { + if (D.size() != N) + { + std::cout<<"size mismatch"<> D3(10, alloc_no_copy_prop2); + //since there is no copy propagation, we only destroy the excess, not all 10 elements, and we copy the N elements + D3 = D2; + if (!verify(D3, N, V)) { + std::cout<<"Failed assign of AllocWithNoCopyProp"< alloc_copy_prop1(dpct::get_default_queue()); + AllocWithCopyPropagation alloc_copy_prop2(dpct::get_default_queue()); + + //Should allocate 5 ints from D3 in the new allocation space + dpct::device_vector> D4(std::move(D3), alloc_copy_prop1); + if (!verify(D4, N, V)) { + std::cout<<"Failed move of AllocWithNoCopyProp to AllocWithCopyProp"<> D5(10, alloc_copy_prop2); + //since there is copy propagation, we destroy all 10 elements, and use the propagated allocator to construct + // 5 new elements + D5 = D4; + if (!verify(D5, N, V)) { + std::cout<<"Failed copy assign of AllocWithCopyProp"< alloc_default_on_copy(dpct::get_default_queue(), 42); + dpct::device_vector> D6(N, V, alloc_default_on_copy); + if (D6.get_allocator().index != 42) { + std::cout<<"index not set correctly for AllocWithDefaultOnCopyConstruction " + <> D7(D6); + if (!verify(D7, N, V)) { + std::cout<<"Failed copy constructor of AllocWithDefaultOnCopyConstruction"< alloc_copy_on_copy(dpct::get_default_queue(), 33); + dpct::device_vector> D8(N, V, alloc_copy_on_copy); + if (!verify(D8, N, V)) { + std::cout<<"Failed creation of AllocWithCopyOnCopyConstruction"<> D9(D8); + if (!verify(D9, N, V)) { + std::cout<<"Failed copy constructor of AllocWithCopyOnCopyConstruction"< #include #include #include #include -bool verify(dpct::device_vector &D, int N, int V) { +template +bool verify(Vector &D, int N, int V) { if (D.size() != N) + { + std::cout<<"size mismatch"<> D4(std::move(D3)); + if (!verify(D4, N, V)) { + std::cout<<"Failed move of AllocWithNoMoveProp"<> D5; + D5 = std::move(D4); + if (!verify(D5, N, V)) { + std::cout<<"Failed move assign of AllocWithNoMoveProp"< alloc_move_prop(dpct::get_default_queue()); + + dpct::device_vector> D6(std::move(D5), alloc_move_prop); + if (!verify(D6, N, V)) { + std::cout<<"Failed move of AllocWithNoMoveProp to AllocWithMoveProp"<> D7; + D7 = std::move(D6); + if (!verify(D7, N, V)) { + std::cout<<"Failed move assign of AllocWithMoveProp"<> D8(std::move(D7)); + if (!verify(D8, N, V)) { + std::cout<<"Failed move of AllocWithMoveProp"< +class my_allocator_with_custom_construct : public sycl::usm_allocator { + public: + my_allocator_with_custom_construct(const sycl::context &Ctxt, const sycl::device &Dev, + const sycl::property_list &PropList = {}) + : sycl::usm_allocator(Ctxt, Dev, PropList) {} + my_allocator_with_custom_construct(const sycl::queue &Q, const sycl::property_list &PropList = {}) + : sycl::usm_allocator(Q, PropList) + {} + + static void construct(T *p) { ::new((void*)p) T(6); } + template + static void construct(T *p, _Arg arg) { ::new((void*)p) T(arg + 3); } + static void destroy(T *p) { p->~T(); } + +}; + + int main() { // used to detect failures @@ -90,6 +110,10 @@ int main() { v4.insert(v4.begin(), 2, *(v2.begin()) - 111); v4.insert(v4.begin(), v2.begin(), v2.begin() + 2); #endif + //insert host side data into the vector + std::vector host_v(2, 79); + v4.insert(v4.begin()+3, host_v.begin(), host_v.end()); + #ifdef _VERBOSE std::cout << "v4.size() = " << v4.size() << std::endl; std::cout << "v4: "; @@ -123,7 +147,7 @@ int main() { //failed_tests += ASSERT_EQUAL("v6.back() = 2", v6.back(), 2); #endif v6.pop_back(); - v6.reserve(20); + v6.reserve(24); #ifdef _VERBOSE if (!v6.empty() && v6.front() == *v6.begin()) { std::cout << "v6.size() = " << v6.size() << ", v6.max_size() = " << @@ -133,19 +157,19 @@ int main() { std::cout << "v6[0] = " << v6[0] << std::endl; // expected: 5 } #else - failed_tests += ASSERT_EQUAL("v6.size() = 12", v6.size(), 12); + failed_tests += ASSERT_EQUAL("v6.size() = 14", v6.size(), 14); failed_tests += ASSERT_EQUAL("v6.max_size()", v6.max_size(), 4611686018427387903); - failed_tests += ASSERT_EQUAL("v6.capacity() = 20", v6.capacity(), 20); + failed_tests += ASSERT_EQUAL("v6.capacity() = 24", v6.capacity(), 24); v6.shrink_to_fit(); - failed_tests += ASSERT_EQUAL("v6.capacity() = 12", v6.capacity(), 12); + failed_tests += ASSERT_EQUAL("v6.capacity() = 14", v6.capacity(), 14); failed_tests += ASSERT_EQUAL("v6[0] = 0", v6[0], 0); #endif v6.resize(20, 99); auto resize_policy = oneapi::dpl::execution::make_device_policy(dpct::get_default_queue()); - auto sum = std::reduce(resize_policy, v6.begin()+12, v6.end(), 0); - failed_tests += ASSERT_EQUAL("sum = 792", sum, 792); + auto sum = std::reduce(resize_policy, v6.begin()+14, v6.end(), 0); + failed_tests += ASSERT_EQUAL("sum = 594", sum, 594); - v6.erase(v6.cbegin() + 10, v6.cend()); + v6.erase(v6.cbegin() + 14, v6.cend()); #ifdef _VERBOSE for (std::size_t i = 0; i < v6.size(); ++i) { std::cout << v6[i] << " "; // expected: 5 4 3 2 1 -111 -111 -111 1 0 @@ -157,17 +181,40 @@ int main() { num_failing += ASSERT_ARRAY_EQUAL(test_name, v6[0], 0); num_failing += ASSERT_ARRAY_EQUAL(test_name, v6[1], 1); num_failing += ASSERT_ARRAY_EQUAL(test_name, v6[2], -111); - num_failing += ASSERT_ARRAY_EQUAL(test_name, v6[3], -111); - num_failing += ASSERT_ARRAY_EQUAL(test_name, v6[4], -111); - num_failing += ASSERT_ARRAY_EQUAL(test_name, v6[5], 1); - num_failing += ASSERT_ARRAY_EQUAL(test_name, v6[6], 2); - num_failing += ASSERT_ARRAY_EQUAL(test_name, v6[7], 3); - num_failing += ASSERT_ARRAY_EQUAL(test_name, v6[8], 4); - num_failing += ASSERT_ARRAY_EQUAL(test_name, v6[9], 5); + num_failing += ASSERT_ARRAY_EQUAL(test_name, v6[3], 79); + num_failing += ASSERT_ARRAY_EQUAL(test_name, v6[4], 79); + num_failing += ASSERT_ARRAY_EQUAL(test_name, v6[5], -111); + num_failing += ASSERT_ARRAY_EQUAL(test_name, v6[6], -111); + num_failing += ASSERT_ARRAY_EQUAL(test_name, v6[7], 1); + num_failing += ASSERT_ARRAY_EQUAL(test_name, v6[8], 2); + num_failing += ASSERT_ARRAY_EQUAL(test_name, v6[9], 3); + num_failing += ASSERT_ARRAY_EQUAL(test_name, v6[10], 4); + num_failing += ASSERT_ARRAY_EQUAL(test_name, v6[11], 5); failed_tests += test_passed(num_failing, test_name); num_failing = 0; #endif + Vector v7(v5.begin()+3, v5.end()-2); + dpct::get_default_queue().wait(); +#ifdef _VERBOSE + std::cout << std::endl << "v7: "; + for (std::size_t i = 0; i < v7.size(); ++i) { + std::cout << v7[i] << " "; // expected: 79 79 -111 -111 -111 1 2 3 + } + std::cout << std::endl; +#else + test_name = "v7 = modified v5"; + + num_failing += ASSERT_ARRAY_EQUAL(test_name, v7[0], 79); + num_failing += ASSERT_ARRAY_EQUAL(test_name, v7[1], 79); + num_failing += ASSERT_ARRAY_EQUAL(test_name, v7[2], -111); + num_failing += ASSERT_ARRAY_EQUAL(test_name, v7[3], -111); + num_failing += ASSERT_ARRAY_EQUAL(test_name, v7[4], 1); + num_failing += ASSERT_ARRAY_EQUAL(test_name, v7[5], 2); + num_failing += ASSERT_ARRAY_EQUAL(test_name, v7[6], 3); +#endif + + { // Simple test of DPCT vector with PSTL algorithm // create buffer std::vector src(8); @@ -209,6 +256,71 @@ int main() { failed_tests += test_passed(num_failing, test_name); } + num_failing = 0; + test_name = "custom_allocator default construction"; + { // test with custom allocator which constructs default constructor of 6 + dpct::device_vector> default_construct(5); + default_construct[4] += 2; + num_failing += ASSERT_ARRAY_EQUAL(test_name, default_construct[0], 6); + num_failing += ASSERT_ARRAY_EQUAL(test_name, default_construct[1], 6); + num_failing += ASSERT_ARRAY_EQUAL(test_name, default_construct[2], 6); + num_failing += ASSERT_ARRAY_EQUAL(test_name, default_construct[3], 6); + num_failing += ASSERT_ARRAY_EQUAL(test_name, default_construct[4], 8); + failed_tests += test_passed(num_failing, test_name); + } + num_failing = 0; + test_name = "custom_allocator construction from input"; + { // test with custom allocator which whos default constructor adds of 3 when constructing from a value or iterator + dpct::device_vector> construct_from_value(5, 2); + construct_from_value[4] += 2; + num_failing += ASSERT_ARRAY_EQUAL(test_name, construct_from_value[0], 5); + num_failing += ASSERT_ARRAY_EQUAL(test_name, construct_from_value[1], 5); + num_failing += ASSERT_ARRAY_EQUAL(test_name, construct_from_value[2], 5); + num_failing += ASSERT_ARRAY_EQUAL(test_name, construct_from_value[3], 5); + num_failing += ASSERT_ARRAY_EQUAL(test_name, construct_from_value[4], 7); + + std::vector src(8); + src[0] = -1; src[1] = 2; src[2] = -3; src[3] = 4; src[4] = -5; src[5] = 6; src[6] = -7; src[7] = 8; + + dpct::device_vector> construct_from_iter(src.begin()+2, src.begin()+7); + construct_from_iter[4] += 2; + num_failing += ASSERT_ARRAY_EQUAL(test_name, construct_from_iter[0], 0); + num_failing += ASSERT_ARRAY_EQUAL(test_name, construct_from_iter[1], 7); + num_failing += ASSERT_ARRAY_EQUAL(test_name, construct_from_iter[2], -2); + num_failing += ASSERT_ARRAY_EQUAL(test_name, construct_from_iter[3], 9); + num_failing += ASSERT_ARRAY_EQUAL(test_name, construct_from_iter[4], -2); + failed_tests += test_passed(num_failing, test_name); + } + + num_failing = 0; + test_name = "inserting host iterators"; + { + std::vector src(8, 99); + std::vector src2(2, 2); + std::vector src3(3, 33); + + dpct::device_vector dst = src; + dst.insert(dst.end(), src2.begin(), src2.end()); + + dst.insert(dst.begin() + 3, src3.begin(), src3.end()); + + num_failing += ASSERT_ARRAY_EQUAL(test_name, dst.size(), 13); + num_failing += ASSERT_ARRAY_EQUAL(test_name, dst[0], 99); + num_failing += ASSERT_ARRAY_EQUAL(test_name, dst[1], 99); + num_failing += ASSERT_ARRAY_EQUAL(test_name, dst[2], 99); + num_failing += ASSERT_ARRAY_EQUAL(test_name, dst[3], 33); + num_failing += ASSERT_ARRAY_EQUAL(test_name, dst[4], 33); + num_failing += ASSERT_ARRAY_EQUAL(test_name, dst[5], 33); + num_failing += ASSERT_ARRAY_EQUAL(test_name, dst[6], 99); + num_failing += ASSERT_ARRAY_EQUAL(test_name, dst[7], 99); + num_failing += ASSERT_ARRAY_EQUAL(test_name, dst[8], 99); + num_failing += ASSERT_ARRAY_EQUAL(test_name, dst[9], 99); + num_failing += ASSERT_ARRAY_EQUAL(test_name, dst[10], 99); + num_failing += ASSERT_ARRAY_EQUAL(test_name, dst[11], 2); + num_failing += ASSERT_ARRAY_EQUAL(test_name, dst[12], 2); + } + + std::cout << std::endl << failed_tests << " failing test(s) detected." << std::endl; if (failed_tests == 0) { diff --git a/help_function/src/swap_device_vector.cpp b/help_function/src/swap_device_vector.cpp new file mode 100644 index 000000000..f0ebaff7a --- /dev/null +++ b/help_function/src/swap_device_vector.cpp @@ -0,0 +1,193 @@ +// ====------ swap_device_vector.cpp---------- -*- C++ -* ----===//// +// +// 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 +#include +#include +#include + +template +bool +verify(Vector& D, int N, int V) +{ + if (D.size() != N) + { + std::cout << "size mismatch" << std::endl; + return false; + } + for (int i = 0; i < N; ++i) + if (D[i] != V) + { + std::cout << "value mismatch " << D[i] << " != " << V << std::endl; + return false; + } + return true; +} + +//adding these global variables to track construction and destruction +// using custom allocators with different settings. +static int num_constructed_prop = 0; +static int num_destroyed_prop = 0; +static int num_constructed_no_prop = 0; +static int num_destroyed_no_prop = 0; + +template +class AllocWithNoSwapPropagation : public sycl::usm_allocator +{ + public: + AllocWithNoSwapPropagation(const sycl::context& Ctxt, const sycl::device& Dev, + const sycl::property_list& PropList = {}) + : sycl::usm_allocator(Ctxt, Dev, PropList) + { + } + AllocWithNoSwapPropagation(const sycl::queue& Q, const sycl::property_list& PropList = {}) + : sycl::usm_allocator(Q, PropList) + { + } + + AllocWithNoSwapPropagation(const AllocWithNoSwapPropagation& other) + : sycl::usm_allocator(other) + { + } + + typedef ::std::false_type propagate_on_container_swap; + static void + construct(T* p) + { + ::new ((void*)p) T(); + num_constructed_no_prop++; + } + template + static void + construct(T* p, _Arg arg) + { + ::new ((void*)p) T(arg); + num_constructed_no_prop++; + } + static void + destroy(T* p) + { + p->~T(); + num_destroyed_no_prop++; + } +}; + +template +class AllocWithSwapPropagation : public sycl::usm_allocator +{ + public: + AllocWithSwapPropagation(const sycl::context& Ctxt, const sycl::device& Dev, + const sycl::property_list& PropList = {}) + : sycl::usm_allocator(Ctxt, Dev, PropList) + { + } + AllocWithSwapPropagation(const sycl::queue& Q, const sycl::property_list& PropList = {}) + : sycl::usm_allocator(Q, PropList) + { + } + + AllocWithSwapPropagation(const AllocWithSwapPropagation& other) + : sycl::usm_allocator(other) + { + } + + typedef ::std::true_type propagate_on_container_swap; + static void + construct(T* p) + { + ::new ((void*)p) T(); + num_constructed_prop++; + } + template + static void + construct(T* p, _Arg arg) + { + ::new ((void*)p) T(arg); + num_constructed_prop++; + } + static void + destroy(T* p) + { + p->~T(); + num_destroyed_prop++; + } +}; + +int +main(void) +{ + constexpr int N1 = 3; + constexpr int V1 = 99; + constexpr int N2 = 7; + constexpr int V2 = 98; + dpct::device_vector D1(N1, V1); + + // check appropriate effect of Allocator::propagate_on_container_swap + + AllocWithNoSwapPropagation alloc_no_swap_prop1(dpct::get_default_queue()); + AllocWithNoSwapPropagation alloc_no_swap_prop2(dpct::get_default_queue()); + + //Should allocate and construct N1 int{} + dpct::device_vector> D2(std::move(D1), alloc_no_swap_prop1); + if (!verify(D2, N1, V1)) + { + std::cout << "Failed move of default allocator to AllocWithNoSwapProp" << std::endl; + return 1; + } + + //Should allocate and construct N2 int{} + dpct::device_vector> D3(N2, V2, alloc_no_swap_prop2); + //since there is no swap propagation, we only destroy the excess + D3.swap(D2); + if (!verify(D3, N1, V1) || !verify(D2, N2, V2)) + { + std::cout << "Failed swap of AllocWithNoSwapProp" << std::endl; + return 1; + } + + if (num_constructed_no_prop != 14 && num_destroyed_no_prop != 4) + { + std::cout << "Allocator without swap propagation is swaping incorrectly: [14, 4] != [" + << num_constructed_no_prop << ", "<< num_destroyed_no_prop << "]" << std::endl; + return 1; + } + + + dpct::device_vector D4(N1, V1); + + AllocWithSwapPropagation alloc_swap_prop1(dpct::get_default_queue()); + AllocWithSwapPropagation alloc_swap_prop2(dpct::get_default_queue()); + + //Should allocate and construct N1 int{} + dpct::device_vector> D5(std::move(D4), alloc_swap_prop1); + if (!verify(D5, N1, V1)) + { + std::cout << "Failed move of default allocator to AllocWithSwapProp" << std::endl; + return 1; + } + + //Should allocate and construct N2 int{} + dpct::device_vector> D6(N2, V2, alloc_swap_prop2); + //since there is no swap propagation, we only destroy the excess + D6.swap(D5); + if (!verify(D6, N1, V1) || !verify(D5, N2, V2)) + { + std::cout << "Failed swap of AllocWithSwapProp" << std::endl; + return 1; + } + + if (num_constructed_prop != 10 && num_destroyed_prop != 0) + { + std::cout << "Allocator with swap propagation is swaping incorrectly: [10, 0] != [" + << num_constructed_prop << ", "<< num_destroyed_prop << "]" << std::endl; + return 1; + } + return 0; +} \ No newline at end of file