diff --git a/.ci/env/apt.sh b/.ci/env/apt.sh index e9a5549841f..2665ed9a65d 100755 --- a/.ci/env/apt.sh +++ b/.ci/env/apt.sh @@ -32,7 +32,7 @@ function add_repo { } function install_dpcpp { - sudo apt-get install -y intel-oneapi-compiler-dpcpp-cpp-2024.1 + sudo apt-get install -y intel-oneapi-compiler-dpcpp-cpp-2024.2 sudo bash -c 'echo libintelocl.so > /etc/OpenCL/vendors/intel-cpu.icd' } @@ -63,7 +63,7 @@ function install_qemu_emulation_apt { } function install_qemu_emulation_deb { - qemu_deb=qemu-user-static_8.2.1+ds-1~bpo12+1_amd64.deb + qemu_deb=qemu-user-static_9.0.1+ds-1~bpo12+1_amd64.deb set -eo pipefail wget http://ftp.debian.org/debian/pool/main/q/qemu/${qemu_deb} sudo dpkg -i ${qemu_deb} diff --git a/.github/CODEOWNERS b/.github/CODEOWNERS index 5871fb84661..d6a47717935 100644 --- a/.github/CODEOWNERS +++ b/.github/CODEOWNERS @@ -1,6 +1,6 @@ # Documentation owners and reviewers -/docs/ @Vika-F @maria-Petrova @Alexsandruss @aepanchi -*.md @Vika-F @maria-Petrova @Alexsandruss @aepanchi +/docs/ @Vika-F @maria-Petrova @Alexsandruss @bdmoore1 +*.md @Vika-F @maria-Petrova @Alexsandruss @bdmoore1 # TTP files third-party* @maria-Petrova @@ -22,7 +22,7 @@ deploy/ @Alexsandruss @napetrov @homksei @ahuber21 @ethanglaser dev/ @Alexsandruss @napetrov @homksei @ahuber21 @ethanglaser # C++ code -cpp/ @Alexsandruss @samir-nasibli @KulikovNikita @Alexandr-Solovev +cpp/ @Alexsandruss @samir-nasibli @Alexandr-Solovev # Tree based methods dtrees @razdoburdin @ahuber21 @avolkov-intel @icfaust diff --git a/.github/workflows/openssf-scorecard.yml b/.github/workflows/openssf-scorecard.yml index 89f5d7b860e..42deb5ad780 100644 --- a/.github/workflows/openssf-scorecard.yml +++ b/.github/workflows/openssf-scorecard.yml @@ -31,7 +31,7 @@ jobs: persist-credentials: false - name: "Run analysis" - uses: ossf/scorecard-action@dc50aa9510b46c811795eb24b2f1ba02a914e534 # v2.3.3 + uses: ossf/scorecard-action@62b2cac7ed8198b15735ed49ab1e5cf35480ba46 # v2.4.0 with: results_file: results.sarif results_format: sarif diff --git a/WORKSPACE b/WORKSPACE index c20ee8fc7e3..3cba5bcd224 100644 --- a/WORKSPACE +++ b/WORKSPACE @@ -33,8 +33,8 @@ micromkl_repo( micromkl_dpc_repo( name = "micromkl_dpc", root_env_var = "MKLGPUFPKROOT", - url = "https://github.com/oneapi-src/oneDAL/releases/download/Dependencies/mklgpufpk_lnx_2024-02-20.tgz", - sha256 = "1c60914461aafa5e5512181c7d5c1fdbdeff83746dbd980fe97074a3b65fc1ed", + url = "https://github.com/oneapi-src/oneDAL/releases/download/Dependencies/mklgpufpk_lnx_20240605.tgz", + sha256 = "0787a92e9580ed6b9fb97d054a0ed77994dbc18b4b3fb099451cb1e6ebdf4f16", ) load("@onedal//dev/bazel/deps:openblas.bzl", "openblas_repo") @@ -115,8 +115,8 @@ http_archive( http_archive( name = "fmt", - url = "https://github.com/fmtlib/fmt/archive/11.0.1.tar.gz", - sha256 = "7d009f7f89ac84c0a83f79ed602463d092fbf66763766a907c97fd02b100f5e9", - strip_prefix = "fmt-11.0.1", + url = "https://github.com/fmtlib/fmt/archive/11.0.2.tar.gz", + sha256 = "6cb1e6d37bdcb756dbbe59be438790db409cdb4868c66e888d5df9f13f7c027f", + strip_prefix = "fmt-11.0.2", build_file = "@onedal//dev/bazel/deps:fmt.tpl.BUILD", ) diff --git a/cpp/daal/include/algorithms/pca/pca_types.h b/cpp/daal/include/algorithms/pca/pca_types.h index 8c8472b140b..9a9daa4f987 100644 --- a/cpp/daal/include/algorithms/pca/pca_types.h +++ b/cpp/daal/include/algorithms/pca/pca_types.h @@ -670,7 +670,7 @@ class DAAL_EXPORT BaseBatchParameter : public daal::algorithms::Parameter BaseBatchParameter(); DAAL_UINT64 resultsToCompute; /*!< 64 bit integer flag that indicates the results to compute */ - size_t nComponents; /*!< number of components for reduced implementation */ + size_t nComponents; /*!< number of components for reduced implementation (applicable for batch mode only) */ bool isDeterministic; /*!< sign flip if required */ bool doScale; /*!< scaling if required */ bool isCorrelation; /*!< correlation is provided */ diff --git a/cpp/oneapi/dal/algo/BUILD b/cpp/oneapi/dal/algo/BUILD index ab93feb58c0..e93804d2e7e 100644 --- a/cpp/oneapi/dal/algo/BUILD +++ b/cpp/oneapi/dal/algo/BUILD @@ -18,6 +18,7 @@ ALGOS = [ "cosine_distance", "dbscan", "decision_tree", + "finiteness_checker", "jaccard", "kmeans", "kmeans_init", diff --git a/cpp/oneapi/dal/algo/finiteness_checker/BUILD b/cpp/oneapi/dal/algo/finiteness_checker/BUILD new file mode 100644 index 00000000000..de390465f9d --- /dev/null +++ b/cpp/oneapi/dal/algo/finiteness_checker/BUILD @@ -0,0 +1,61 @@ +package(default_visibility = ["//visibility:public"]) +load("@onedal//dev/bazel:dal.bzl", + "dal_module", + "dal_test_suite", +) + +dal_module( + name = "finiteness_checker", + auto = True, + dal_deps = [ + "@onedal//cpp/oneapi/dal:core", + "@onedal//cpp/oneapi/dal/backend/primitives:reduction", + ], + extra_deps = [ + "@onedal//cpp/daal:data_management", + ] +) + +dal_test_suite( + name = "cpu_tests", + private = True, + compile_as = [ "c++" ], + srcs = glob([ + "backend/cpu/test/*.cpp", + ]), + dal_deps = [ + ":finiteness_checker", + ], +) + +dal_test_suite( + name = "gpu_tests", + private = True, + compile_as = [ "dpc++" ], + srcs = glob([ + "backend/gpu/test/*.cpp", + ]), + dal_deps = [ + ":finiteness_checker", + ], +) + +dal_test_suite( + name = "interface_tests", + framework = "catch2", + srcs = glob([ + "test/*.cpp", + ]), + dal_deps = [ + ":finiteness_checker", + ], +) + +dal_test_suite( + name = "tests", + tests = [ + ":cpu_tests", + ":gpu_tests", + ":interface_tests", + ], +) diff --git a/cpp/oneapi/dal/algo/finiteness_checker/backend/cpu/compute_kernel.hpp b/cpp/oneapi/dal/algo/finiteness_checker/backend/cpu/compute_kernel.hpp new file mode 100644 index 00000000000..c1a1b8fc3c4 --- /dev/null +++ b/cpp/oneapi/dal/algo/finiteness_checker/backend/cpu/compute_kernel.hpp @@ -0,0 +1,39 @@ +/******************************************************************************* +* Copyright contributors to the oneDAL project +* +* Licensed under the Apache License, Version 2.0 (the "License"); +* you may not use this file except in compliance with the License. +* You may obtain a copy of the License at +* +* http://www.apache.org/licenses/LICENSE-2.0 +* +* Unless required by applicable law or agreed to in writing, software +* distributed under the License is distributed on an "AS IS" BASIS, +* WITHOUT WARRANTIES OR CONDITIONS OF ANY KIND, either express or implied. +* See the License for the specific language governing permissions and +* limitations under the License. +*******************************************************************************/ + +#pragma once + +#include "oneapi/dal/algo/finiteness_checker/compute_types.hpp" +#include "oneapi/dal/backend/dispatcher.hpp" +#include "oneapi/dal/table/homogen.hpp" + +namespace oneapi::dal::finiteness_checker::backend { + +template +struct compute_kernel_cpu { + compute_result operator()(const dal::backend::context_cpu& ctx, + const detail::descriptor_base& params, + const compute_input& input) const; + +#ifdef ONEDAL_DATA_PARALLEL + void operator()(const dal::backend::context_cpu& ctx, + const detail::descriptor_base& params, + const table& data, + bool& res) const; +#endif +}; + +} // namespace oneapi::dal::finiteness_checker::backend diff --git a/cpp/oneapi/dal/algo/finiteness_checker/backend/cpu/compute_kernel_dense.cpp b/cpp/oneapi/dal/algo/finiteness_checker/backend/cpu/compute_kernel_dense.cpp new file mode 100644 index 00000000000..60d965046ec --- /dev/null +++ b/cpp/oneapi/dal/algo/finiteness_checker/backend/cpu/compute_kernel_dense.cpp @@ -0,0 +1,73 @@ +/******************************************************************************* +* Copyright contributors to the oneDAL project +* +* Licensed under the Apache License, Version 2.0 (the "License"); +* you may not use this file except in compliance with the License. +* You may obtain a copy of the License at +* +* http://www.apache.org/licenses/LICENSE-2.0 +* +* Unless required by applicable law or agreed to in writing, software +* distributed under the License is distributed on an "AS IS" BASIS, +* WITHOUT WARRANTIES OR CONDITIONS OF ANY KIND, either express or implied. +* See the License for the specific language governing permissions and +* limitations under the License. +*******************************************************************************/ + +#include + +#include "oneapi/dal/algo/finiteness_checker/backend/cpu/compute_kernel.hpp" +#include "oneapi/dal/backend/interop/common.hpp" +#include "oneapi/dal/backend/interop/error_converter.hpp" +#include "oneapi/dal/backend/interop/table_conversion.hpp" +#include "oneapi/dal/exceptions.hpp" + +#include "oneapi/dal/table/row_accessor.hpp" + +namespace oneapi::dal::finiteness_checker::backend { + +using dal::backend::context_cpu; +using input_t = compute_input; +using result_t = compute_result; +using descriptor_t = detail::descriptor_base; + +namespace interop = dal::backend::interop; + +template +static result_t call_daal_kernel(const context_cpu& ctx, + const descriptor_t& desc, + const table& data) { + const auto daal_data = interop::convert_to_daal_table(data); + + return result_t().set_finite( + daal::data_management::internal::allValuesAreFinite(*daal_data.get(), + desc.get_allow_NaN())); +} + +template +static result_t compute(const context_cpu& ctx, const descriptor_t& desc, const input_t& input) { + return call_daal_kernel(ctx, desc, input.get_data()); +} + +template +struct compute_kernel_cpu { + result_t operator()(const context_cpu& ctx, + const descriptor_t& desc, + const input_t& input) const { + return compute(ctx, desc, input); + } + +#ifdef ONEDAL_DATA_PARALLEL + void operator()(const context_cpu& ctx, + const descriptor_t& desc, + const table& data, + bool& res) const { + throw unimplemented(dal::detail::error_messages::method_not_implemented()); + } +#endif +}; + +template struct compute_kernel_cpu; +template struct compute_kernel_cpu; + +} // namespace oneapi::dal::finiteness_checker::backend diff --git a/cpp/oneapi/dal/algo/finiteness_checker/backend/gpu/compute_kernel.hpp b/cpp/oneapi/dal/algo/finiteness_checker/backend/gpu/compute_kernel.hpp new file mode 100644 index 00000000000..51fc6d4a35e --- /dev/null +++ b/cpp/oneapi/dal/algo/finiteness_checker/backend/gpu/compute_kernel.hpp @@ -0,0 +1,39 @@ +/******************************************************************************* +* Copyright contributors to the oneDAL project +* +* Licensed under the Apache License, Version 2.0 (the "License"); +* you may not use this file except in compliance with the License. +* You may obtain a copy of the License at +* +* http://www.apache.org/licenses/LICENSE-2.0 +* +* Unless required by applicable law or agreed to in writing, software +* distributed under the License is distributed on an "AS IS" BASIS, +* WITHOUT WARRANTIES OR CONDITIONS OF ANY KIND, either express or implied. +* See the License for the specific language governing permissions and +* limitations under the License. +*******************************************************************************/ + +#pragma once + +#include "oneapi/dal/algo/finiteness_checker/compute_types.hpp" +#include "oneapi/dal/backend/dispatcher.hpp" +#include "oneapi/dal/table/homogen.hpp" + +namespace oneapi::dal::finiteness_checker::backend { + +template +struct compute_kernel_gpu { + compute_result operator()(const dal::backend::context_gpu& ctx, + const detail::descriptor_base& params, + const compute_input& input) const; + +#ifdef ONEDAL_DATA_PARALLEL + void operator()(const dal::backend::context_gpu& ctx, + const detail::descriptor_base& params, + const table& data, + bool& res); +#endif +}; + +} // namespace oneapi::dal::finiteness_checker::backend diff --git a/cpp/oneapi/dal/algo/finiteness_checker/backend/gpu/compute_kernel_dense_dpc.cpp b/cpp/oneapi/dal/algo/finiteness_checker/backend/gpu/compute_kernel_dense_dpc.cpp new file mode 100644 index 00000000000..09389a2c122 --- /dev/null +++ b/cpp/oneapi/dal/algo/finiteness_checker/backend/gpu/compute_kernel_dense_dpc.cpp @@ -0,0 +1,81 @@ +/******************************************************************************* +* Copyright contributors to the oneDAL project +* +* Licensed under the Apache License, Version 2.0 (the "License"); +* you may not use this file except in compliance with the License. +* You may obtain a copy of the License at +* +* http://www.apache.org/licenses/LICENSE-2.0 +* +* Unless required by applicable law or agreed to in writing, software +* distributed under the License is distributed on an "AS IS" BASIS, +* WITHOUT WARRANTIES OR CONDITIONS OF ANY KIND, either express or implied. +* See the License for the specific language governing permissions and +* limitations under the License. +*******************************************************************************/ + +#include "oneapi/dal/algo/finiteness_checker/backend/gpu/compute_kernel.hpp" +#include "oneapi/dal/backend/primitives/reduction.hpp" +#include "oneapi/dal/backend/primitives/utils.hpp" +#include "oneapi/dal/detail/profiler.hpp" + +namespace oneapi::dal::finiteness_checker::backend { + +using dal::backend::context_gpu; +using input_t = compute_input; +using result_t = compute_result; +using descriptor_t = detail::descriptor_base; + +namespace pr = dal::backend::primitives; + +template +bool compute_finiteness(sycl::queue& queue, + const pr::ndview& data_1d, + bool allowNaN, + const dal::backend::event_vector& deps = {}) { + Float out; + + if (allowNaN) { + ONEDAL_PROFILER_TASK(finiteness_checker.reduce, queue); + out = pr::reduce_1d(queue, data_1d, pr::logical_or{}, pr::isinf{}, deps); + } + else { + ONEDAL_PROFILER_TASK(finiteness_checker.reduce, queue); + out = pr::reduce_1d(queue, data_1d, pr::logical_or{}, pr::isinfornan{}, deps); + } + // invert out to match daal implementation (assert result is finite) + return !static_cast(out); +} + +template +static result_t compute(const context_gpu& ctx, const descriptor_t& desc, const input_t& input) { + auto& queue = ctx.get_queue(); + const auto data = input.get_data(); + const auto data_1d = pr::table2ndarray_1d(queue, data, sycl::usm::alloc::device); + return result_t{}.set_finite(compute_finiteness(queue, data_1d, desc.get_allow_NaN())); +} + +template +struct compute_kernel_gpu { + result_t operator()(const context_gpu& ctx, + const descriptor_t& desc, + const input_t& input) const { + return compute(ctx, desc, input); + } + +#ifdef ONEDAL_DATA_PARALLEL + void operator()(const context_gpu& ctx, + const descriptor_t& desc, + const table& data, + bool& res) { + auto& queue = ctx.get_queue(); + const auto data_1d = pr::table2ndarray_1d(queue, data, sycl::usm::alloc::device); + res = compute_finiteness(queue, data_1d, desc.get_allow_NaN()); + } +#endif +}; + +template struct compute_kernel_gpu; +template struct compute_kernel_gpu; + +} // namespace oneapi::dal::finiteness_checker::backend diff --git a/cpp/oneapi/dal/algo/finiteness_checker/common.cpp b/cpp/oneapi/dal/algo/finiteness_checker/common.cpp new file mode 100644 index 00000000000..29b8ea6aa21 --- /dev/null +++ b/cpp/oneapi/dal/algo/finiteness_checker/common.cpp @@ -0,0 +1,45 @@ +/******************************************************************************* +* Copyright contributors to the oneDAL project +* +* Licensed under the Apache License, Version 2.0 (the "License"); +* you may not use this file except in compliance with the License. +* You may obtain a copy of the License at +* +* http://www.apache.org/licenses/LICENSE-2.0 +* +* Unless required by applicable law or agreed to in writing, software +* distributed under the License is distributed on an "AS IS" BASIS, +* WITHOUT WARRANTIES OR CONDITIONS OF ANY KIND, either express or implied. +* See the License for the specific language governing permissions and +* limitations under the License. +*******************************************************************************/ + +#include "oneapi/dal/algo/finiteness_checker/common.hpp" +#include "oneapi/dal/exceptions.hpp" + +namespace oneapi::dal::finiteness_checker::detail { +namespace v1 { + +template +class descriptor_impl : public base { +public: + bool allowNaN = false; +}; + +template +descriptor_base::descriptor_base() : impl_(new descriptor_impl{}) {} + +template +bool descriptor_base::get_allow_NaN() const { + return impl_->allowNaN; +} + +template +void descriptor_base::set_allow_NaN(bool value) { + impl_->allowNaN = value; +} + +template class ONEDAL_EXPORT descriptor_base; + +} // namespace v1 +} // namespace oneapi::dal::finiteness_checker::detail diff --git a/cpp/oneapi/dal/algo/finiteness_checker/common.hpp b/cpp/oneapi/dal/algo/finiteness_checker/common.hpp new file mode 100644 index 00000000000..e9e5b36930c --- /dev/null +++ b/cpp/oneapi/dal/algo/finiteness_checker/common.hpp @@ -0,0 +1,140 @@ +/******************************************************************************* +* Copyright contributors to the oneDAL project +* +* Licensed under the Apache License, Version 2.0 (the "License"); +* you may not use this file except in compliance with the License. +* You may obtain a copy of the License at +* +* http://www.apache.org/licenses/LICENSE-2.0 +* +* Unless required by applicable law or agreed to in writing, software +* distributed under the License is distributed on an "AS IS" BASIS, +* WITHOUT WARRANTIES OR CONDITIONS OF ANY KIND, either express or implied. +* See the License for the specific language governing permissions and +* limitations under the License. +*******************************************************************************/ + +#pragma once + +#include "oneapi/dal/detail/common.hpp" +#include "oneapi/dal/table/common.hpp" + +namespace oneapi::dal::finiteness_checker { + +namespace task { +namespace v1 { + +/// Tag-type that parameterizes entities that are used to compute statistics, distance, and so on. +struct compute {}; +/// Alias tag-type for the dense method. +using by_default = compute; +} // namespace v1 + +using v1::compute; +using v1::by_default; + +} // namespace task + +namespace method { +namespace v1 { +struct dense {}; +using by_default = dense; +} // namespace v1 + +using v1::dense; +using v1::by_default; + +} // namespace method + +namespace detail { +namespace v1 { +struct descriptor_tag {}; + +template +class descriptor_impl; + +template +constexpr bool is_valid_float_v = dal::detail::is_one_of_v; + +template +constexpr bool is_valid_method_v = dal::detail::is_one_of_v; + +template +constexpr bool is_valid_task_v = dal::detail::is_one_of_v; + +template +class descriptor_base : public base { + static_assert(is_valid_task_v); + +public: + using tag_t = descriptor_tag; + using float_t = float; + using method_t = method::by_default; + using task_t = Task; + + descriptor_base(); + + bool get_allow_NaN() const; + +protected: + void set_allow_NaN(bool); + +private: + dal::detail::pimpl> impl_; +}; + +} // namespace v1 + +using v1::descriptor_tag; +using v1::descriptor_impl; +using v1::descriptor_base; + +using v1::is_valid_float_v; +using v1::is_valid_method_v; +using v1::is_valid_task_v; + +} // namespace detail + +namespace v1 { + +/// @tparam Float The floating-point type that the algorithm uses for +/// intermediate computations. Can be :expr:`float` or +/// :expr:`double`. +/// @tparam Method Tag-type that specifies an implementation of algorithm. Can +/// be :expr:`method::dense`. +/// @tparam Task Tag-type that specifies the type of the problem to solve. Can +/// be :expr:`task::compute`. +template +class descriptor : public detail::descriptor_base { + static_assert(detail::is_valid_float_v); + static_assert(detail::is_valid_method_v); + static_assert(detail::is_valid_task_v); + + using base_t = detail::descriptor_base; + +public: + using float_t = Float; + using method_t = Method; + using task_t = Task; + + /// Creates a new instance of the class with the default property values. + descriptor() = default; + + /// @remark default = False + bool get_allow_NaN() const { + return base_t::get_allow_NaN(); + } + + auto& set_allow_NaN(bool value) { + base_t::set_allow_NaN(value); + return *this; + } +}; + +} // namespace v1 + +using v1::descriptor; + +} // namespace oneapi::dal::finiteness_checker diff --git a/cpp/oneapi/dal/algo/finiteness_checker/compute.hpp b/cpp/oneapi/dal/algo/finiteness_checker/compute.hpp new file mode 100644 index 00000000000..92252303610 --- /dev/null +++ b/cpp/oneapi/dal/algo/finiteness_checker/compute.hpp @@ -0,0 +1,31 @@ +/******************************************************************************* +* Copyright contributors to the oneDAL project +* +* Licensed under the Apache License, Version 2.0 (the "License"); +* you may not use this file except in compliance with the License. +* You may obtain a copy of the License at +* +* http://www.apache.org/licenses/LICENSE-2.0 +* +* Unless required by applicable law or agreed to in writing, software +* distributed under the License is distributed on an "AS IS" BASIS, +* WITHOUT WARRANTIES OR CONDITIONS OF ANY KIND, either express or implied. +* See the License for the specific language governing permissions and +* limitations under the License. +*******************************************************************************/ + +#pragma once + +#include "oneapi/dal/algo/finiteness_checker/compute_types.hpp" +#include "oneapi/dal/algo/finiteness_checker/detail/compute_ops.hpp" +#include "oneapi/dal/compute.hpp" + +namespace oneapi::dal::detail { +namespace v1 { + +template +struct compute_ops + : dal::finiteness_checker::detail::compute_ops {}; + +} // namespace v1 +} // namespace oneapi::dal::detail diff --git a/cpp/oneapi/dal/algo/finiteness_checker/compute_types.cpp b/cpp/oneapi/dal/algo/finiteness_checker/compute_types.cpp new file mode 100644 index 00000000000..65c29d6630a --- /dev/null +++ b/cpp/oneapi/dal/algo/finiteness_checker/compute_types.cpp @@ -0,0 +1,71 @@ +/******************************************************************************* +* Copyright contributors to the oneDAL project +* +* Licensed under the Apache License, Version 2.0 (the "License"); +* you may not use this file except in compliance with the License. +* You may obtain a copy of the License at +* +* http://www.apache.org/licenses/LICENSE-2.0 +* +* Unless required by applicable law or agreed to in writing, software +* distributed under the License is distributed on an "AS IS" BASIS, +* WITHOUT WARRANTIES OR CONDITIONS OF ANY KIND, either express or implied. +* See the License for the specific language governing permissions and +* limitations under the License. +*******************************************************************************/ + +#include "oneapi/dal/algo/finiteness_checker/compute_types.hpp" +#include "oneapi/dal/detail/common.hpp" + +namespace oneapi::dal::finiteness_checker { + +template +class detail::v1::compute_input_impl : public base { +public: + compute_input_impl(const table& data) : data(data) {} + table data; +}; + +template +class detail::v1::compute_result_impl : public base { +public: + bool finite; +}; + +using detail::v1::compute_input_impl; +using detail::v1::compute_result_impl; + +namespace v1 { + +template +compute_input::compute_input(const table& data) : impl_(new compute_input_impl(data)) {} + +template +const table& compute_input::get_data() const { + return impl_->data; +} + +template +void compute_input::set_data_impl(const table& value) { + impl_->data = value; +} + +template class ONEDAL_EXPORT compute_input; + +template +compute_result::compute_result() : impl_(new compute_result_impl{}) {} + +template +bool compute_result::get_finite() const { + return impl_->finite; +} + +template +void compute_result::set_finite_impl(const bool& value) { + impl_->finite = value; +} + +template class ONEDAL_EXPORT compute_result; + +} // namespace v1 +} // namespace oneapi::dal::finiteness_checker diff --git a/cpp/oneapi/dal/algo/finiteness_checker/compute_types.hpp b/cpp/oneapi/dal/algo/finiteness_checker/compute_types.hpp new file mode 100644 index 00000000000..30091328498 --- /dev/null +++ b/cpp/oneapi/dal/algo/finiteness_checker/compute_types.hpp @@ -0,0 +1,98 @@ +/******************************************************************************* +* Copyright contributors to the oneDAL project +* +* Licensed under the Apache License, Version 2.0 (the "License"); +* you may not use this file except in compliance with the License. +* You may obtain a copy of the License at +* +* http://www.apache.org/licenses/LICENSE-2.0 +* +* Unless required by applicable law or agreed to in writing, software +* distributed under the License is distributed on an "AS IS" BASIS, +* WITHOUT WARRANTIES OR CONDITIONS OF ANY KIND, either express or implied. +* See the License for the specific language governing permissions and +* limitations under the License. +*******************************************************************************/ + +#pragma once + +#include "oneapi/dal/algo/finiteness_checker/common.hpp" + +namespace oneapi::dal::finiteness_checker { + +namespace detail { +namespace v1 { +template +class compute_input_impl; + +template +class compute_result_impl; +} // namespace v1 + +using v1::compute_input_impl; +using v1::compute_result_impl; + +} // namespace detail + +namespace v1 { + +/// @tparam Task Tag-type that specifies the type of the problem to solve. Can +/// be :expr:`task::compute`. +template +class compute_input : public base { + static_assert(detail::is_valid_task_v); + +public: + using task_t = Task; + + /// Creates a new instance of the class with the given :literal:`data`. + compute_input(const table& data); + + /// @remark default = table{} + const table& get_data() const; + + auto& set_data(const table& data) { + set_data_impl(data); + return *this; + } + +protected: + void set_data_impl(const table& data); + +private: + dal::detail::pimpl> impl_; +}; + +/// @tparam Task Tag-type that specifies the type of the problem to solve. Can +/// be :expr:`task::compute`. +template +class compute_result : public base { + static_assert(detail::is_valid_task_v); + +public: + using task_t = Task; + + /// Creates a new instance of the class with the default property values. + compute_result(); + + /// A boolean with the result finiteness. + bool get_finite() const; + + auto& set_finite(const bool& value) { + set_finite_impl(value); + return *this; + } + +protected: + void set_finite_impl(const bool&); + +private: + dal::detail::pimpl> impl_; +}; + +} // namespace v1 + +using v1::compute_input; +using v1::compute_result; + +} // namespace oneapi::dal::finiteness_checker diff --git a/cpp/oneapi/dal/algo/finiteness_checker/detail/compute_ops.cpp b/cpp/oneapi/dal/algo/finiteness_checker/detail/compute_ops.cpp new file mode 100644 index 00000000000..03822b8ca9e --- /dev/null +++ b/cpp/oneapi/dal/algo/finiteness_checker/detail/compute_ops.cpp @@ -0,0 +1,44 @@ +/******************************************************************************* +* Copyright contributors to the oneDAL project +* +* Licensed under the Apache License, Version 2.0 (the "License"); +* you may not use this file except in compliance with the License. +* You may obtain a copy of the License at +* +* http://www.apache.org/licenses/LICENSE-2.0 +* +* Unless required by applicable law or agreed to in writing, software +* distributed under the License is distributed on an "AS IS" BASIS, +* WITHOUT WARRANTIES OR CONDITIONS OF ANY KIND, either express or implied. +* See the License for the specific language governing permissions and +* limitations under the License. +*******************************************************************************/ + +#include "oneapi/dal/algo/finiteness_checker/detail/compute_ops.hpp" +#include "oneapi/dal/algo/finiteness_checker/backend/cpu/compute_kernel.hpp" +#include "oneapi/dal/backend/dispatcher.hpp" + +namespace oneapi::dal::finiteness_checker::detail { +namespace v1 { + +using dal::detail::host_policy; + +template +struct compute_ops_dispatcher { + compute_result operator()(const host_policy& ctx, + const descriptor_base& desc, + const compute_input& input) const { + using kernel_dispatcher_t = dal::backend::kernel_dispatcher< // + KERNEL_SINGLE_NODE_CPU(backend::compute_kernel_cpu)>; + return kernel_dispatcher_t()(ctx, desc, input); + } +}; + +#define INSTANTIATE(F, M, T) \ + template struct ONEDAL_EXPORT compute_ops_dispatcher; + +INSTANTIATE(float, method::dense, task::compute) +INSTANTIATE(double, method::dense, task::compute) + +} // namespace v1 +} // namespace oneapi::dal::finiteness_checker::detail diff --git a/cpp/oneapi/dal/algo/finiteness_checker/detail/compute_ops.hpp b/cpp/oneapi/dal/algo/finiteness_checker/detail/compute_ops.hpp new file mode 100644 index 00000000000..a974d9dc57f --- /dev/null +++ b/cpp/oneapi/dal/algo/finiteness_checker/detail/compute_ops.hpp @@ -0,0 +1,77 @@ +/******************************************************************************* +* Copyright contributors to the oneDAL project +* +* Licensed under the Apache License, Version 2.0 (the "License"); +* you may not use this file except in compliance with the License. +* You may obtain a copy of the License at +* +* http://www.apache.org/licenses/LICENSE-2.0 +* +* Unless required by applicable law or agreed to in writing, software +* distributed under the License is distributed on an "AS IS" BASIS, +* WITHOUT WARRANTIES OR CONDITIONS OF ANY KIND, either express or implied. +* See the License for the specific language governing permissions and +* limitations under the License. +*******************************************************************************/ + +#pragma once + +#include "oneapi/dal/algo/finiteness_checker/compute_types.hpp" +#include "oneapi/dal/detail/error_messages.hpp" +#include "oneapi/dal/table/homogen.hpp" + +namespace oneapi::dal::finiteness_checker::detail { +namespace v1 { + +template +struct compute_ops_dispatcher { + compute_result operator()(const Context&, + const descriptor_base& desc, + const compute_input&) const; + +#ifdef ONEDAL_DATA_PARALLEL + void operator()(const Context&, + const descriptor_base& desc, + const table& data, + const bool&); +#endif +}; + +template +struct compute_ops { + using float_t = typename Descriptor::float_t; + using method_t = typename Descriptor::method_t; + using task_t = typename Descriptor::task_t; + using input_t = compute_input; + using result_t = compute_result; + using descriptor_base_t = descriptor_base; + + void check_preconditions(const Descriptor& params, const input_t& input) const { + using msg = dal::detail::error_messages; + + if (!input.get_data().has_data()) { + throw domain_error(msg::input_data_is_empty()); + } + } + + template + auto operator()(const Context& ctx, const Descriptor& desc, const input_t& input) const { + check_preconditions(desc, input); + const auto result = + compute_ops_dispatcher()(ctx, desc, input); + return result; + } + +#ifdef ONEDAL_DATA_PARALLEL + template + void operator()(const Context& ctx, const Descriptor& desc, const table& data, bool& res) { + compute_ops_dispatcher()(ctx, desc, data, res); + } +#endif +}; + +} // namespace v1 + +using v1::compute_ops; + +} // namespace oneapi::dal::finiteness_checker::detail diff --git a/cpp/oneapi/dal/algo/finiteness_checker/detail/compute_ops_dpc.cpp b/cpp/oneapi/dal/algo/finiteness_checker/detail/compute_ops_dpc.cpp new file mode 100644 index 00000000000..6e2b210043d --- /dev/null +++ b/cpp/oneapi/dal/algo/finiteness_checker/detail/compute_ops_dpc.cpp @@ -0,0 +1,58 @@ +/******************************************************************************* +* Copyright contributors to the oneDAL project +* +* Licensed under the Apache License, Version 2.0 (the "License"); +* you may not use this file except in compliance with the License. +* You may obtain a copy of the License at +* +* http://www.apache.org/licenses/LICENSE-2.0 +* +* Unless required by applicable law or agreed to in writing, software +* distributed under the License is distributed on an "AS IS" BASIS, +* WITHOUT WARRANTIES OR CONDITIONS OF ANY KIND, either express or implied. +* See the License for the specific language governing permissions and +* limitations under the License. +*******************************************************************************/ + +#include "oneapi/dal/algo/finiteness_checker/backend/cpu/compute_kernel.hpp" +#include "oneapi/dal/algo/finiteness_checker/backend/gpu/compute_kernel.hpp" +#include "oneapi/dal/algo/finiteness_checker/detail/compute_ops.hpp" +#include "oneapi/dal/backend/dispatcher.hpp" + +namespace oneapi::dal::finiteness_checker::detail { +namespace v1 { + +using dal::detail::data_parallel_policy; + +template +struct compute_ops_dispatcher { + compute_result operator()(const data_parallel_policy& ctx, + const descriptor_base& params, + const compute_input& input) const { + using kernel_dispatcher_t = dal::backend::kernel_dispatcher< + KERNEL_SINGLE_NODE_CPU(backend::compute_kernel_cpu), + KERNEL_SINGLE_NODE_GPU(backend::compute_kernel_gpu)>; + return kernel_dispatcher_t{}(ctx, params, input); + } + +#ifdef ONEDAL_DATA_PARALLEL + void operator()(const data_parallel_policy& ctx, + const descriptor_base& params, + const table& data, + bool& res) { + using kernel_dispatcher_t = dal::backend::kernel_dispatcher< + KERNEL_SINGLE_NODE_CPU(backend::compute_kernel_cpu), + KERNEL_SINGLE_NODE_GPU(backend::compute_kernel_gpu)>; + kernel_dispatcher_t{}(ctx, params, data, res); + } +#endif +}; + +#define INSTANTIATE(F, M, T) \ + template struct ONEDAL_EXPORT compute_ops_dispatcher; + +INSTANTIATE(float, method::dense, task::compute) +INSTANTIATE(double, method::dense, task::compute) + +} // namespace v1 +} // namespace oneapi::dal::finiteness_checker::detail diff --git a/cpp/oneapi/dal/algo/finiteness_checker/test/batch.cpp b/cpp/oneapi/dal/algo/finiteness_checker/test/batch.cpp new file mode 100644 index 00000000000..e99b65cda38 --- /dev/null +++ b/cpp/oneapi/dal/algo/finiteness_checker/test/batch.cpp @@ -0,0 +1,104 @@ +/******************************************************************************* +* Copyright contributors to the oneDAL project +* +* Licensed under the Apache License, Version 2.0 (the "License"); +* you may not use this file except in compliance with the License. +* You may obtain a copy of the License at +* +* http://www.apache.org/licenses/LICENSE-2.0 +* +* Unless required by applicable law or agreed to in writing, software +* distributed under the License is distributed on an "AS IS" BASIS, +* WITHOUT WARRANTIES OR CONDITIONS OF ANY KIND, either express or implied. +* See the License for the specific language governing permissions and +* limitations under the License. +*******************************************************************************/ + +#include +#include + +#include "oneapi/dal/algo/finiteness_checker/compute.hpp" + +#include "oneapi/dal/test/engine/fixtures.hpp" +#include "oneapi/dal/test/engine/math.hpp" + +namespace oneapi::dal::finiteness_checker::test { + +namespace te = dal::test::engine; + +template +class finite_checker_batch_test : public te::float_algo_fixture> { +public: + using Float = std::tuple_element_t<0, TestType>; + using Method = std::tuple_element_t<1, TestType>; + + void check_finiteness(const te::dataframe& x_data, + bool allowNaN, + double value, + const te::table_id& x_data_table_id) { + const table x = x_data.get_table(this->get_policy(), x_data_table_id); + + INFO("create descriptor"); + const auto finiteness_desc = + finiteness_checker::descriptor{}.set_allow_NaN(allowNaN); + + INFO("run compute"); + const bool compute_result = this->compute(finiteness_desc, x).get_finite(); + if (compute_result == (std::isinf(value) || (std::isnan(value) && !allowNaN))) { + CAPTURE(compute_result, value, allowNaN); + FAIL(); + } + SUCCEED(); + } +}; + +using finiteness_types = COMBINE_TYPES((float, double), (finiteness_checker::method::dense)); + +TEMPLATE_LIST_TEST_M(finite_checker_batch_test, + "finiteness checker typical", + "[finiteness_checker][integration][batch]", + finiteness_types) { + SKIP_IF(this->not_float64_friendly()); + + // Initialize values + const te::dataframe x_data = + GENERATE_DATAFRAME(te::dataframe_builder{ 50, 50 }.fill_normal(0, 1, 7777), + te::dataframe_builder{ 100, 50 }.fill_normal(0, 1, 7777), + te::dataframe_builder{ 250, 50 }.fill_normal(0, 1, 7777), + te::dataframe_builder{ 1100, 50 }.fill_normal(0, 1, 7777)); + auto x_data_mutable = x_data.get_array().get_mutable_data(); + const double value = GENERATE(0.0, + -std::numeric_limits::infinity(), + std::numeric_limits::infinity(), + std::numeric_limits::quiet_NaN()); + const bool allowNaN = GENERATE(0, 1); + x_data_mutable[45] = value; + + // Homogen floating point type is the same as algorithm's floating point type + const auto x_data_table_id = this->get_homogen_table_id(); + + this->check_finiteness(x_data, allowNaN, value, x_data_table_id); +} + +TEMPLATE_LIST_TEST_M(finite_checker_batch_test, + "finiteness_checker compute one element matrix", + "[finiteness_checker][integration][batch]", + finiteness_types) { + SKIP_IF(this->not_float64_friendly()); + + // Initialize values to doubles + const double value = GENERATE(0.0, + -std::numeric_limits::infinity(), + std::numeric_limits::infinity(), + std::numeric_limits::quiet_NaN()); + const bool allowNaN = GENERATE(0, 1); + + const te::dataframe x_data = GENERATE_DATAFRAME(te::dataframe_builder{ 1, 1 }.fill(value)); + + // Homogen floating point type is the same as algorithm's floating point type + const auto x_data_table_id = this->get_homogen_table_id(); + + this->check_finiteness(x_data, allowNaN, value, x_data_table_id); +} + +} // namespace oneapi::dal::finiteness_checker::test diff --git a/cpp/oneapi/dal/algo/linear_regression/backend/gpu/finalize_train_kernel_norm_eq_dpc.cpp b/cpp/oneapi/dal/algo/linear_regression/backend/gpu/finalize_train_kernel_norm_eq_dpc.cpp index d3431663249..a74723e1b00 100644 --- a/cpp/oneapi/dal/algo/linear_regression/backend/gpu/finalize_train_kernel_norm_eq_dpc.cpp +++ b/cpp/oneapi/dal/algo/linear_regression/backend/gpu/finalize_train_kernel_norm_eq_dpc.cpp @@ -14,129 +14,32 @@ * limitations under the License. *******************************************************************************/ -#include "oneapi/dal/detail/common.hpp" -#include "oneapi/dal/backend/dispatcher.hpp" -#include "oneapi/dal/backend/primitives/ndarray.hpp" -#include "oneapi/dal/backend/primitives/lapack.hpp" -#include "oneapi/dal/backend/primitives/utils.hpp" - -#include "oneapi/dal/table/row_accessor.hpp" - -#include "oneapi/dal/algo/linear_regression/common.hpp" -#include "oneapi/dal/algo/linear_regression/train_types.hpp" -#include "oneapi/dal/algo/linear_regression/backend/model_impl.hpp" #include "oneapi/dal/algo/linear_regression/backend/gpu/finalize_train_kernel.hpp" -#include "oneapi/dal/algo/linear_regression/backend/gpu/update_kernel.hpp" -#include "oneapi/dal/algo/linear_regression/backend/gpu/misc.hpp" - -namespace oneapi::dal::linear_regression::backend { - -using dal::backend::context_gpu; - -namespace be = dal::backend; -namespace pr = be::primitives; - -template -static train_result call_dal_kernel(const context_gpu& ctx, - const detail::descriptor_base& desc, - const detail::train_parameters& params, - const partial_train_result& input) { - using dal::detail::check_mul_overflow; - - using model_t = model; - using model_impl_t = detail::model_impl; - - auto& queue = ctx.get_queue(); - - const bool compute_intercept = desc.get_compute_intercept(); - - constexpr auto uplo = pr::mkl::uplo::upper; - constexpr auto alloc = sycl::usm::alloc::device; - - const auto response_count = input.get_partial_xty().get_row_count(); - const auto ext_feature_count = input.get_partial_xty().get_column_count(); - const auto feature_count = ext_feature_count - compute_intercept; - - const pr::ndshape<2> xtx_shape{ ext_feature_count, ext_feature_count }; - - const auto xtx_nd = - pr::table2ndarray(queue, input.get_partial_xtx(), sycl::usm::alloc::device); - const auto xty_nd = pr::table2ndarray(queue, - input.get_partial_xty(), - sycl::usm::alloc::device); - - const pr::ndshape<2> betas_shape{ response_count, feature_count + 1 }; +#include "oneapi/dal/algo/linear_regression/backend/gpu/finalize_train_kernel_norm_eq_impl.hpp" - const auto betas_size = check_mul_overflow(response_count, feature_count + 1); - auto betas_arr = array::zeros(queue, betas_size, alloc); - - double alpha = desc.get_alpha(); - sycl::event ridge_event; - if (alpha != 0.0) { - ridge_event = add_ridge_penalty(queue, xtx_nd, compute_intercept, alpha); - } - - auto nxtx = pr::ndarray::empty(queue, xtx_shape, alloc); - auto nxty = pr::ndview::wrap_mutable(betas_arr, betas_shape); - auto solve_event = pr::solve_system(queue, - compute_intercept, - xtx_nd, - xty_nd, - nxtx, - nxty, - { ridge_event }); - sycl::event::wait_and_throw({ solve_event }); - - auto betas = homogen_table::wrap(betas_arr, response_count, feature_count + 1); - - const auto model_impl = std::make_shared(betas); - const auto model = dal::detail::make_private(model_impl); - - const auto options = desc.get_result_options(); - auto result = train_result().set_model(model).set_result_options(options); - - if (options.test(result_options::intercept)) { - auto arr = array::zeros(queue, response_count, alloc); - auto dst = pr::ndview::wrap_mutable(arr, { 1l, response_count }); - const auto src = nxty.get_col_slice(0l, 1l).t(); - - pr::copy(queue, dst, src).wait_and_throw(); - - auto intercept = homogen_table::wrap(arr, 1l, response_count); - result.set_intercept(intercept); - } - - if (options.test(result_options::coefficients)) { - const auto size = check_mul_overflow(response_count, feature_count); - - auto arr = array::zeros(queue, size, alloc); - const auto src = nxty.get_col_slice(1l, feature_count + 1); - auto dst = pr::ndview::wrap_mutable(arr, { response_count, feature_count }); +#include "oneapi/dal/detail/common.hpp" - pr::copy(queue, dst, src).wait_and_throw(); +#include "oneapi/dal/backend/dispatcher.hpp" - auto coefficients = homogen_table::wrap(arr, response_count, feature_count); - result.set_coefficients(coefficients); - } +namespace oneapi::dal::linear_regression::backend { - return result; -} +namespace bk = dal::backend; template -static train_result train(const context_gpu& ctx, - const detail::descriptor_base& desc, - const detail::train_parameters& params, - const partial_train_result& input) { - return call_dal_kernel(ctx, desc, params, input); +static train_result finalize_train(const bk::context_gpu& ctx, + const detail::descriptor_base& desc, + const detail::train_parameters& params, + const partial_train_result& input) { + return finalize_train_kernel_norm_eq_impl(ctx)(desc, params, input); } template struct finalize_train_kernel_gpu { - train_result operator()(const context_gpu& ctx, + train_result operator()(const bk::context_gpu& ctx, const detail::descriptor_base& desc, const detail::train_parameters& params, const partial_train_result& input) const { - return train(ctx, desc, params, input); + return finalize_train(ctx, desc, params, input); } }; diff --git a/cpp/oneapi/dal/algo/linear_regression/backend/gpu/finalize_train_kernel_norm_eq_impl.hpp b/cpp/oneapi/dal/algo/linear_regression/backend/gpu/finalize_train_kernel_norm_eq_impl.hpp new file mode 100644 index 00000000000..6eeaf17c0da --- /dev/null +++ b/cpp/oneapi/dal/algo/linear_regression/backend/gpu/finalize_train_kernel_norm_eq_impl.hpp @@ -0,0 +1,51 @@ +/******************************************************************************* +* Copyright contributors to the oneDAL project +* +* Licensed under the Apache License, Version 2.0 (the "License"); +* you may not use this file except in compliance with the License. +* You may obtain a copy of the License at +* +* http://www.apache.org/licenses/LICENSE-2.0 +* +* Unless required by applicable law or agreed to in writing, software +* distributed under the License is distributed on an "AS IS" BASIS, +* WITHOUT WARRANTIES OR CONDITIONS OF ANY KIND, either express or implied. +* See the License for the specific language governing permissions and +* limitations under the License. +*******************************************************************************/ + +#pragma once + +#include "oneapi/dal/algo/linear_regression/backend/gpu/finalize_train_kernel.hpp" +#include "oneapi/dal/backend/primitives/utils.hpp" + +#ifdef ONEDAL_DATA_PARALLEL + +namespace oneapi::dal::linear_regression::backend { + +namespace bk = dal::backend; + +template +class finalize_train_kernel_norm_eq_impl { + using comm_t = bk::communicator; + using input_t = partial_train_result; + using result_t = train_result; + using descriptor_t = detail::descriptor_base; + using train_parameters_t = detail::train_parameters; + +public: + finalize_train_kernel_norm_eq_impl(const bk::context_gpu& ctx) + : q(ctx.get_queue()), + comm_(ctx.get_communicator()) {} + result_t operator()(const descriptor_t& desc, + const train_parameters_t& params, + const input_t& input); + +private: + sycl::queue q; + comm_t comm_; +}; + +} // namespace oneapi::dal::linear_regression::backend + +#endif // ONEDAL_DATA_PARALLEL diff --git a/cpp/oneapi/dal/algo/linear_regression/backend/gpu/finalize_train_kernel_norm_eq_impl_dpc.cpp b/cpp/oneapi/dal/algo/linear_regression/backend/gpu/finalize_train_kernel_norm_eq_impl_dpc.cpp new file mode 100644 index 00000000000..c470f45403e --- /dev/null +++ b/cpp/oneapi/dal/algo/linear_regression/backend/gpu/finalize_train_kernel_norm_eq_impl_dpc.cpp @@ -0,0 +1,127 @@ +/******************************************************************************* +* Copyright contributors to the oneDAL project +* +* Licensed under the Apache License, Version 2.0 (the "License"); +* you may not use this file except in compliance with the License. +* You may obtain a copy of the License at +* +* http://www.apache.org/licenses/LICENSE-2.0 +* +* Unless required by applicable law or agreed to in writing, software +* distributed under the License is distributed on an "AS IS" BASIS, +* WITHOUT WARRANTIES OR CONDITIONS OF ANY KIND, either express or implied. +* See the License for the specific language governing permissions and +* limitations under the License. +*******************************************************************************/ + +#include "oneapi/dal/algo/linear_regression/backend/gpu/finalize_train_kernel_norm_eq_impl.hpp" +#include "oneapi/dal/algo/linear_regression/backend/gpu/misc.hpp" +#include "oneapi/dal/algo/linear_regression/backend/model_impl.hpp" + +#include "oneapi/dal/backend/primitives/lapack.hpp" + +namespace oneapi::dal::linear_regression::backend { + +namespace be = dal::backend; +namespace pr = be::primitives; + +using be::context_gpu; + +template +train_result finalize_train_kernel_norm_eq_impl::operator()( + const detail::descriptor_base& desc, + const detail::train_parameters& params, + const partial_train_result& input) { + using dal::detail::check_mul_overflow; + + using model_t = model; + using model_impl_t = detail::model_impl; + + const bool compute_intercept = desc.get_compute_intercept(); + + constexpr auto uplo = pr::mkl::uplo::upper; + constexpr auto alloc = sycl::usm::alloc::device; + + const auto response_count = input.get_partial_xty().get_row_count(); + const auto ext_feature_count = input.get_partial_xty().get_column_count(); + const auto feature_count = ext_feature_count - compute_intercept; + + const pr::ndshape<2> xtx_shape{ ext_feature_count, ext_feature_count }; + + const auto xtx_nd = + pr::table2ndarray(q, input.get_partial_xtx(), sycl::usm::alloc::device); + const auto xty_nd = pr::table2ndarray(q, + input.get_partial_xty(), + sycl::usm::alloc::device); + + const pr::ndshape<2> betas_shape{ response_count, feature_count + 1 }; + + const auto betas_size = check_mul_overflow(response_count, feature_count + 1); + auto betas_arr = array::zeros(q, betas_size, alloc); + + if (comm_.get_rank_count() > 1) { + { + ONEDAL_PROFILER_TASK(xtx_allreduce); + auto xtx_arr = + dal::array::wrap(q, xtx_nd.get_mutable_data(), xtx_nd.get_count()); + comm_.allreduce(xtx_arr).wait(); + } + { + ONEDAL_PROFILER_TASK(xty_allreduce); + auto xty_arr = + dal::array::wrap(q, xty_nd.get_mutable_data(), xty_nd.get_count()); + comm_.allreduce(xty_arr).wait(); + } + } + + double alpha = desc.get_alpha(); + sycl::event ridge_event; + if (alpha != 0.0) { + ridge_event = add_ridge_penalty(q, xtx_nd, compute_intercept, alpha); + } + + auto nxtx = pr::ndarray::empty(q, xtx_shape, alloc); + auto nxty = pr::ndview::wrap_mutable(betas_arr, betas_shape); + auto solve_event = + pr::solve_system(q, compute_intercept, xtx_nd, xty_nd, nxtx, nxty, { ridge_event }); + sycl::event::wait_and_throw({ solve_event }); + + auto betas = homogen_table::wrap(betas_arr, response_count, feature_count + 1); + + const auto model_impl = std::make_shared(betas); + const auto model = dal::detail::make_private(model_impl); + + const auto options = desc.get_result_options(); + auto result = train_result().set_model(model).set_result_options(options); + + if (options.test(result_options::intercept)) { + auto arr = array::zeros(q, response_count, alloc); + auto dst = pr::ndview::wrap_mutable(arr, { 1l, response_count }); + const auto src = nxty.get_col_slice(0l, 1l).t(); + + pr::copy(q, dst, src).wait_and_throw(); + + auto intercept = homogen_table::wrap(arr, 1l, response_count); + result.set_intercept(intercept); + } + + if (options.test(result_options::coefficients)) { + const auto size = check_mul_overflow(response_count, feature_count); + + auto arr = array::zeros(q, size, alloc); + const auto src = nxty.get_col_slice(1l, feature_count + 1); + auto dst = pr::ndview::wrap_mutable(arr, { response_count, feature_count }); + + pr::copy(q, dst, src).wait_and_throw(); + + auto coefficients = homogen_table::wrap(arr, response_count, feature_count); + result.set_coefficients(coefficients); + } + + return result; +} + +template class finalize_train_kernel_norm_eq_impl; +template class finalize_train_kernel_norm_eq_impl; + +} // namespace oneapi::dal::linear_regression::backend diff --git a/cpp/oneapi/dal/algo/linear_regression/backend/gpu/misc.hpp b/cpp/oneapi/dal/algo/linear_regression/backend/gpu/misc.hpp index 5ad5ba647ec..723fde68fb9 100644 --- a/cpp/oneapi/dal/algo/linear_regression/backend/gpu/misc.hpp +++ b/cpp/oneapi/dal/algo/linear_regression/backend/gpu/misc.hpp @@ -44,7 +44,7 @@ sycl::event add_ridge_penalty(sycl::queue& q, Float alpha, const bk::event_vector& deps = {}) { ONEDAL_ASSERT(xtx.has_mutable_data()); - ONEDAL_ASSERT(be::is_known_usm(q, xtx.get_mutable_data())); + ONEDAL_ASSERT(bk::is_known_usm(q, xtx.get_mutable_data())); ONEDAL_ASSERT(xtx.get_dimension(0) == xtx.get_dimension(1)); Float* xtx_ptr = xtx.get_mutable_data(); @@ -52,7 +52,7 @@ sycl::event add_ridge_penalty(sycl::queue& q, std::int64_t original_feature_count = feature_count - compute_intercept; return q.submit([&](sycl::handler& cgh) { - const auto range = be::make_range_1d(original_feature_count); + const auto range = bk::make_range_1d(original_feature_count); cgh.depends_on(deps); std::int64_t step = feature_count + 1; cgh.parallel_for(range, [=](sycl::id<1> idx) { diff --git a/cpp/oneapi/dal/algo/linear_regression/backend/gpu/train_kernel_norm_eq_dpc.cpp b/cpp/oneapi/dal/algo/linear_regression/backend/gpu/train_kernel_norm_eq_dpc.cpp index 25b08aa7710..04d76fe86b7 100644 --- a/cpp/oneapi/dal/algo/linear_regression/backend/gpu/train_kernel_norm_eq_dpc.cpp +++ b/cpp/oneapi/dal/algo/linear_regression/backend/gpu/train_kernel_norm_eq_dpc.cpp @@ -104,17 +104,9 @@ static train_result call_dal_kernel(const context_gpu& ctx, old_x_arr = std::move(x_arr), old_y_arr = std::move(y_arr); } - const be::event_vector solve_deps{ last_xty_event, last_xtx_event }; - - double alpha = desc.get_alpha(); - if (alpha != 0.0) { - last_xtx_event = - add_ridge_penalty(queue, xtx, compute_intercept, alpha, { last_xtx_event }); - } - auto& comm = ctx.get_communicator(); if (comm.get_rank_count() > 1) { - sycl::event::wait_and_throw(solve_deps); + sycl::event::wait_and_throw({ last_xty_event, last_xtx_event }); { ONEDAL_PROFILER_TASK(xtx_allreduce); auto xtx_arr = dal::array::wrap(queue, xtx.get_mutable_data(), xtx.get_count()); @@ -127,6 +119,13 @@ static train_result call_dal_kernel(const context_gpu& ctx, } } + double alpha = desc.get_alpha(); + if (alpha != 0.0) { + last_xtx_event = + add_ridge_penalty(queue, xtx, compute_intercept, alpha, { last_xtx_event }); + } + const be::event_vector solve_deps{ last_xty_event, last_xtx_event }; + auto nxtx = pr::ndarray::empty(queue, xtx_shape, alloc); auto nxty = pr::ndview::wrap_mutable(betas_arr, betas_shape); auto solve_event = diff --git a/cpp/oneapi/dal/algo/linear_regression/detail/finalize_train_ops_dpc.cpp b/cpp/oneapi/dal/algo/linear_regression/detail/finalize_train_ops_dpc.cpp index 3592aeefccb..21a5ce8108d 100644 --- a/cpp/oneapi/dal/algo/linear_regression/detail/finalize_train_ops_dpc.cpp +++ b/cpp/oneapi/dal/algo/linear_regression/detail/finalize_train_ops_dpc.cpp @@ -38,7 +38,7 @@ struct finalize_train_ops_dispatcher { const partial_train_result& input) const { using kernel_dispatcher_t = dal::backend::kernel_dispatcher< KERNEL_SINGLE_NODE_CPU(parameters::train_parameters_cpu), - KERNEL_SINGLE_NODE_GPU(parameters::train_parameters_gpu)>; + KERNEL_UNIVERSAL_SPMD_GPU(parameters::train_parameters_gpu)>; return kernel_dispatcher_t{}(ctx, desc, input); } @@ -56,14 +56,16 @@ struct finalize_train_ops_dispatcher { const partial_train_result& input) const { using kernel_dispatcher_t = dal::backend::kernel_dispatcher< KERNEL_SINGLE_NODE_CPU(backend::finalize_train_kernel_cpu), - KERNEL_SINGLE_NODE_GPU(backend::finalize_train_kernel_gpu)>; + KERNEL_UNIVERSAL_SPMD_GPU(backend::finalize_train_kernel_gpu)>; return kernel_dispatcher_t{}(ctx, desc, params, input); } }; -#define INSTANTIATE(F, M, T) \ - template struct ONEDAL_EXPORT \ - finalize_train_ops_dispatcher; +#define INSTANTIATE(F, M, T) \ + template struct ONEDAL_EXPORT \ + finalize_train_ops_dispatcher; \ + template struct ONEDAL_EXPORT \ + finalize_train_ops_dispatcher; INSTANTIATE(float, method::norm_eq, task::regression) INSTANTIATE(double, method::norm_eq, task::regression) diff --git a/cpp/oneapi/dal/algo/linear_regression/test/fixture.hpp b/cpp/oneapi/dal/algo/linear_regression/test/fixture.hpp index aedf0165454..fb935174cfe 100644 --- a/cpp/oneapi/dal/algo/linear_regression/test/fixture.hpp +++ b/cpp/oneapi/dal/algo/linear_regression/test/fixture.hpp @@ -54,6 +54,9 @@ class lr_test : public te::crtp_algo_fixture { using test_input_t = infer_input; using test_result_t = infer_result; + using partial_input_t = partial_train_input<>; + using partial_result_t = partial_train_result<>; + te::table_id get_homogen_table_id() const { return te::table_id::homogen(); } diff --git a/cpp/oneapi/dal/algo/linear_regression/test/online_spmd.cpp b/cpp/oneapi/dal/algo/linear_regression/test/online_spmd.cpp new file mode 100644 index 00000000000..c0f7968adfc --- /dev/null +++ b/cpp/oneapi/dal/algo/linear_regression/test/online_spmd.cpp @@ -0,0 +1,126 @@ +/******************************************************************************* +* Copyright contributors to the oneDAL project +* +* Licensed under the Apache License, Version 2.0 (the "License"); +* you may not use this file except in compliance with the License. +* You may obtain a copy of the License at +* +* http://www.apache.org/licenses/LICENSE-2.0 +* +* Unless required by applicable law or agreed to in writing, software +* distributed under the License is distributed on an "AS IS" BASIS, +* WITHOUT WARRANTIES OR CONDITIONS OF ANY KIND, either express or implied. +* See the License for the specific language governing permissions and +* limitations under the License. +*******************************************************************************/ + +#include "oneapi/dal/algo/linear_regression/test/fixture.hpp" +#include "oneapi/dal/test/engine/tables.hpp" +#include "oneapi/dal/test/engine/io.hpp" + +namespace oneapi::dal::linear_regression::test { + +namespace te = dal::test::engine; +namespace la = te::linalg; +namespace linear_regression = oneapi::dal::linear_regression; + +template +class lr_online_spmd_test : public lr_test> { +public: + using base_t = lr_test>; + using float_t = typename base_t::float_t; + using input_t = typename base_t::train_input_t; + using partial_input_t = typename base_t::partial_input_t; + using partial_result_t = typename base_t::partial_result_t; + using result_t = typename base_t::train_result_t; + + void set_rank_count(std::int64_t rank_count) { + n_rank = rank_count; + } + + std::int64_t get_rank_count() { + return n_rank; + } + + void generate_dimensions() { + this->t_count_ = GENERATE(307, 12999); + this->s_count_ = GENERATE(10000); + this->f_count_ = GENERATE(2, 17); + this->r_count_ = GENERATE(2, 15); + this->intercept_ = GENERATE(0, 1); + } + + template + result_t finalize_train_override(Args&&... args) { + return this->finalize_train_via_spmd_threads_and_merge(n_rank, std::forward(args)...); + } + + result_t merge_finalize_train_result_override(const std::vector& results) { + return results[0]; + } + + template + std::vector split_finalize_train_input_override(std::int64_t split_count, + Args&&... args) { + ONEDAL_ASSERT(split_count == n_rank); + const std::vector input{ std::forward(args)... }; + + return input; + } + + void run_and_check_linear_online_spmd(std::int64_t n_rank, + std::int64_t n_blocks, + std::int64_t seed = 888, + double tol = 1e-2) { + table x_train, y_train, x_test, y_test; + std::tie(x_train, y_train, x_test, y_test) = this->prepare_inputs(seed, tol); + + const auto desc = this->get_descriptor(); + std::vector partial_results; + auto input_table_x = base_t::template split_table_by_rows(x_train, n_rank); + auto input_table_y = base_t::template split_table_by_rows(y_train, n_rank); + for (int64_t i = 0; i < n_rank; i++) { + partial_result_t partial_result; + auto input_table_x_blocks = + base_t::template split_table_by_rows(input_table_x[i], n_blocks); + auto input_table_y_blocks = + base_t::template split_table_by_rows(input_table_y[i], n_blocks); + for (int64_t j = 0; j < n_blocks; j++) { + partial_result = this->partial_train(desc, + partial_result, + input_table_x_blocks[j], + input_table_y_blocks[j]); + } + partial_results.push_back(partial_result); + } + + const auto train_result = this->finalize_train_override(desc, partial_results); + + SECTION("Checking intercept values") { + if (desc.get_result_options().test(result_options::intercept)) + base_t::check_if_close(train_result.get_intercept(), base_t::bias_, tol); + } + + SECTION("Checking coefficient values") { + if (desc.get_result_options().test(result_options::coefficients)) + base_t::check_if_close(train_result.get_coefficients(), base_t::beta_, tol); + } + } + +private: + std::int64_t n_rank; +}; + +TEMPLATE_LIST_TEST_M(lr_online_spmd_test, "lr common flow", "[lr][integration][spmd]", lr_types) { + SKIP_IF(this->get_policy().is_cpu()); + SKIP_IF(this->not_float64_friendly()); + + this->generate(777); + + this->set_rank_count(GENERATE(1, 2, 4)); + std::int64_t n_blocks = GENERATE(1, 3, 10); + + this->run_and_check_linear_online_spmd(this->get_rank_count(), n_blocks); +} + +} // namespace oneapi::dal::linear_regression::test diff --git a/dev/download_micromkl.bat b/dev/download_micromkl.bat index 6a30a2a44e3..a38515735a5 100755 --- a/dev/download_micromkl.bat +++ b/dev/download_micromkl.bat @@ -20,7 +20,7 @@ powershell.exe -command "if ($PSVersionTable.PSVersion.Major -ge 3) {exit 1} els set MKLURLROOT=https://github.com/oneapi-src/oneDAL/releases/download/Dependencies/ set MKLVERSION=20230413 -set MKLGPUVERSION="2024-02-20" +set MKLGPUVERSION=20240605 set MKLPACKAGE=mklfpk_win_%MKLVERSION% set MKLGPUPACKAGE=mklgpufpk_win_%MKLGPUVERSION% diff --git a/dev/download_micromkl.sh b/dev/download_micromkl.sh index 0aaef938e9a..6eb52ddca76 100755 --- a/dev/download_micromkl.sh +++ b/dev/download_micromkl.sh @@ -18,7 +18,7 @@ MKLFPK_URL_ROOT="https://github.com/oneapi-src/oneDAL/releases/download/Dependencies/" MKLFPK_VERSION="20230413" MKLFPK_VERSION_MAC="20210426" -MKLGPUFPK_VERSION="2024-02-20" +MKLGPUFPK_VERSION="20240605" WITH_GPU=true while true ; do diff --git a/makefile.lst b/makefile.lst index de7afb1090c..92dc52ff521 100755 --- a/makefile.lst +++ b/makefile.lst @@ -228,6 +228,7 @@ ONEAPI.ALGOS := \ dbscan \ decision_forest \ decision_tree \ + finiteness_checker \ kmeans \ kmeans_init \ knn \ diff --git a/samples/cmake/setup_samples.cmake b/samples/cmake/setup_samples.cmake index afaf9042417..dd7d3cc758b 100644 --- a/samples/cmake/setup_samples.cmake +++ b/samples/cmake/setup_samples.cmake @@ -111,7 +111,7 @@ function(add_samples samples_paths) set_target_properties(${sample} PROPERTIES RUNTIME_OUTPUT_DIRECTORY "${PROJECT_SOURCE_DIR}/_cmake_results/${CPU_ARCHITECTURE}_${LINK_TYPE}") add_custom_target(run_${sample} - COMMAND ${MPIEXEC_EXECUTABLE} ${MPIEXEC_NUMPROC_FLAG} \\ + COMMAND ${MPIEXEC_EXECUTABLE} ${MPIEXEC_NUMPROC_FLAG} ${MPIEXEC_MAX_NUMPROCS} -ppn ${MPIEXEC_NUMPROCS_PER_NODE} $ DEPENDS ${sample} WORKING_DIRECTORY ${CMAKE_CURRENT_SOURCE_DIR}