diff --git a/CONTRIBUTING.md b/CONTRIBUTING.md index ad49b73a314..0174e52e313 100644 --- a/CONTRIBUTING.md +++ b/CONTRIBUTING.md @@ -2,12 +2,12 @@ We welcome community contributions to oneAPI DPC++ Library (oneDPL). You can: -- Submit your changes directly with a [pull request](https://github.com/oneapi-src/oneDPL/pulls). -- Log a bug or feedback with an [issue](https://github.com/oneapi-src/oneDPL/issues). +- Submit your changes directly with a [pull request](https://github.com/uxlfoundation/oneDPL/pulls). +- Log a bug or feedback with an [issue](https://github.com/uxlfoundation/oneDPL/issues). # License -oneDPL is licensed under the terms in [LICENSE](https://github.com/oneapi-src/oneDPL/blob/main/LICENSE.txt). +oneDPL is licensed under the terms in [LICENSE](https://github.com/uxlfoundation/oneDPL/blob/main/LICENSE.txt). By contributing to the project, you agree to the license and copyright terms therein and release your contribution under these terms. @@ -23,7 +23,7 @@ your change directly to the repository: [validation testing](#validation-testing). - Submit a [pull request](https://docs.github.com/en/free-pro-team@latest/github/collaborating-with-issues-and-pull-requests/creating-a-pull-request) into the - main branch. You may add a description of your contribution into [CREDITS.txt](https://github.com/oneapi-src/oneDPL/blob/main/CREDITS.txt). + main branch. You may add a description of your contribution into [CREDITS.txt](https://github.com/uxlfoundation/oneDPL/blob/main/CREDITS.txt). - Contributors that would like to open branches directly in the oneDPL repo instead of working via fork may request write access to the repository by contacting project maintainers on [UXL Foundation Slack](https://slack-invite.uxlfoundation.org/) using the @@ -31,7 +31,7 @@ your change directly to the repository: # Coding Conventions -Running clang-format is required, except in the [test folder](https://github.com/oneapi-src/oneDPL/tree/main/test). +Running clang-format is required, except in the [test folder](https://github.com/uxlfoundation/oneDPL/tree/main/test). # Validation Testing @@ -66,6 +66,6 @@ ctest --output-on-failure --timeout 1200 -R ^reduce.pass$ # Add -R testname (e.g Before submitting a PR to the oneDPL repository please run the tests that exercise the code you have updated. If you need help identifying those tests please check with the maintainers on [UXL Foundation Slack](https://slack-invite.uxlfoundation.org/) using the [#onedpl](https://uxlfoundation.slack.com/channels/onedpl) channel -or ask a question through [GitHub issues](https://github.com/oneapi-src/oneDPL/issues). +or ask a question through [GitHub issues](https://github.com/uxlfoundation/oneDPL/issues). -For more details on configurations available for oneDPL testing see the [CMake README](https://github.com/oneapi-src/oneDPL/blob/main/cmake/README.md). +For more details on configurations available for oneDPL testing see the [CMake README](https://github.com/uxlfoundation/oneDPL/blob/main/cmake/README.md). diff --git a/README.md b/README.md index 056ffd4e261..2b9bd12e8f0 100644 --- a/README.md +++ b/README.md @@ -1,7 +1,7 @@ # oneAPI DPC++ Library (oneDPL) -[![Apache License Version 2.0](https://img.shields.io/badge/license-Apache_2.0-green.svg)](LICENSE.txt) [![oneDPL CI](https://github.com/oneapi-src/oneDPL/actions/workflows/ci-testing.yml/badge.svg)](https://github.com/oneapi-src/oneDPL/actions/workflows/ci-testing.yml?query=branch%3Amaster) -[![Join the community on GitHub Discussions](https://badgen.net/badge/join%20the%20discussion/on%20github/blue?icon=github)](https://github.com/oneapi-src/oneDPL/discussions) -[![OpenSSF Scorecard](https://api.securityscorecards.dev/projects/github.com/oneapi-src/oneDPL/badge)](https://securityscorecards.dev/viewer/?uri=github.com/oneapi-src/oneDPL) +[![Apache License Version 2.0](https://img.shields.io/badge/license-Apache_2.0-green.svg)](LICENSE.txt) [![oneDPL CI](https://github.com/uxlfoundation/oneDPL/actions/workflows/ci-testing.yml/badge.svg)](https://github.com/uxlfoundation/oneDPL/actions/workflows/ci-testing.yml?query=branch%3Amaster) +[![Join the community on GitHub Discussions](https://badgen.net/badge/join%20the%20discussion/on%20github/blue?icon=github)](https://github.com/uxlfoundation/oneDPL/discussions) +[![OpenSSF Scorecard](https://api.securityscorecards.dev/projects/github.com/uxlfoundation/oneDPL/badge)](https://securityscorecards.dev/viewer/?uri=github.com/uxlfoundation/oneDPL) oneDPL is part of the [UXL Foundation] and is an implementation of the [oneAPI specification] for the oneDPL component. @@ -19,10 +19,10 @@ Install the Intel® oneAPI Base Toolkit (Base Kit) to use oneDPL. Refer to the s for more information. ## Release Information -Visit the latest [Release Notes](https://github.com/oneapi-src/oneDPL/blob/main/documentation/release_notes.rst). +Visit the latest [Release Notes](https://github.com/uxlfoundation/oneDPL/blob/main/documentation/release_notes.rst). ## License -oneDPL is licensed under [Apache License Version 2.0 with LLVM exceptions](https://github.com/oneapi-src/oneDPL/blob/main/LICENSE.txt). +oneDPL is licensed under [Apache License Version 2.0 with LLVM exceptions](https://github.com/uxlfoundation/oneDPL/blob/main/LICENSE.txt). Refer to the [LICENSE](licensing/LICENSE.txt) file for the full license text and copyright notice. ## Security @@ -31,17 +31,17 @@ for information on how to report a potential security issue or vulnerability. You can also view the [Security Policy](SECURITY.md). ## Contributing -See [CONTRIBUTING.md](https://github.com/oneapi-src/oneDPL/blob/main/CONTRIBUTING.md) for details. +See [CONTRIBUTING.md](https://github.com/uxlfoundation/oneDPL/blob/main/CONTRIBUTING.md) for details. ## Documentation -See the full documentation set for [oneDPL](https://oneapi-src.github.io/oneDPL). +See the full documentation set for [oneDPL](https://uxlfoundation.github.io/oneDPL). ## Samples You can find oneDPL samples at the [oneDPL Samples](https://github.com/oneapi-src/oneAPI-samples/tree/master/Libraries/oneDPL) page. ## Support and Contribution -Please report issues and suggestions via [GitHub issues](https://github.com/oneapi-src/oneDPL/issues). +Please report issues and suggestions via [GitHub issues](https://github.com/uxlfoundation/oneDPL/issues). You can also contact oneDPL developers via [UXL Foundation Slack](https://slack-invite.uxlfoundation.org/) using the [#onedpl](https://uxlfoundation.slack.com/channels/onedpl) channel. diff --git a/SECURITY.md b/SECURITY.md index 847606a51d2..75d54e41198 100644 --- a/SECURITY.md +++ b/SECURITY.md @@ -61,6 +61,6 @@ If you have any suggestions on how this Policy could be improved, submit an issue or a pull request to this repository. **Do not** report potential vulnerabilities or security flaws via a pull request. -[1]: https://github.com/oneapi-src/oneDPL/releases/latest -[2]: https://github.com/oneapi-src/oneDPL/security/advisories/new -[3]: https://github.com/oneapi-src/oneDPL/security/advisories +[1]: https://github.com/uxlfoundation/oneDPL/releases/latest +[2]: https://github.com/uxlfoundation/oneDPL/security/advisories/new +[3]: https://github.com/uxlfoundation/oneDPL/security/advisories diff --git a/documentation/CHANGES.rst b/documentation/CHANGES.rst index dca9c8d0c0f..b7a175b3889 100644 --- a/documentation/CHANGES.rst +++ b/documentation/CHANGES.rst @@ -577,6 +577,6 @@ Known Issues and Limitations .. [#fnote1] The sorting algorithms in oneDPL use Radix sort for arithmetic data types compared with ``std::less`` or ``std::greater``, otherwise Merge sort. .. _`Intel® oneAPI Threading Building Blocks (oneTBB) Release Notes`: https://www.intel.com/content/www/us/en/developer/articles/release-notes/intel-oneapi-threading-building-blocks-release-notes.html -.. _`oneDPL Guide`: https://oneapi-src.github.io/oneDPL/index.html -.. _`Tested Standard C++ API`: https://oneapi-src.github.io/oneDPL/api_for_dpcpp_kernels/tested_standard_cpp_api.html#tested-standard-c-api-reference -.. _`Macros`: https://oneapi-src.github.io/oneDPL/macros.html +.. _`oneDPL Guide`: https://uxlfoundation.github.io/oneDPL/index.html +.. _`Tested Standard C++ API`: https://uxlfoundation.github.io/oneDPL/api_for_dpcpp_kernels/tested_standard_cpp_api.html#tested-standard-c-api-reference +.. _`Macros`: https://uxlfoundation.github.io/oneDPL/macros.html diff --git a/documentation/library_guide/_templates/layout.html b/documentation/library_guide/_templates/layout.html index 567eddaf3f1..4d8d4e23407 100644 --- a/documentation/library_guide/_templates/layout.html +++ b/documentation/library_guide/_templates/layout.html @@ -7,7 +7,7 @@ var wapLocalCode = 'us-en'; // Dynamically set per localized site, see mapping table for values var wapSection = "oneapi-dpl"; // WAP team will give you a unique section for your site // Load TMS - if(document.location.href.includes("oneapi-src.github.io")){ + if(document.location.href.includes("uxlfoundation.github.io")){ (function () { var url = 'https://www.intel.com/content/dam/www/global/wap/tms-loader.js'; // WAP file URL var po = document.createElement('script'); po.type = 'text/javascript'; po.async = true; po.src = url; @@ -50,4 +50,4 @@ {% include "searchbox.html" %} - {% endblock %} \ No newline at end of file + {% endblock %} diff --git a/documentation/library_guide/conf.py b/documentation/library_guide/conf.py index f24b48a7028..2450a451f28 100644 --- a/documentation/library_guide/conf.py +++ b/documentation/library_guide/conf.py @@ -113,7 +113,7 @@ #'collapse_navigation': False, # Collapse navigation (False makes it tree-like) #'navigation_depth': 4 # Depth of the headers shown in the navigation bar #'display_version': True, # Display the docs version - 'repository_url': 'https://github.com/oneapi-src/oneDPL', + 'repository_url': 'https://github.com/uxlfoundation/oneDPL', 'path_to_docs': 'documentation/library_guide', 'use_issues_button': True, 'use_edit_page_button': True, diff --git a/documentation/library_guide/index.rst b/documentation/library_guide/index.rst index f460357ad1f..026fa570f72 100644 --- a/documentation/library_guide/index.rst +++ b/documentation/library_guide/index.rst @@ -6,7 +6,7 @@ provide high-productivity APIs to developers, which can minimize SYCL* programmi efforts across devices for high performance parallel applications. For general information, refer to the `oneDPL GitHub* repository -`_. +`_. .. toctree:: :maxdepth: 2 diff --git a/documentation/library_guide/introduction/onedpl_gsg.rst b/documentation/library_guide/introduction/onedpl_gsg.rst index 761b7d677b2..d15dbf755a8 100644 --- a/documentation/library_guide/introduction/onedpl_gsg.rst +++ b/documentation/library_guide/introduction/onedpl_gsg.rst @@ -12,7 +12,7 @@ programming efforts across devices for high performance parallel applications. * Macros -For general information about |onedpl_short|, visit the `oneDPL GitHub* repository `_, +For general information about |onedpl_short|, visit the `oneDPL GitHub* repository `_, or visit the |onedpl_library_guide|_ and the `Intel® oneAPI DPC++ Library main page `_. @@ -220,4 +220,4 @@ Find More - Add oneAPI components to a Yocto project build using the meta-intel layers. * - `oneAPI Samples Catalog `_ - Explore the complete list of oneAPI code samples in the oneAPI Samples Catalog (GitHub*). - These samples were designed to help you develop, offload, and optimize multiarchitecture applications targeting CPUs, GPUs, and FPGAs. \ No newline at end of file + These samples were designed to help you develop, offload, and optimize multiarchitecture applications targeting CPUs, GPUs, and FPGAs. diff --git a/documentation/library_guide/notices_disclaimers.rst b/documentation/library_guide/notices_disclaimers.rst index 0cd36d35357..08cd323b3e7 100644 --- a/documentation/library_guide/notices_disclaimers.rst +++ b/documentation/library_guide/notices_disclaimers.rst @@ -20,7 +20,7 @@ License oneDPL is licensed under Apache License Version 2.0 with LLVM exceptions. -Refer to the `LICENSE `_ file for the full license text and copyright notice. +Refer to the `LICENSE `_ file for the full license text and copyright notice. diff --git a/documentation/library_guide/onedpl_gsg.rst b/documentation/library_guide/onedpl_gsg.rst index 13aea171737..0fc76ce6f58 100644 --- a/documentation/library_guide/onedpl_gsg.rst +++ b/documentation/library_guide/onedpl_gsg.rst @@ -12,7 +12,7 @@ programming efforts across devices for high performance parallel applications. * Macros -For general information about |onedpl_short|, visit the `oneDPL GitHub* repository `_, +For general information about |onedpl_short|, visit the `oneDPL GitHub* repository `_, or visit the |onedpl_library_guide|_ and the `Intel® oneAPI DPC++ Library main page `_. diff --git a/documentation/library_guide/requirements.txt b/documentation/library_guide/requirements.txt index fea4cdc7936..91d79101464 100644 --- a/documentation/library_guide/requirements.txt +++ b/documentation/library_guide/requirements.txt @@ -1,6 +1,6 @@ breathe==4.9.1 sphinx==4.4.0 -jinja2==3.1.4 +jinja2==3.1.5 # https://github.com/sphinx-doc/sphinx/issues/9923 docutils==0.15 # Pin versions to avoid a compatibility issue: "The sphinxcontrib. extension used by this project needs at least Sphinx v5.0" diff --git a/documentation/release_notes.rst b/documentation/release_notes.rst index 1c462a3a967..10bc53ae39a 100644 --- a/documentation/release_notes.rst +++ b/documentation/release_notes.rst @@ -848,11 +848,11 @@ Known Issues and Limitations .. [#fnote1] The sorting algorithms in oneDPL use Radix sort for arithmetic data types and ``sycl::half`` (since oneDPL 2022.6) compared with ``std::less`` or ``std::greater``, otherwise Merge sort. -.. _`oneDPL Guide`: https://oneapi-src.github.io/oneDPL/index.html +.. _`oneDPL Guide`: https://uxlfoundation.github.io/oneDPL/index.html .. _`Intel® oneAPI Threading Building Blocks (oneTBB) Release Notes`: https://www.intel.com/content/www/us/en/developer/articles/release-notes/intel-oneapi-threading-building-blocks-release-notes.html -.. _`restrictions and known limitations`: https://oneapi-src.github.io/oneDPL/introduction.html#restrictions. -.. _`Tested Standard C++ API`: https://oneapi-src.github.io/oneDPL/api_for_sycl_kernels/tested_standard_cpp_api.html#tested-standard-c-api-reference -.. _`Macros`: https://oneapi-src.github.io/oneDPL/macros.html -.. _`2022.0 Changes`: https://oneapi-src.github.io/oneDPL/oneDPL_2022.0_changes.html +.. _`restrictions and known limitations`: https://uxlfoundation.github.io/oneDPL/introduction.html#restrictions. +.. _`Tested Standard C++ API`: https://uxlfoundation.github.io/oneDPL/api_for_sycl_kernels/tested_standard_cpp_api.html#tested-standard-c-api-reference +.. _`Macros`: https://uxlfoundation.github.io/oneDPL/macros.html +.. _`2022.0 Changes`: https://uxlfoundation.github.io/oneDPL/oneDPL_2022.0_changes.html .. _`sycl device copyable`: https://registry.khronos.org/SYCL/specs/sycl-2020/html/sycl-2020.html#sec::device.copyable .. _`oneAPI DPC++ Library Manual Migration Guide`: https://www.intel.com/content/www/us/en/developer/articles/guide/oneapi-dpcpp-library-manual-migration.html diff --git a/examples/convex_hull/README.md b/examples/convex_hull/README.md index 109f687257b..be4cee608ff 100644 --- a/examples/convex_hull/README.md +++ b/examples/convex_hull/README.md @@ -25,7 +25,7 @@ Correctness of the convex hull is checked by `std::any_of` algorithm using `coun ## License -This code example is licensed under [Apache License Version 2.0 with LLVM exceptions](https://github.com/oneapi-src/oneDPL/blob/main/LICENSE.txt). Refer to the "[LICENSE](licensing/LICENSE.txt)" file for the full license text and copyright notice. +This code example is licensed under [Apache License Version 2.0 with LLVM exceptions](https://github.com/uxlfoundation/oneDPL/blob/main/LICENSE.txt). Refer to the "[LICENSE](licensing/LICENSE.txt)" file for the full license text and copyright notice. ## Building the 'Convex hull' Program diff --git a/examples/dot_product/README.md b/examples/dot_product/README.md index 4c4732894d3..c3e4c4d1577 100644 --- a/examples/dot_product/README.md +++ b/examples/dot_product/README.md @@ -11,7 +11,7 @@ This example contains the oneDPL-based implementation of dot product based on `s ## License -This code example is licensed under [Apache License Version 2.0 with LLVM exceptions](https://github.com/oneapi-src/oneDPL/blob/main/LICENSE.txt). Refer to the "[LICENSE](licensing/LICENSE.txt)" file for the full license text and copyright notice. +This code example is licensed under [Apache License Version 2.0 with LLVM exceptions](https://github.com/uxlfoundation/oneDPL/blob/main/LICENSE.txt). Refer to the "[LICENSE](licensing/LICENSE.txt)" file for the full license text and copyright notice. ## Building the 'Dot product' Program diff --git a/include/oneapi/dpl/pstl/hetero/dpcpp/parallel_backend_sycl.h b/include/oneapi/dpl/pstl/hetero/dpcpp/parallel_backend_sycl.h index c0568a312ca..7e38c1c4c3f 100644 --- a/include/oneapi/dpl/pstl/hetero/dpcpp/parallel_backend_sycl.h +++ b/include/oneapi/dpl/pstl/hetero/dpcpp/parallel_backend_sycl.h @@ -486,65 +486,29 @@ struct __parallel_transform_scan_static_single_group_submitter<_Inclusive, _Elem __hdl.parallel_for<_ScanKernelName...>( sycl::nd_range<1>(_WGSize, _WGSize), [=](sycl::nd_item<1> __self_item) { const auto& __group = __self_item.get_group(); - const auto& __subgroup = __self_item.get_sub_group(); // This kernel is only launched for sizes less than 2^16 const ::std::uint16_t __item_id = __self_item.get_local_linear_id(); - const ::std::uint16_t __subgroup_id = __subgroup.get_group_id(); - const ::std::uint16_t __subgroup_size = __subgroup.get_local_linear_range(); - -#if _ONEDPL_SYCL_SUB_GROUP_LOAD_STORE_PRESENT - constexpr bool __can_use_subgroup_load_store = - _IsFullGroup && oneapi::dpl::__internal::__range_has_raw_ptr_iterator_v<::std::decay_t<_InRng>>; -#else - constexpr bool __can_use_subgroup_load_store = false; -#endif auto __lacc_ptr = __dpl_sycl::__get_accessor_ptr(__lacc); - if constexpr (__can_use_subgroup_load_store) - { - _ONEDPL_PRAGMA_UNROLL - for (::std::uint16_t __i = 0; __i < _ElemsPerItem; ++__i) - { - auto __idx = __i * _WGSize + __subgroup_id * __subgroup_size; - auto __val = __unary_op(__subgroup.load(__in_rng.begin() + __idx)); - __subgroup.store(__lacc_ptr + __idx, __val); - } - } - else + for (std::uint16_t __idx = __item_id; __idx < __n; __idx += _WGSize) { - for (::std::uint16_t __idx = __item_id; __idx < __n; __idx += _WGSize) - { - __lacc[__idx] = __unary_op(__in_rng[__idx]); - } + __lacc[__idx] = __unary_op(__in_rng[__idx]); } __scan_work_group<_ValueType, _Inclusive>(__group, __lacc_ptr, __lacc_ptr + __n, __lacc_ptr, __bin_op, __init); - if constexpr (__can_use_subgroup_load_store) + for (std::uint16_t __idx = __item_id; __idx < __n; __idx += _WGSize) { - _ONEDPL_PRAGMA_UNROLL - for (::std::uint16_t __i = 0; __i < _ElemsPerItem; ++__i) - { - auto __idx = __i * _WGSize + __subgroup_id * __subgroup_size; - auto __val = __subgroup.load(__lacc_ptr + __idx); - __subgroup.store(__out_rng.begin() + __idx, __val); - } + __out_rng[__idx] = __lacc[__idx]; } - else - { - for (::std::uint16_t __idx = __item_id; __idx < __n; __idx += _WGSize) - { - __out_rng[__idx] = __lacc[__idx]; - } - const ::std::uint16_t __residual = __n % _WGSize; - const ::std::uint16_t __residual_start = __n - __residual; - if (__item_id < __residual) - { - auto __idx = __residual_start + __item_id; - __out_rng[__idx] = __lacc[__idx]; - } + const std::uint16_t __residual = __n % _WGSize; + const std::uint16_t __residual_start = __n - __residual; + if (__item_id < __residual) + { + auto __idx = __residual_start + __item_id; + __out_rng[__idx] = __lacc[__idx]; } }); }); @@ -592,35 +556,12 @@ struct __parallel_copy_if_static_single_group_submitter<_Size, _ElemsPerItem, _W sycl::nd_range<1>(_WGSize, _WGSize), [=](sycl::nd_item<1> __self_item) { auto __res_ptr = __result_and_scratch_storage_t::__get_usm_or_buffer_accessor_ptr(__res_acc); const auto& __group = __self_item.get_group(); - const auto& __subgroup = __self_item.get_sub_group(); // This kernel is only launched for sizes less than 2^16 const ::std::uint16_t __item_id = __self_item.get_local_linear_id(); - const ::std::uint16_t __subgroup_id = __subgroup.get_group_id(); - const ::std::uint16_t __subgroup_size = __subgroup.get_local_linear_range(); - -#if _ONEDPL_SYCL_SUB_GROUP_LOAD_STORE_PRESENT - constexpr bool __can_use_subgroup_load_store = - _IsFullGroup && oneapi::dpl::__internal::__range_has_raw_ptr_iterator_v<::std::decay_t<_InRng>>; -#else - constexpr bool __can_use_subgroup_load_store = false; -#endif auto __lacc_ptr = __dpl_sycl::__get_accessor_ptr(__lacc); - if constexpr (__can_use_subgroup_load_store) - { - _ONEDPL_PRAGMA_UNROLL - for (::std::uint16_t __i = 0; __i < _ElemsPerItem; ++__i) - { - auto __idx = __i * _WGSize + __subgroup_id * __subgroup_size; - uint16_t __val = __unary_op(__subgroup.load(__in_rng.begin() + __idx)); - __subgroup.store(__lacc_ptr + __idx, __val); - } - } - else + for (std::uint16_t __idx = __item_id; __idx < __n; __idx += _WGSize) { - for (::std::uint16_t __idx = __item_id; __idx < __n; __idx += _WGSize) - { - __lacc[__idx] = __unary_op(__in_rng[__idx]); - } + __lacc[__idx] = __unary_op(__in_rng[__idx]); } __scan_work_group<_ValueType, /* _Inclusive */ false>( @@ -939,12 +880,10 @@ struct __write_red_by_seg using std::get; auto __out_keys = get<0>(__out_rng.tuple()); auto __out_values = get<1>(__out_rng.tuple()); - using _KeyType = oneapi::dpl::__internal::__value_t; - using _ValType = oneapi::dpl::__internal::__value_t; - const _KeyType& __next_key = get<2>(__tup); - const _KeyType& __current_key = get<3>(__tup); - const _ValType& __current_value = get<1>(get<0>(__tup)); + const auto& __next_key = get<2>(__tup); + const auto& __current_key = get<3>(__tup); + const auto& __current_value = get<1>(get<0>(__tup)); const bool __is_seg_end = get<1>(__tup); const std::size_t __out_idx = get<0>(get<0>(__tup)); @@ -2492,8 +2431,6 @@ __parallel_reduce_by_segment(oneapi::dpl::__internal::__device_backend_tag, _Exe const auto __n = __keys.size(); - using __diff_type = oneapi::dpl::__internal::__difference_t<_Range1>; - using __key_type = oneapi::dpl::__internal::__value_t<_Range1>; using __val_type = oneapi::dpl::__internal::__value_t<_Range2>; // Prior to icpx 2025.0, the reduce-then-scan path performs poorly and should be avoided. #if !defined(__INTEL_LLVM_COMPILER) || __INTEL_LLVM_COMPILER >= 20250000 diff --git a/include/oneapi/dpl/pstl/hetero/dpcpp/parallel_backend_sycl_merge.h b/include/oneapi/dpl/pstl/hetero/dpcpp/parallel_backend_sycl_merge.h index df7e31ba501..5e405b34a54 100644 --- a/include/oneapi/dpl/pstl/hetero/dpcpp/parallel_backend_sycl_merge.h +++ b/include/oneapi/dpl/pstl/hetero/dpcpp/parallel_backend_sycl_merge.h @@ -227,11 +227,6 @@ struct __parallel_merge_submitter_large<_IdType, _CustomName, nd_range_params eval_nd_range_params(_ExecutionPolicy&& __exec, const _Range1& __rng1, const _Range2& __rng2) const { - using _Range1ValueType = oneapi::dpl::__internal::__value_t<_Range1>; - using _Range2ValueType = oneapi::dpl::__internal::__value_t<_Range2>; - using _RangeValueType = std::conditional_t<(sizeof(_Range1ValueType) > sizeof(_Range2ValueType)), - _Range1ValueType, _Range2ValueType>; - const std::size_t __n = __rng1.size() + __rng2.size(); // Empirical number of values to process per work-item diff --git a/include/oneapi/dpl/pstl/hetero/dpcpp/parallel_backend_sycl_utils.h b/include/oneapi/dpl/pstl/hetero/dpcpp/parallel_backend_sycl_utils.h index f5275b501a9..3d9d923d80d 100644 --- a/include/oneapi/dpl/pstl/hetero/dpcpp/parallel_backend_sycl_utils.h +++ b/include/oneapi/dpl/pstl/hetero/dpcpp/parallel_backend_sycl_utils.h @@ -74,7 +74,7 @@ __slm_adjusted_work_group_size(const _ExecutionPolicy& __policy, _Size __local_m if (__wg_size == 0) __wg_size = __max_work_group_size(__policy); auto __local_mem_size = __policy.queue().get_device().template get_info(); - return sycl::min(__local_mem_size / __local_mem_per_wi, __wg_size); + return sycl::min<_Size>(__local_mem_size / __local_mem_per_wi, __wg_size); } #if _ONEDPL_USE_SUB_GROUPS diff --git a/include/oneapi/dpl/pstl/hetero/dpcpp/sycl_defs.h b/include/oneapi/dpl/pstl/hetero/dpcpp/sycl_defs.h index 80d6fc3fa44..72540c492b3 100644 --- a/include/oneapi/dpl/pstl/hetero/dpcpp/sycl_defs.h +++ b/include/oneapi/dpl/pstl/hetero/dpcpp/sycl_defs.h @@ -75,7 +75,7 @@ #define _ONEDPL_SYCL2020_LOCAL_ACCESSOR_PRESENT (!_ONEDPL_LIBSYCL_VERSION_LESS_THAN(60000)) // The unified future supporting USM memory and buffers is only supported after DPC++ 2023.1 but not by 2023.2. #define _ONEDPL_SYCL2020_DEFAULT_ACCESSOR_CONSTRUCTOR_PRESENT \ - (!_ONEDPL_LIBSYCL_VERSION_LESS_THAN(60100) || _ONEDPL_LIBSYCL_VERSION != 60200) + (!_ONEDPL_LIBSYCL_VERSION_LESS_THAN(60100) && _ONEDPL_LIBSYCL_VERSION != 60200) #define _ONEDPL_SYCL2020_HOST_TARGET_PRESENT (!_ONEDPL_LIBSYCL_VERSION_LESS_THAN(60200)) #define _ONEDPL_SYCL2020_HOST_ACCESSOR_PRESENT (!_ONEDPL_LIBSYCL_VERSION_LESS_THAN(60200)) #define _ONEDPL_SYCL2020_GET_HOST_ACCESS_PRESENT (!_ONEDPL_LIBSYCL_VERSION_LESS_THAN(60200)) @@ -89,9 +89,6 @@ #define _ONEDPL_SYCL_DEVICE_COPYABLE_SPECIALIZATION_BROKEN (_ONEDPL_LIBSYCL_VERSION_LESS_THAN(70100)) -// TODO: determine which compiler configurations provide subgroup load/store -#define _ONEDPL_SYCL_SUB_GROUP_LOAD_STORE_PRESENT false - // Macro to check if we are compiling for SPIR-V devices. This macro must only be used within // SYCL kernels for determining SPIR-V compilation. Using this macro on the host may lead to incorrect behavior. #ifndef _ONEDPL_DETECT_SPIRV_COMPILATION // Check if overridden for testing diff --git a/jenkinsfiles/RHEL_Check.groovy b/jenkinsfiles/RHEL_Check.groovy index 967df926148..476c0a3309e 100644 --- a/jenkinsfiles/RHEL_Check.groovy +++ b/jenkinsfiles/RHEL_Check.groovy @@ -104,7 +104,7 @@ pipeline { parameters { string(name: 'Commit_id', defaultValue: 'None', description: '',) string(name: 'PR_number', defaultValue: 'None', description: '',) - string(name: 'Repository', defaultValue: 'oneapi-src/oneDPL', description: '',) + string(name: 'Repository', defaultValue: 'uxlfoundation/oneDPL', description: '',) string(name: 'User', defaultValue: 'None', description: '',) string(name: 'OneAPI_Package_Date', defaultValue: 'Default', description: '',) string(name: 'Base_branch', defaultValue: 'main', description: '',) diff --git a/jenkinsfiles/UB18_Check.groovy b/jenkinsfiles/UB18_Check.groovy index cb1da9ced2e..12e6751e72b 100644 --- a/jenkinsfiles/UB18_Check.groovy +++ b/jenkinsfiles/UB18_Check.groovy @@ -104,7 +104,7 @@ pipeline { parameters { string(name: 'Commit_id', defaultValue: 'None', description: '',) string(name: 'PR_number', defaultValue: 'None', description: '',) - string(name: 'Repository', defaultValue: 'oneapi-src/oneDPL', description: '',) + string(name: 'Repository', defaultValue: 'uxlfoundation/oneDPL', description: '',) string(name: 'User', defaultValue: 'None', description: '',) string(name: 'OneAPI_Package_Date', defaultValue: 'Default', description: '',) string(name: 'Base_branch', defaultValue: 'main', description: '',) diff --git a/jenkinsfiles/UB20_Check.groovy b/jenkinsfiles/UB20_Check.groovy index 0e4da5798c0..23f85b592a6 100644 --- a/jenkinsfiles/UB20_Check.groovy +++ b/jenkinsfiles/UB20_Check.groovy @@ -88,7 +88,7 @@ pipeline { parameters { string(name: 'Commit_id', defaultValue: 'None', description: '',) string(name: 'PR_number', defaultValue: 'None', description: '',) - string(name: 'Repository', defaultValue: 'oneapi-src/oneDPL', description: '',) + string(name: 'Repository', defaultValue: 'uxlfoundation/oneDPL', description: '',) string(name: 'User', defaultValue: 'None', description: '',) string(name: 'OneAPI_Package_Date', defaultValue: 'Default', description: '',) string(name: 'Base_branch', defaultValue: 'main', description: '',) diff --git a/jenkinsfiles/Windows_Check.groovy b/jenkinsfiles/Windows_Check.groovy index bf871f24bff..268460728d7 100644 --- a/jenkinsfiles/Windows_Check.groovy +++ b/jenkinsfiles/Windows_Check.groovy @@ -113,7 +113,7 @@ pipeline { parameters { string(name: 'Commit_id', defaultValue: 'None', description: '',) string(name: 'PR_number', defaultValue: 'None', description: '',) - string(name: 'Repository', defaultValue: 'oneapi-src/oneDPL', description: '',) + string(name: 'Repository', defaultValue: 'uxlfoundation/oneDPL', description: '',) string(name: 'User', defaultValue: 'None', description: '',) string(name: 'OneAPI_Package_Date', defaultValue: 'Default', description: '',) string(name: 'Base_branch', defaultValue: 'main', description: '',) @@ -200,7 +200,7 @@ pipeline { bat script: """ d: cd ${env.WORKSPACE} - git clone https://github.com/oneapi-src/oneDPL.git src + git clone https://github.com/uxlfoundation/oneDPL.git src cd src git config --local --add remote.origin.fetch +refs/pull/${env.PR_number}/head:refs/remotes/origin/pr/${env.PR_number} git pull origin diff --git a/test/parallel_api/numeric/numeric.ops/reduce_by_segment_zip.pass.cpp b/test/parallel_api/numeric/numeric.ops/reduce_by_segment_zip.pass.cpp index dba44019d00..3670eb3a910 100644 --- a/test/parallel_api/numeric/numeric.ops/reduce_by_segment_zip.pass.cpp +++ b/test/parallel_api/numeric/numeric.ops/reduce_by_segment_zip.pass.cpp @@ -108,6 +108,59 @@ test_with_usm() EXPECT_EQ_N(exp_values1, output_values1, n, "wrong values1 from reduce_by_segment"); EXPECT_EQ_N(exp_values2, output_values2, n, "wrong values2 from reduce_by_segment"); } + +template +void +test_zip_with_discard() +{ + constexpr sycl::usm::alloc alloc_type = sycl::usm::alloc::device; + sycl::queue q = TestUtils::get_test_queue(); + + constexpr int n = 5; + + //data initialization + int keys1[n] = {1, 1, 2, 2, 3}; + int keys2[n] = {1, 1, 2, 2, 3}; + int values1[n] = {1, 1, 1, 1, 1}; + int values2[n] = {2, 2, 2, 2, 2}; + int output_keys[n] = {}; + int output_values[n] = {}; + + // allocate USM memory and copying data to USM shared/device memory + TestUtils::usm_data_transfer dt_helper1(q, keys1, n); + TestUtils::usm_data_transfer dt_helper2(q, keys2, n); + TestUtils::usm_data_transfer dt_helper3(q, values1, n); + TestUtils::usm_data_transfer dt_helper4(q, values2, n); + TestUtils::usm_data_transfer dt_helper5(q, output_keys, n); + TestUtils::usm_data_transfer dt_helper6(q, output_values, n); + auto d_keys1 = dt_helper1.get_data(); + auto d_keys2 = dt_helper2.get_data(); + auto d_values1 = dt_helper3.get_data(); + auto d_values2 = dt_helper4.get_data(); + auto d_output_keys = dt_helper5.get_data(); + auto d_output_values = dt_helper6.get_data(); + + //make zip iterators + auto begin_keys_in = oneapi::dpl::make_zip_iterator(d_keys1, d_keys2); + auto end_keys_in = oneapi::dpl::make_zip_iterator(d_keys1 + n, d_keys2 + n); + auto begin_vals_in = oneapi::dpl::make_zip_iterator(d_values1, d_values2); + auto begin_keys_out = oneapi::dpl::make_zip_iterator(d_output_keys, oneapi::dpl::discard_iterator()); + auto begin_vals_out = oneapi::dpl::make_zip_iterator(oneapi::dpl::discard_iterator(), d_output_values); + + //run reduce_by_segment algorithm + auto new_last = oneapi::dpl::reduce_by_segment(TestUtils::make_device_policy(q), begin_keys_in, + end_keys_in, begin_vals_in, begin_keys_out, begin_vals_out, + std::equal_to<>(), TestUtils::TupleAddFunctor()); + + //retrieve result on the host and check the result + dt_helper5.retrieve_data(output_keys); + dt_helper6.retrieve_data(output_values); + + const int exp_keys[n] = {1, 2, 3}; + const int exp_values[n] = {4, 4, 2}; + EXPECT_EQ_N(exp_keys, output_keys, n, "wrong keys from reduce_by_segment"); + EXPECT_EQ_N(exp_values, output_values, n, "wrong values from reduce_by_segment"); +} #endif //The code below for test a call of reduce_by_segment with zip iterators was kept "as is", as an example reported by a user; just "memory deallocation" added. @@ -118,6 +171,8 @@ int main() test_with_usm(); // Run tests for USM device memory test_with_usm(); + + test_zip_with_discard(); #endif return TestUtils::done(TEST_DPCPP_BACKEND_PRESENT);