diff --git a/cmake/get_info.cmake b/cmake/get_info.cmake index 94f898a24c4..dcaf07f334b 100644 --- a/cmake/get_info.cmake +++ b/cmake/get_info.cmake @@ -72,6 +72,21 @@ function(ginkgo_print_variable log_type var_name) FILE(APPEND ${log_type} "${upd_string}") endfunction() + +function(ginkgo_print_env_variable log_type var_name) + string(SUBSTRING + " +-- ${var_name}: " 0 55 upd_string) + if(NOT ENV{${var_name}}) + set(str_value "") + else() + set(str_value "$ENV{${var_name}}") + endif() + string(APPEND upd_string "${str_value}") + FILE(APPEND ${log_type} "${upd_string}") +endfunction() + + macro(ginkgo_print_foreach_variable variables) foreach(var ${variables}) ginkgo_print_variable(${${log_type}} ${var} ) diff --git a/core/test/reorder/rcm.cpp b/core/test/reorder/rcm.cpp index 030db177fa3..4d05bb0574f 100644 --- a/core/test/reorder/rcm.cpp +++ b/core/test/reorder/rcm.cpp @@ -1,5 +1,5 @@ /************************************************************* -Copyright (c) 2017-2019, the Ginkgo authors +Copyright (c) 2017-2020, the Ginkgo authors All rights reserved. Redistribution and use in source and binary forms, with or without diff --git a/dpcpp/base/executor.dp.cpp b/dpcpp/base/executor.dp.cpp index 208d7255fb3..7e311d489c7 100644 --- a/dpcpp/base/executor.dp.cpp +++ b/dpcpp/base/executor.dp.cpp @@ -33,6 +33,8 @@ OF THIS SOFTWARE, EVEN IF ADVISED OF THE POSSIBILITY OF SUCH DAMAGE. #include +#include +#include #include #include #include @@ -56,6 +58,8 @@ const std::vector get_devices(std::string device_type) {"all", sycl::info::device_type::all}, {"cpu", sycl::info::device_type::cpu}, {"gpu", sycl::info::device_type::gpu}}; + std::for_each(device_type.begin(), device_type.end(), + [](char &c) { c = std::tolower(c); }); return sycl::device::get_devices(device_type_map.at(device_type)); } @@ -123,7 +127,6 @@ void DpcppExecutor::raw_copy_to(const DpcppExecutor *dest, size_type num_bytes, const void *src_ptr, void *dest_ptr) const { if (num_bytes > 0) { - // TODO: does this work? Or is it needed to go through host? dest->get_queue()->memcpy(dest_ptr, src_ptr, num_bytes).wait(); } } @@ -147,6 +150,9 @@ int DpcppExecutor::get_num_devices(std::string device_type) } +namespace detail { + + void delete_queue(sycl::queue *queue) { queue->wait(); @@ -154,6 +160,9 @@ void delete_queue(sycl::queue *queue) } +} // namespace detail + + void DpcppExecutor::set_device_property() { assert(device_id_ < DpcppExecutor::get_num_devices(device_type_)); @@ -174,8 +183,13 @@ void DpcppExecutor::set_device_property() } max_workgroup_size_ = device.get_info(); + // Here we declare the queue with the property `in_order` which ensures the + // kernels are executed in the submission order. Otherwise, calls to + // `wait()` would be needed after every call to a DPC++ function or kernel. + // For example, without `in_order`, doing a copy, a kernel, and a copy, will + // not necessarily happen in that order by default, which we need to avoid. auto *queue = new sycl::queue{device, sycl::property::queue::in_order{}}; - queue_ = std::move(queue_manager{queue, delete_queue}); + queue_ = std::move(queue_manager{queue, detail::delete_queue}); } diff --git a/dpcpp/components/absolute_array.dp.cpp b/dpcpp/components/absolute_array.dp.cpp index 9603dacbc8b..830752de343 100644 --- a/dpcpp/components/absolute_array.dp.cpp +++ b/dpcpp/components/absolute_array.dp.cpp @@ -51,7 +51,7 @@ void inplace_absolute_array(std::shared_ptr exec, { exec->get_queue()->submit([&](sycl::handler &cgh) { cgh.parallel_for(sycl::range<1>{n}, [=](sycl::id<1> idx_id) { - const int idx = idx_id[0]; + const auto idx = idx_id[0]; data[idx] = dpcpp::abs(data[idx]); }); }); @@ -67,7 +67,7 @@ void outplace_absolute_array(std::shared_ptr exec, { exec->get_queue()->submit([&](sycl::handler &cgh) { cgh.parallel_for(sycl::range<1>{n}, [=](sycl::id<1> idx_id) { - const int idx = idx_id[0]; + const auto idx = idx_id[0]; out[idx] = dpcpp::abs(in[idx]); }); }); diff --git a/dpcpp/components/csr_spgeam.dp.hpp b/dpcpp/components/csr_spgeam.dp.hpp deleted file mode 100644 index 013ee10bbf8..00000000000 --- a/dpcpp/components/csr_spgeam.dp.hpp +++ /dev/null @@ -1,76 +0,0 @@ -/************************************************************* -Copyright (c) 2017-2020, the Ginkgo authors -All rights reserved. - -Redistribution and use in source and binary forms, with or without -modification, are permitted provided that the following conditions -are met: - -1. Redistributions of source code must retain the above copyright -notice, this list of conditions and the following disclaimer. - -2. Redistributions in binary form must reproduce the above copyright -notice, this list of conditions and the following disclaimer in the -documentation and/or other materials provided with the distribution. - -3. Neither the name of the copyright holder nor the names of its -contributors may be used to endorse or promote products derived from -this software without specific prior written permission. - -THIS SOFTWARE IS PROVIDED BY THE COPYRIGHT HOLDERS AND CONTRIBUTORS "AS -IS" AND ANY EXPRESS OR IMPLIED WARRANTIES, INCLUDING, BUT NOT LIMITED -TO, THE IMPLIED WARRANTIES OF MERCHANTABILITY AND FITNESS FOR A -PARTICULAR PURPOSE ARE DISCLAIMED. IN NO EVENT SHALL THE COPYRIGHT -HOLDER OR CONTRIBUTORS BE LIABLE FOR ANY DIRECT, INDIRECT, INCIDENTAL, -SPECIAL, EXEMPLARY, OR CONSEQUENTIAL DAMAGES (INCLUDING, BUT NOT -LIMITED TO, PROCUREMENT OF SUBSTITUTE GOODS OR SERVICES; LOSS OF USE, -DATA, OR PROFITS; OR BUSINESS INTERRUPTION) HOWEVER CAUSED AND ON ANY -THEORY OF LIABILITY, WHETHER IN CONTRACT, STRICT LIABILITY, OR TORT -(INCLUDING NEGLIGENCE OR OTHERWISE) ARISING IN ANY WAY OUT OF THE USE -OF THIS SOFTWARE, EVEN IF ADVISED OF THE POSSIBILITY OF SUCH DAMAGE. -*************************************************************/ - -#ifndef GKO_DPCPP_COMPONENTS_CSR_SPGEAM_DP_HPP_ -#define GKO_DPCPP_COMPONENTS_CSR_SPGEAM_DP_HPP_ - - -#include - - -#include - - -#include "core/base/utils.hpp" - - -namespace gko { -namespace kernels { -namespace dpcpp { - - -/** - * Adds two (sorted) sparse matrices. - * - * Calls begin_cb(row) on each row to initialize row-local data - * Calls entry_cb(row, col, a_val, b_val, local_data) on each output non-zero - * Calls end_cb(row, local_data) on each row to finalize row-local data - * - * If the three functions are thread-safe, the whole invocation is. - */ -template -void abstract_spgeam(const matrix::Csr *a, - const matrix::Csr *b, - BeginCallback begin_cb, EntryCallback entry_cb, - EndCallback end_cb) -{ - GKO_NOT_IMPLEMENTED; -} - - -} // namespace dpcpp -} // namespace kernels -} // namespace gko - - -#endif // GKO_DPCPP_COMPONENTS_CSR_SPGEAM_DP_HPP_ diff --git a/dpcpp/components/fill_array.dp.cpp b/dpcpp/components/fill_array.dp.cpp index f8c260cdab6..98328484be2 100644 --- a/dpcpp/components/fill_array.dp.cpp +++ b/dpcpp/components/fill_array.dp.cpp @@ -48,7 +48,7 @@ void fill_array(std::shared_ptr exec, ValueType *array, { exec->get_queue()->submit([&](sycl::handler &cgh) { cgh.parallel_for(sycl::range<1>{n}, [=](sycl::id<1> idx_id) { - const int idx = idx_id[0]; + const auto idx = idx_id[0]; array[idx] = val; }); }); diff --git a/dpcpp/components/format_conversion.dp.hpp b/dpcpp/components/format_conversion.dp.hpp index f1dba5eed78..0b603e7bcf2 100644 --- a/dpcpp/components/format_conversion.dp.hpp +++ b/dpcpp/components/format_conversion.dp.hpp @@ -56,10 +56,8 @@ namespace dpcpp { template inline void convert_unsorted_idxs_to_ptrs(const IndexType *idxs, size_type num_nonzeros, - IndexType *ptrs, size_type length) -{ - GKO_NOT_IMPLEMENTED; -} + IndexType *ptrs, + size_type length) GKO_NOT_IMPLEMENTED; /** @@ -72,18 +70,12 @@ inline void convert_unsorted_idxs_to_ptrs(const IndexType *idxs, template inline void convert_sorted_idxs_to_ptrs(const IndexType *idxs, size_type num_nonzeros, IndexType *ptrs, - size_type length) -{ - GKO_NOT_IMPLEMENTED; -} + size_type length) GKO_NOT_IMPLEMENTED; template inline void convert_ptrs_to_idxs(const IndexType *ptrs, size_type num_rows, - IndexType *idxs) -{ - GKO_NOT_IMPLEMENTED; -} + IndexType *idxs) GKO_NOT_IMPLEMENTED; } // namespace dpcpp diff --git a/dpcpp/components/matrix_operations.dp.hpp b/dpcpp/components/matrix_operations.dp.hpp index d4cbcd245b8..7c39735f485 100644 --- a/dpcpp/components/matrix_operations.dp.hpp +++ b/dpcpp/components/matrix_operations.dp.hpp @@ -48,13 +48,9 @@ namespace dpcpp { * Computes the infinity norm of a column-major matrix. */ template -remove_complex compute_inf_norm(size_type num_rows, - size_type num_cols, - const ValueType *matrix, - size_type stride) -{ - GKO_NOT_IMPLEMENTED; -} +remove_complex compute_inf_norm( + size_type num_rows, size_type num_cols, const ValueType *matrix, + size_type stride) GKO_NOT_IMPLEMENTED; } // namespace dpcpp diff --git a/dpcpp/components/precision_conversion.dp.cpp b/dpcpp/components/precision_conversion.dp.cpp index c508e853ad6..41f9c66833a 100644 --- a/dpcpp/components/precision_conversion.dp.cpp +++ b/dpcpp/components/precision_conversion.dp.cpp @@ -48,7 +48,7 @@ void convert_precision(std::shared_ptr exec, { exec->get_queue()->submit([&](sycl::handler &cgh) { cgh.parallel_for(sycl::range<1>{size}, [=](sycl::id<1> idx_id) { - const int idx = idx_id[0]; + const auto idx = idx_id[0]; out[idx] = in[idx]; }); }); diff --git a/dpcpp/components/prefix_sum.dp.cpp b/dpcpp/components/prefix_sum.dp.cpp index 21ca279b696..6740a7e966c 100644 --- a/dpcpp/components/prefix_sum.dp.cpp +++ b/dpcpp/components/prefix_sum.dp.cpp @@ -47,13 +47,9 @@ namespace components { template void prefix_sum(std::shared_ptr exec, IndexType *counts, - size_type num_entries) -{ - GKO_NOT_IMPLEMENTED; -} + size_type num_entries) GKO_NOT_IMPLEMENTED; GKO_INSTANTIATE_FOR_EACH_INDEX_TYPE(GKO_DECLARE_PREFIX_SUM_KERNEL); - // instantiate for size_type as well, as this is used in the Sellp format template GKO_DECLARE_PREFIX_SUM_KERNEL(size_type); diff --git a/dpcpp/factorization/factorization_kernels.dp.cpp b/dpcpp/factorization/factorization_kernels.dp.cpp index 74386b58376..5e33d368733 100644 --- a/dpcpp/factorization/factorization_kernels.dp.cpp +++ b/dpcpp/factorization/factorization_kernels.dp.cpp @@ -59,29 +59,21 @@ namespace factorization { template void find_missing_diagonal_elements( const matrix::Csr *mtx, - IndexType *elements_to_add_per_row, bool *changes_required) -{ - GKO_NOT_IMPLEMENTED; -} + IndexType *elements_to_add_per_row, + bool *changes_required) GKO_NOT_IMPLEMENTED; template -void add_missing_diagonal_elements(const matrix::Csr *mtx, - ValueType *new_values, - IndexType *new_col_idxs, - const IndexType *row_ptrs_addition) -{ - GKO_NOT_IMPLEMENTED; -} +void add_missing_diagonal_elements( + const matrix::Csr *mtx, ValueType *new_values, + IndexType *new_col_idxs, + const IndexType *row_ptrs_addition) GKO_NOT_IMPLEMENTED; template void add_diagonal_elements(std::shared_ptr exec, matrix::Csr *mtx, - bool is_sorted) -{ - GKO_NOT_IMPLEMENTED; -} + bool is_sorted) GKO_NOT_IMPLEMENTED; GKO_INSTANTIATE_FOR_EACH_VALUE_AND_INDEX_TYPE( GKO_DECLARE_FACTORIZATION_ADD_DIAGONAL_ELEMENTS_KERNEL); @@ -91,10 +83,7 @@ template void initialize_row_ptrs_l_u( std::shared_ptr exec, const matrix::Csr *system_matrix, - IndexType *l_row_ptrs, IndexType *u_row_ptrs) -{ - GKO_NOT_IMPLEMENTED; -} + IndexType *l_row_ptrs, IndexType *u_row_ptrs) GKO_NOT_IMPLEMENTED; GKO_INSTANTIATE_FOR_EACH_VALUE_AND_INDEX_TYPE( GKO_DECLARE_FACTORIZATION_INITIALIZE_ROW_PTRS_L_U_KERNEL); @@ -105,9 +94,7 @@ void initialize_l_u(std::shared_ptr exec, const matrix::Csr *system_matrix, matrix::Csr *csr_l, matrix::Csr *csr_u) -{ GKO_NOT_IMPLEMENTED; -} GKO_INSTANTIATE_FOR_EACH_VALUE_AND_INDEX_TYPE( GKO_DECLARE_FACTORIZATION_INITIALIZE_L_U_KERNEL); @@ -117,10 +104,7 @@ template void initialize_row_ptrs_l( std::shared_ptr exec, const matrix::Csr *system_matrix, - IndexType *l_row_ptrs) -{ - GKO_NOT_IMPLEMENTED; -} + IndexType *l_row_ptrs) GKO_NOT_IMPLEMENTED; GKO_INSTANTIATE_FOR_EACH_VALUE_AND_INDEX_TYPE( GKO_DECLARE_FACTORIZATION_INITIALIZE_ROW_PTRS_L_KERNEL); @@ -129,10 +113,8 @@ GKO_INSTANTIATE_FOR_EACH_VALUE_AND_INDEX_TYPE( template void initialize_l(std::shared_ptr exec, const matrix::Csr *system_matrix, - matrix::Csr *csr_l, bool diag_sqrt) -{ - GKO_NOT_IMPLEMENTED; -} + matrix::Csr *csr_l, + bool diag_sqrt) GKO_NOT_IMPLEMENTED; GKO_INSTANTIATE_FOR_EACH_VALUE_AND_INDEX_TYPE( GKO_DECLARE_FACTORIZATION_INITIALIZE_L_KERNEL); diff --git a/dpcpp/factorization/par_ict_kernels.dp.cpp b/dpcpp/factorization/par_ict_kernels.dp.cpp index 6e7dd7e5563..c99d4e813cd 100644 --- a/dpcpp/factorization/par_ict_kernels.dp.cpp +++ b/dpcpp/factorization/par_ict_kernels.dp.cpp @@ -48,7 +48,6 @@ OF THIS SOFTWARE, EVEN IF ADVISED OF THE POSSIBILITY OF SUCH DAMAGE. #include "core/base/utils.hpp" #include "core/components/prefix_sum.hpp" #include "core/matrix/csr_builder.hpp" -#include "dpcpp/components/csr_spgeam.dp.hpp" namespace gko { @@ -67,9 +66,7 @@ void compute_factor(std::shared_ptr exec, const matrix::Csr *a, matrix::Csr *l, const matrix::Coo *) -{ GKO_NOT_IMPLEMENTED; -} GKO_INSTANTIATE_FOR_EACH_VALUE_AND_INDEX_TYPE( GKO_DECLARE_PAR_ICT_COMPUTE_FACTOR_KERNEL); @@ -81,9 +78,7 @@ void add_candidates(std::shared_ptr exec, const matrix::Csr *a, const matrix::Csr *l, matrix::Csr *l_new) -{ GKO_NOT_IMPLEMENTED; -} GKO_INSTANTIATE_FOR_EACH_VALUE_AND_INDEX_TYPE( GKO_DECLARE_PAR_ICT_ADD_CANDIDATES_KERNEL); diff --git a/dpcpp/factorization/par_ilu_kernels.dp.cpp b/dpcpp/factorization/par_ilu_kernels.dp.cpp index baa8b4a4481..cd1ae44ce37 100644 --- a/dpcpp/factorization/par_ilu_kernels.dp.cpp +++ b/dpcpp/factorization/par_ilu_kernels.dp.cpp @@ -53,14 +53,11 @@ namespace par_ilu_factorization { template -void compute_l_u_factors(std::shared_ptr exec, - size_type iterations, - const matrix::Coo *system_matrix, - matrix::Csr *l_factor, - matrix::Csr *u_factor) -{ - GKO_NOT_IMPLEMENTED; -} +void compute_l_u_factors( + std::shared_ptr exec, size_type iterations, + const matrix::Coo *system_matrix, + matrix::Csr *l_factor, + matrix::Csr *u_factor) GKO_NOT_IMPLEMENTED; GKO_INSTANTIATE_FOR_EACH_VALUE_AND_INDEX_TYPE( GKO_DECLARE_PAR_ILU_COMPUTE_L_U_FACTORS_KERNEL); diff --git a/dpcpp/factorization/par_ilut_kernels.dp.cpp b/dpcpp/factorization/par_ilut_kernels.dp.cpp index 6b7f5b0ac17..80b4e11c219 100644 --- a/dpcpp/factorization/par_ilut_kernels.dp.cpp +++ b/dpcpp/factorization/par_ilut_kernels.dp.cpp @@ -52,7 +52,6 @@ OF THIS SOFTWARE, EVEN IF ADVISED OF THE POSSIBILITY OF SUCH DAMAGE. #include "core/components/prefix_sum.hpp" #include "core/matrix/coo_builder.hpp" #include "core/matrix/csr_builder.hpp" -#include "dpcpp/components/csr_spgeam.dp.hpp" namespace gko { @@ -71,10 +70,7 @@ void threshold_select(std::shared_ptr exec, const matrix::Csr *m, IndexType rank, Array &tmp, Array> &, - remove_complex &threshold) -{ - GKO_NOT_IMPLEMENTED; -} + remove_complex &threshold) GKO_NOT_IMPLEMENTED; GKO_INSTANTIATE_FOR_EACH_VALUE_AND_INDEX_TYPE( GKO_DECLARE_PAR_ILUT_THRESHOLD_SELECT_KERNEL); @@ -91,10 +87,7 @@ void abstract_filter(std::shared_ptr exec, const matrix::Csr *m, matrix::Csr *m_out, matrix::Coo *m_out_coo, - Predicate pred) -{ - GKO_NOT_IMPLEMENTED; -} + Predicate pred) GKO_NOT_IMPLEMENTED; template @@ -102,10 +95,8 @@ void threshold_filter(std::shared_ptr exec, const matrix::Csr *m, remove_complex threshold, matrix::Csr *m_out, - matrix::Coo *m_out_coo, bool) -{ - GKO_NOT_IMPLEMENTED; -} + matrix::Coo *m_out_coo, + bool) GKO_NOT_IMPLEMENTED; GKO_INSTANTIATE_FOR_EACH_VALUE_AND_INDEX_TYPE( GKO_DECLARE_PAR_ILUT_THRESHOLD_FILTER_KERNEL); @@ -116,15 +107,12 @@ constexpr auto sample_size = bucket_count * sampleselect_oversampling; template -void threshold_filter_approx(std::shared_ptr exec, - const matrix::Csr *m, - IndexType rank, Array &tmp, - remove_complex &threshold, - matrix::Csr *m_out, - matrix::Coo *m_out_coo) -{ - GKO_NOT_IMPLEMENTED; -} +void threshold_filter_approx( + std::shared_ptr exec, + const matrix::Csr *m, IndexType rank, + Array &tmp, remove_complex &threshold, + matrix::Csr *m_out, + matrix::Coo *m_out_coo) GKO_NOT_IMPLEMENTED; GKO_INSTANTIATE_FOR_EACH_VALUE_AND_INDEX_TYPE( GKO_DECLARE_PAR_ILUT_THRESHOLD_FILTER_APPROX_KERNEL); @@ -138,9 +126,7 @@ void compute_l_u_factors(std::shared_ptr exec, matrix::Csr *u, const matrix::Coo *, matrix::Csr *u_csc) -{ GKO_NOT_IMPLEMENTED; -} GKO_INSTANTIATE_FOR_EACH_VALUE_AND_INDEX_TYPE( GKO_DECLARE_PAR_ILUT_COMPUTE_LU_FACTORS_KERNEL); @@ -154,9 +140,7 @@ void add_candidates(std::shared_ptr exec, const matrix::Csr *u, matrix::Csr *l_new, matrix::Csr *u_new) -{ GKO_NOT_IMPLEMENTED; -} GKO_INSTANTIATE_FOR_EACH_VALUE_AND_INDEX_TYPE( GKO_DECLARE_PAR_ILUT_ADD_CANDIDATES_KERNEL); diff --git a/dpcpp/get_info.cmake b/dpcpp/get_info.cmake index 475a53f5dcf..7cbfa34f7d0 100644 --- a/dpcpp/get_info.cmake +++ b/dpcpp/get_info.cmake @@ -1,4 +1,7 @@ ginkgo_print_module_header(${detailed_log} "DPCPP") ginkgo_print_module_footer(${detailed_log} "DPCPP variables:") ginkgo_print_variable(${detailed_log} "GINKGO_DPCPP_FLAGS") +ginkgo_print_module_footer(${detailed_log} "DPCPP environment variables:") +ginkgo_print_env_variable(${detailed_log} "SYCL_DEVICE_TYPE") +ginkgo_print_env_variable(${detailed_log} "SYCL_BE") ginkgo_print_module_footer(${detailed_log} "") diff --git a/dpcpp/matrix/coo_kernels.dp.cpp b/dpcpp/matrix/coo_kernels.dp.cpp index b72b24d1fa5..57f70edf499 100644 --- a/dpcpp/matrix/coo_kernels.dp.cpp +++ b/dpcpp/matrix/coo_kernels.dp.cpp @@ -64,10 +64,8 @@ namespace coo { template void spmv(std::shared_ptr exec, const matrix::Coo *a, - const matrix::Dense *b, matrix::Dense *c) -{ - GKO_NOT_IMPLEMENTED; -} + const matrix::Dense *b, + matrix::Dense *c) GKO_NOT_IMPLEMENTED; GKO_INSTANTIATE_FOR_EACH_VALUE_AND_INDEX_TYPE(GKO_DECLARE_COO_SPMV_KERNEL); @@ -78,10 +76,7 @@ void advanced_spmv(std::shared_ptr exec, const matrix::Coo *a, const matrix::Dense *b, const matrix::Dense *beta, - matrix::Dense *c) -{ - GKO_NOT_IMPLEMENTED; -} + matrix::Dense *c) GKO_NOT_IMPLEMENTED; GKO_INSTANTIATE_FOR_EACH_VALUE_AND_INDEX_TYPE( GKO_DECLARE_COO_ADVANCED_SPMV_KERNEL); @@ -90,10 +85,8 @@ GKO_INSTANTIATE_FOR_EACH_VALUE_AND_INDEX_TYPE( template void spmv2(std::shared_ptr exec, const matrix::Coo *a, - const matrix::Dense *b, matrix::Dense *c) -{ - GKO_NOT_IMPLEMENTED; -} + const matrix::Dense *b, + matrix::Dense *c) GKO_NOT_IMPLEMENTED; GKO_INSTANTIATE_FOR_EACH_VALUE_AND_INDEX_TYPE(GKO_DECLARE_COO_SPMV2_KERNEL); @@ -103,10 +96,7 @@ void advanced_spmv2(std::shared_ptr exec, const matrix::Dense *alpha, const matrix::Coo *a, const matrix::Dense *b, - matrix::Dense *c) -{ - GKO_NOT_IMPLEMENTED; -} + matrix::Dense *c) GKO_NOT_IMPLEMENTED; GKO_INSTANTIATE_FOR_EACH_VALUE_AND_INDEX_TYPE( GKO_DECLARE_COO_ADVANCED_SPMV2_KERNEL); @@ -115,19 +105,15 @@ GKO_INSTANTIATE_FOR_EACH_VALUE_AND_INDEX_TYPE( template void convert_row_idxs_to_ptrs(std::shared_ptr exec, const IndexType *idxs, size_type num_nonzeros, - IndexType *ptrs, size_type length) -{ - GKO_NOT_IMPLEMENTED; -} + IndexType *ptrs, + size_type length) GKO_NOT_IMPLEMENTED; template void convert_to_csr(std::shared_ptr exec, const matrix::Coo *source, matrix::Csr *result) -{ GKO_NOT_IMPLEMENTED; -} GKO_INSTANTIATE_FOR_EACH_VALUE_AND_INDEX_TYPE( GKO_DECLARE_COO_CONVERT_TO_CSR_KERNEL); @@ -136,10 +122,7 @@ GKO_INSTANTIATE_FOR_EACH_VALUE_AND_INDEX_TYPE( template void convert_to_dense(std::shared_ptr exec, const matrix::Coo *source, - matrix::Dense *result) -{ - GKO_NOT_IMPLEMENTED; -} + matrix::Dense *result) GKO_NOT_IMPLEMENTED; GKO_INSTANTIATE_FOR_EACH_VALUE_AND_INDEX_TYPE( GKO_DECLARE_COO_CONVERT_TO_DENSE_KERNEL); @@ -148,10 +131,7 @@ GKO_INSTANTIATE_FOR_EACH_VALUE_AND_INDEX_TYPE( template void extract_diagonal(std::shared_ptr exec, const matrix::Coo *orig, - matrix::Diagonal *diag) -{ - GKO_NOT_IMPLEMENTED; -} + matrix::Diagonal *diag) GKO_NOT_IMPLEMENTED; GKO_INSTANTIATE_FOR_EACH_VALUE_AND_INDEX_TYPE( GKO_DECLARE_COO_EXTRACT_DIAGONAL_KERNEL); diff --git a/dpcpp/matrix/csr_kernels.dp.cpp b/dpcpp/matrix/csr_kernels.dp.cpp index 3516b7efc04..7863539f7a2 100644 --- a/dpcpp/matrix/csr_kernels.dp.cpp +++ b/dpcpp/matrix/csr_kernels.dp.cpp @@ -53,7 +53,6 @@ OF THIS SOFTWARE, EVEN IF ADVISED OF THE POSSIBILITY OF SUCH DAMAGE. #include "core/base/iterator_factory.hpp" #include "core/components/prefix_sum.hpp" #include "core/matrix/csr_builder.hpp" -#include "dpcpp/components/csr_spgeam.dp.hpp" #include "dpcpp/components/format_conversion.dp.hpp" @@ -71,10 +70,8 @@ namespace csr { template void spmv(std::shared_ptr exec, const matrix::Csr *a, - const matrix::Dense *b, matrix::Dense *c) -{ - GKO_NOT_IMPLEMENTED; -} + const matrix::Dense *b, + matrix::Dense *c) GKO_NOT_IMPLEMENTED; GKO_INSTANTIATE_FOR_EACH_VALUE_AND_INDEX_TYPE(GKO_DECLARE_CSR_SPMV_KERNEL); @@ -85,10 +82,7 @@ void advanced_spmv(std::shared_ptr exec, const matrix::Csr *a, const matrix::Dense *b, const matrix::Dense *beta, - matrix::Dense *c) -{ - GKO_NOT_IMPLEMENTED; -} + matrix::Dense *c) GKO_NOT_IMPLEMENTED; GKO_INSTANTIATE_FOR_EACH_VALUE_AND_INDEX_TYPE( GKO_DECLARE_CSR_ADVANCED_SPMV_KERNEL); @@ -97,49 +91,34 @@ GKO_INSTANTIATE_FOR_EACH_VALUE_AND_INDEX_TYPE( template void spgemm_insert_row(unordered_set &cols, const matrix::Csr *c, - size_type row) -{ - GKO_NOT_IMPLEMENTED; -} + size_type row) GKO_NOT_IMPLEMENTED; template void spgemm_insert_row2(unordered_set &cols, const matrix::Csr *a, const matrix::Csr *b, - size_type row) -{ - GKO_NOT_IMPLEMENTED; -} + size_type row) GKO_NOT_IMPLEMENTED; template void spgemm_accumulate_row(map &cols, const matrix::Csr *c, - ValueType scale, size_type row) -{ - GKO_NOT_IMPLEMENTED; -} + ValueType scale, size_type row) GKO_NOT_IMPLEMENTED; template void spgemm_accumulate_row2(map &cols, const matrix::Csr *a, const matrix::Csr *b, - ValueType scale, size_type row) -{ - GKO_NOT_IMPLEMENTED; -} + ValueType scale, size_type row) GKO_NOT_IMPLEMENTED; template void spgemm(std::shared_ptr exec, const matrix::Csr *a, const matrix::Csr *b, - matrix::Csr *c) -{ - GKO_NOT_IMPLEMENTED; -} + matrix::Csr *c) GKO_NOT_IMPLEMENTED; GKO_INSTANTIATE_FOR_EACH_VALUE_AND_INDEX_TYPE(GKO_DECLARE_CSR_SPGEMM_KERNEL); @@ -151,10 +130,7 @@ void advanced_spgemm(std::shared_ptr exec, const matrix::Csr *b, const matrix::Dense *beta, const matrix::Csr *d, - matrix::Csr *c) -{ - GKO_NOT_IMPLEMENTED; -} + matrix::Csr *c) GKO_NOT_IMPLEMENTED; GKO_INSTANTIATE_FOR_EACH_VALUE_AND_INDEX_TYPE( GKO_DECLARE_CSR_ADVANCED_SPGEMM_KERNEL); @@ -166,10 +142,7 @@ void spgeam(std::shared_ptr exec, const matrix::Csr *a, const matrix::Dense *beta, const matrix::Csr *b, - matrix::Csr *c) -{ - GKO_NOT_IMPLEMENTED; -} + matrix::Csr *c) GKO_NOT_IMPLEMENTED; GKO_INSTANTIATE_FOR_EACH_VALUE_AND_INDEX_TYPE(GKO_DECLARE_CSR_SPGEAM_KERNEL); @@ -177,19 +150,14 @@ GKO_INSTANTIATE_FOR_EACH_VALUE_AND_INDEX_TYPE(GKO_DECLARE_CSR_SPGEAM_KERNEL); template void convert_row_ptrs_to_idxs(std::shared_ptr exec, const IndexType *ptrs, size_type num_rows, - IndexType *idxs) -{ - GKO_NOT_IMPLEMENTED; -} + IndexType *idxs) GKO_NOT_IMPLEMENTED; template void convert_to_coo(std::shared_ptr exec, const matrix::Csr *source, matrix::Coo *result) -{ GKO_NOT_IMPLEMENTED; -} GKO_INSTANTIATE_FOR_EACH_VALUE_AND_INDEX_TYPE( GKO_DECLARE_CSR_CONVERT_TO_COO_KERNEL); @@ -198,10 +166,7 @@ GKO_INSTANTIATE_FOR_EACH_VALUE_AND_INDEX_TYPE( template void convert_to_dense(std::shared_ptr exec, const matrix::Csr *source, - matrix::Dense *result) -{ - GKO_NOT_IMPLEMENTED; -} + matrix::Dense *result) GKO_NOT_IMPLEMENTED; GKO_INSTANTIATE_FOR_EACH_VALUE_AND_INDEX_TYPE( GKO_DECLARE_CSR_CONVERT_TO_DENSE_KERNEL); @@ -232,29 +197,20 @@ inline void convert_csr_to_csc(size_type num_rows, const IndexType *row_ptrs, const IndexType *col_idxs, const ValueType *csr_vals, IndexType *row_idxs, IndexType *col_ptrs, ValueType *csc_vals, - UnaryOperator op) -{ - GKO_NOT_IMPLEMENTED; -} + UnaryOperator op) GKO_NOT_IMPLEMENTED; template void transpose_and_transform(std::shared_ptr exec, matrix::Csr *trans, const matrix::Csr *orig, - UnaryOperator op) -{ - GKO_NOT_IMPLEMENTED; -} + UnaryOperator op) GKO_NOT_IMPLEMENTED; template void transpose(std::shared_ptr exec, const matrix::Csr *orig, - matrix::Csr *trans) -{ - GKO_NOT_IMPLEMENTED; -} + matrix::Csr *trans) GKO_NOT_IMPLEMENTED; GKO_INSTANTIATE_FOR_EACH_VALUE_AND_INDEX_TYPE(GKO_DECLARE_CSR_TRANSPOSE_KERNEL); @@ -263,9 +219,7 @@ template void conj_transpose(std::shared_ptr exec, const matrix::Csr *orig, matrix::Csr *trans) -{ GKO_NOT_IMPLEMENTED; -} GKO_INSTANTIATE_FOR_EACH_VALUE_AND_INDEX_TYPE( GKO_DECLARE_CSR_CONJ_TRANSPOSE_KERNEL); @@ -294,80 +248,45 @@ template void convert_to_hybrid(std::shared_ptr exec, const matrix::Csr *source, matrix::Hybrid *result) -{ GKO_NOT_IMPLEMENTED; -} GKO_INSTANTIATE_FOR_EACH_VALUE_AND_INDEX_TYPE( GKO_DECLARE_CSR_CONVERT_TO_HYBRID_KERNEL); -template -void row_permute_impl(std::shared_ptr exec, - const Array *permutation_indices, - const matrix::Csr *orig, - matrix::Csr *row_permuted) -{ - GKO_NOT_IMPLEMENTED; -} +template +void invert_permutation(std::shared_ptr exec, + size_type size, const IndexType *permutation_indices, + IndexType *inv_permutation) GKO_NOT_IMPLEMENTED; + +GKO_INSTANTIATE_FOR_EACH_INDEX_TYPE(GKO_DECLARE_INVERT_PERMUTATION_KERNEL); template -void row_permute(std::shared_ptr exec, - const Array *permutation_indices, - const matrix::Csr *orig, - matrix::Csr *row_permuted) -{ - GKO_NOT_IMPLEMENTED; -} +void row_permute( + std::shared_ptr exec, const IndexType *perm, + const matrix::Csr *orig, + matrix::Csr *row_permuted) GKO_NOT_IMPLEMENTED; GKO_INSTANTIATE_FOR_EACH_VALUE_AND_INDEX_TYPE( GKO_DECLARE_CSR_ROW_PERMUTE_KERNEL); template -void inverse_row_permute(std::shared_ptr exec, - const Array *permutation_indices, - const matrix::Csr *orig, - matrix::Csr *row_permuted) -{ - GKO_NOT_IMPLEMENTED; -} +void inverse_row_permute( + std::shared_ptr exec, const IndexType *perm, + const matrix::Csr *orig, + matrix::Csr *row_permuted) GKO_NOT_IMPLEMENTED; GKO_INSTANTIATE_FOR_EACH_VALUE_AND_INDEX_TYPE( GKO_DECLARE_CSR_INVERSE_ROW_PERMUTE_KERNEL); template -void column_permute_impl(const Array *permutation_indices, - const matrix::Csr *orig, - matrix::Csr *column_permuted) -{ - GKO_NOT_IMPLEMENTED; -} - - -template -void column_permute(std::shared_ptr exec, - const Array *permutation_indices, - const matrix::Csr *orig, - matrix::Csr *column_permuted) -{ - GKO_NOT_IMPLEMENTED; -} - -GKO_INSTANTIATE_FOR_EACH_VALUE_AND_INDEX_TYPE( - GKO_DECLARE_CSR_COLUMN_PERMUTE_KERNEL); - - -template -void inverse_column_permute(std::shared_ptr exec, - const Array *permutation_indices, - const matrix::Csr *orig, - matrix::Csr *column_permuted) -{ - GKO_NOT_IMPLEMENTED; -} +void inverse_column_permute( + std::shared_ptr exec, const IndexType *perm, + const matrix::Csr *orig, + matrix::Csr *column_permuted) GKO_NOT_IMPLEMENTED; GKO_INSTANTIATE_FOR_EACH_VALUE_AND_INDEX_TYPE( GKO_DECLARE_CSR_INVERSE_COLUMN_PERMUTE_KERNEL); @@ -376,10 +295,7 @@ GKO_INSTANTIATE_FOR_EACH_VALUE_AND_INDEX_TYPE( template void calculate_nonzeros_per_row(std::shared_ptr exec, const matrix::Csr *source, - Array *result) -{ - GKO_NOT_IMPLEMENTED; -} + Array *result) GKO_NOT_IMPLEMENTED; GKO_INSTANTIATE_FOR_EACH_VALUE_AND_INDEX_TYPE( GKO_DECLARE_CSR_CALCULATE_NONZEROS_PER_ROW_KERNEL); @@ -388,9 +304,7 @@ GKO_INSTANTIATE_FOR_EACH_VALUE_AND_INDEX_TYPE( template void sort_by_column_index(std::shared_ptr exec, matrix::Csr *to_sort) -{ GKO_NOT_IMPLEMENTED; -} GKO_INSTANTIATE_FOR_EACH_VALUE_AND_INDEX_TYPE( GKO_DECLARE_CSR_SORT_BY_COLUMN_INDEX); @@ -399,10 +313,8 @@ GKO_INSTANTIATE_FOR_EACH_VALUE_AND_INDEX_TYPE( template void is_sorted_by_column_index( std::shared_ptr exec, - const matrix::Csr *to_check, bool *is_sorted) -{ - GKO_NOT_IMPLEMENTED; -} + const matrix::Csr *to_check, + bool *is_sorted) GKO_NOT_IMPLEMENTED; GKO_INSTANTIATE_FOR_EACH_VALUE_AND_INDEX_TYPE( GKO_DECLARE_CSR_IS_SORTED_BY_COLUMN_INDEX); @@ -411,10 +323,7 @@ GKO_INSTANTIATE_FOR_EACH_VALUE_AND_INDEX_TYPE( template void extract_diagonal(std::shared_ptr exec, const matrix::Csr *orig, - matrix::Diagonal *diag) -{ - GKO_NOT_IMPLEMENTED; -} + matrix::Diagonal *diag) GKO_NOT_IMPLEMENTED; GKO_INSTANTIATE_FOR_EACH_VALUE_AND_INDEX_TYPE(GKO_DECLARE_CSR_EXTRACT_DIAGONAL); diff --git a/dpcpp/matrix/dense_kernels.dp.cpp b/dpcpp/matrix/dense_kernels.dp.cpp index 5f76e1b5b97..e84519e8192 100644 --- a/dpcpp/matrix/dense_kernels.dp.cpp +++ b/dpcpp/matrix/dense_kernels.dp.cpp @@ -69,10 +69,7 @@ template void simple_apply(std::shared_ptr exec, const matrix::Dense *a, const matrix::Dense *b, - matrix::Dense *c) -{ - GKO_NOT_IMPLEMENTED; -} + matrix::Dense *c) GKO_NOT_IMPLEMENTED; GKO_INSTANTIATE_FOR_EACH_VALUE_TYPE(GKO_DECLARE_DENSE_SIMPLE_APPLY_KERNEL); @@ -81,20 +78,16 @@ template void apply(std::shared_ptr exec, const matrix::Dense *alpha, const matrix::Dense *a, const matrix::Dense *b, - const matrix::Dense *beta, matrix::Dense *c) -{ - GKO_NOT_IMPLEMENTED; -} + const matrix::Dense *beta, + matrix::Dense *c) GKO_NOT_IMPLEMENTED; GKO_INSTANTIATE_FOR_EACH_VALUE_TYPE(GKO_DECLARE_DENSE_APPLY_KERNEL); template void scale(std::shared_ptr exec, - const matrix::Dense *alpha, matrix::Dense *x) -{ - GKO_NOT_IMPLEMENTED; -} + const matrix::Dense *alpha, + matrix::Dense *x) GKO_NOT_IMPLEMENTED; GKO_INSTANTIATE_FOR_EACH_VALUE_TYPE(GKO_DECLARE_DENSE_SCALE_KERNEL); @@ -102,10 +95,8 @@ GKO_INSTANTIATE_FOR_EACH_VALUE_TYPE(GKO_DECLARE_DENSE_SCALE_KERNEL); template void add_scaled(std::shared_ptr exec, const matrix::Dense *alpha, - const matrix::Dense *x, matrix::Dense *y) -{ - GKO_NOT_IMPLEMENTED; -} + const matrix::Dense *x, + matrix::Dense *y) GKO_NOT_IMPLEMENTED; GKO_INSTANTIATE_FOR_EACH_VALUE_TYPE(GKO_DECLARE_DENSE_ADD_SCALED_KERNEL); @@ -114,10 +105,7 @@ template void add_scaled_diag(std::shared_ptr exec, const matrix::Dense *alpha, const matrix::Diagonal *x, - matrix::Dense *y) -{ - GKO_NOT_IMPLEMENTED; -} + matrix::Dense *y) GKO_NOT_IMPLEMENTED; GKO_INSTANTIATE_FOR_EACH_VALUE_TYPE(GKO_DECLARE_DENSE_ADD_SCALED_DIAG_KERNEL); @@ -126,10 +114,7 @@ template void compute_dot(std::shared_ptr exec, const matrix::Dense *x, const matrix::Dense *y, - matrix::Dense *result) -{ - GKO_NOT_IMPLEMENTED; -} + matrix::Dense *result) GKO_NOT_IMPLEMENTED; GKO_INSTANTIATE_FOR_EACH_VALUE_TYPE(GKO_DECLARE_DENSE_COMPUTE_DOT_KERNEL); @@ -138,9 +123,7 @@ template void compute_norm2(std::shared_ptr exec, const matrix::Dense *x, matrix::Dense> *result) -{ GKO_NOT_IMPLEMENTED; -} GKO_INSTANTIATE_FOR_EACH_VALUE_TYPE(GKO_DECLARE_DENSE_COMPUTE_NORM2_KERNEL); @@ -149,9 +132,7 @@ template void convert_to_coo(std::shared_ptr exec, const matrix::Dense *source, matrix::Coo *result) -{ GKO_NOT_IMPLEMENTED; -} GKO_INSTANTIATE_FOR_EACH_VALUE_AND_INDEX_TYPE( GKO_DECLARE_DENSE_CONVERT_TO_COO_KERNEL); @@ -161,9 +142,7 @@ template void convert_to_csr(std::shared_ptr exec, const matrix::Dense *source, matrix::Csr *result) -{ GKO_NOT_IMPLEMENTED; -} GKO_INSTANTIATE_FOR_EACH_VALUE_AND_INDEX_TYPE( GKO_DECLARE_DENSE_CONVERT_TO_CSR_KERNEL); @@ -173,9 +152,7 @@ template void convert_to_ell(std::shared_ptr exec, const matrix::Dense *source, matrix::Ell *result) -{ GKO_NOT_IMPLEMENTED; -} GKO_INSTANTIATE_FOR_EACH_VALUE_AND_INDEX_TYPE( GKO_DECLARE_DENSE_CONVERT_TO_ELL_KERNEL); @@ -185,9 +162,7 @@ template void convert_to_hybrid(std::shared_ptr exec, const matrix::Dense *source, matrix::Hybrid *result) -{ GKO_NOT_IMPLEMENTED; -} GKO_INSTANTIATE_FOR_EACH_VALUE_AND_INDEX_TYPE( GKO_DECLARE_DENSE_CONVERT_TO_HYBRID_KERNEL); @@ -197,9 +172,7 @@ template void convert_to_sellp(std::shared_ptr exec, const matrix::Dense *source, matrix::Sellp *result) -{ GKO_NOT_IMPLEMENTED; -} GKO_INSTANTIATE_FOR_EACH_VALUE_AND_INDEX_TYPE( GKO_DECLARE_DENSE_CONVERT_TO_SELLP_KERNEL); @@ -209,9 +182,7 @@ template void convert_to_sparsity_csr(std::shared_ptr exec, const matrix::Dense *source, matrix::SparsityCsr *result) -{ GKO_NOT_IMPLEMENTED; -} GKO_INSTANTIATE_FOR_EACH_VALUE_AND_INDEX_TYPE( GKO_DECLARE_DENSE_CONVERT_TO_SPARSITY_CSR_KERNEL); @@ -219,10 +190,8 @@ GKO_INSTANTIATE_FOR_EACH_VALUE_AND_INDEX_TYPE( template void count_nonzeros(std::shared_ptr exec, - const matrix::Dense *source, size_type *result) -{ - GKO_NOT_IMPLEMENTED; -} + const matrix::Dense *source, + size_type *result) GKO_NOT_IMPLEMENTED; GKO_INSTANTIATE_FOR_EACH_VALUE_TYPE(GKO_DECLARE_DENSE_COUNT_NONZEROS_KERNEL); @@ -230,10 +199,7 @@ GKO_INSTANTIATE_FOR_EACH_VALUE_TYPE(GKO_DECLARE_DENSE_COUNT_NONZEROS_KERNEL); template void calculate_max_nnz_per_row(std::shared_ptr exec, const matrix::Dense *source, - size_type *result) -{ - GKO_NOT_IMPLEMENTED; -} + size_type *result) GKO_NOT_IMPLEMENTED; GKO_INSTANTIATE_FOR_EACH_VALUE_TYPE( GKO_DECLARE_DENSE_CALCULATE_MAX_NNZ_PER_ROW_KERNEL); @@ -242,10 +208,7 @@ GKO_INSTANTIATE_FOR_EACH_VALUE_TYPE( template void calculate_nonzeros_per_row(std::shared_ptr exec, const matrix::Dense *source, - Array *result) -{ - GKO_NOT_IMPLEMENTED; -} + Array *result) GKO_NOT_IMPLEMENTED; GKO_INSTANTIATE_FOR_EACH_VALUE_TYPE( GKO_DECLARE_DENSE_CALCULATE_NONZEROS_PER_ROW_KERNEL); @@ -255,10 +218,7 @@ template void calculate_total_cols(std::shared_ptr exec, const matrix::Dense *source, size_type *result, size_type stride_factor, - size_type slice_size) -{ - GKO_NOT_IMPLEMENTED; -} + size_type slice_size) GKO_NOT_IMPLEMENTED; GKO_INSTANTIATE_FOR_EACH_VALUE_TYPE( GKO_DECLARE_DENSE_CALCULATE_TOTAL_COLS_KERNEL); @@ -267,35 +227,27 @@ GKO_INSTANTIATE_FOR_EACH_VALUE_TYPE( template void transpose(std::shared_ptr exec, const matrix::Dense *orig, - matrix::Dense *trans) -{ - GKO_NOT_IMPLEMENTED; -} + matrix::Dense *trans) GKO_NOT_IMPLEMENTED; -GKO_INSTANTIATE_FOR_EACH_VALUE_TYPE(GKO_DECLARE_TRANSPOSE_KERNEL); +GKO_INSTANTIATE_FOR_EACH_VALUE_TYPE(GKO_DECLARE_DENSE_TRANSPOSE_KERNEL); template void conj_transpose(std::shared_ptr exec, const matrix::Dense *orig, - matrix::Dense *trans) -{ - GKO_NOT_IMPLEMENTED; -} + matrix::Dense *trans) GKO_NOT_IMPLEMENTED; -GKO_INSTANTIATE_FOR_EACH_VALUE_TYPE(GKO_DECLARE_CONJ_TRANSPOSE_KERNEL); +GKO_INSTANTIATE_FOR_EACH_VALUE_TYPE(GKO_DECLARE_DENSE_CONJ_TRANSPOSE_KERNEL); template void row_permute(std::shared_ptr exec, const Array *permutation_indices, const matrix::Dense *orig, - matrix::Dense *row_permuted) -{ - GKO_NOT_IMPLEMENTED; -} + matrix::Dense *row_permuted) GKO_NOT_IMPLEMENTED; -GKO_INSTANTIATE_FOR_EACH_VALUE_AND_INDEX_TYPE(GKO_DECLARE_ROW_PERMUTE_KERNEL); +GKO_INSTANTIATE_FOR_EACH_VALUE_AND_INDEX_TYPE( + GKO_DECLARE_DENSE_ROW_PERMUTE_KERNEL); template @@ -303,12 +255,10 @@ void column_permute(std::shared_ptr exec, const Array *permutation_indices, const matrix::Dense *orig, matrix::Dense *column_permuted) -{ GKO_NOT_IMPLEMENTED; -} GKO_INSTANTIATE_FOR_EACH_VALUE_AND_INDEX_TYPE( - GKO_DECLARE_COLUMN_PERMUTE_KERNEL); + GKO_DECLARE_DENSE_COLUMN_PERMUTE_KERNEL); template @@ -316,12 +266,10 @@ void inverse_row_permute(std::shared_ptr exec, const Array *permutation_indices, const matrix::Dense *orig, matrix::Dense *row_permuted) -{ GKO_NOT_IMPLEMENTED; -} GKO_INSTANTIATE_FOR_EACH_VALUE_AND_INDEX_TYPE( - GKO_DECLARE_INVERSE_ROW_PERMUTE_KERNEL); + GKO_DECLARE_DENSE_INV_ROW_PERMUTE_KERNEL); template @@ -329,31 +277,24 @@ void inverse_column_permute(std::shared_ptr exec, const Array *permutation_indices, const matrix::Dense *orig, matrix::Dense *column_permuted) -{ GKO_NOT_IMPLEMENTED; -} GKO_INSTANTIATE_FOR_EACH_VALUE_AND_INDEX_TYPE( - GKO_DECLARE_INVERSE_COLUMN_PERMUTE_KERNEL); + GKO_DECLARE_DENSE_INV_COLUMN_PERMUTE_KERNEL); template void extract_diagonal(std::shared_ptr exec, const matrix::Dense *orig, - matrix::Diagonal *diag) -{ - GKO_NOT_IMPLEMENTED; -} + matrix::Diagonal *diag) GKO_NOT_IMPLEMENTED; -GKO_INSTANTIATE_FOR_EACH_VALUE_TYPE(GKO_DECLARE_EXTRACT_DIAGONAL_KERNEL); +GKO_INSTANTIATE_FOR_EACH_VALUE_TYPE(GKO_DECLARE_DENSE_EXTRACT_DIAGONAL_KERNEL); template void inplace_absolute_dense(std::shared_ptr exec, matrix::Dense *source) -{ GKO_NOT_IMPLEMENTED; -} GKO_INSTANTIATE_FOR_EACH_VALUE_TYPE(GKO_DECLARE_INPLACE_ABSOLUTE_DENSE_KERNEL); @@ -362,9 +303,7 @@ template void outplace_absolute_dense(std::shared_ptr exec, const matrix::Dense *source, matrix::Dense> *result) -{ GKO_NOT_IMPLEMENTED; -} GKO_INSTANTIATE_FOR_EACH_VALUE_TYPE(GKO_DECLARE_OUTPLACE_ABSOLUTE_DENSE_KERNEL); diff --git a/dpcpp/matrix/diagonal_kernels.dp.cpp b/dpcpp/matrix/diagonal_kernels.dp.cpp index c2090a17dc4..8825daf65c7 100644 --- a/dpcpp/matrix/diagonal_kernels.dp.cpp +++ b/dpcpp/matrix/diagonal_kernels.dp.cpp @@ -55,10 +55,7 @@ template void apply_to_dense(std::shared_ptr exec, const matrix::Diagonal *a, const matrix::Dense *b, - matrix::Dense *c) -{ - GKO_NOT_IMPLEMENTED; -} + matrix::Dense *c) GKO_NOT_IMPLEMENTED; GKO_INSTANTIATE_FOR_EACH_VALUE_TYPE(GKO_DECLARE_DIAGONAL_APPLY_TO_DENSE_KERNEL); @@ -67,10 +64,7 @@ template void right_apply_to_dense(std::shared_ptr exec, const matrix::Diagonal *a, const matrix::Dense *b, - matrix::Dense *c) -{ - GKO_NOT_IMPLEMENTED; -} + matrix::Dense *c) GKO_NOT_IMPLEMENTED; GKO_INSTANTIATE_FOR_EACH_VALUE_TYPE( GKO_DECLARE_DIAGONAL_RIGHT_APPLY_TO_DENSE_KERNEL); @@ -80,10 +74,7 @@ template void apply_to_csr(std::shared_ptr exec, const matrix::Diagonal *a, const matrix::Csr *b, - matrix::Csr *c) -{ - GKO_NOT_IMPLEMENTED; -} + matrix::Csr *c) GKO_NOT_IMPLEMENTED; GKO_INSTANTIATE_FOR_EACH_VALUE_AND_INDEX_TYPE( GKO_DECLARE_DIAGONAL_APPLY_TO_CSR_KERNEL); @@ -94,9 +85,7 @@ void right_apply_to_csr(std::shared_ptr exec, const matrix::Diagonal *a, const matrix::Csr *b, matrix::Csr *c) -{ GKO_NOT_IMPLEMENTED; -} GKO_INSTANTIATE_FOR_EACH_VALUE_AND_INDEX_TYPE( GKO_DECLARE_DIAGONAL_RIGHT_APPLY_TO_CSR_KERNEL); @@ -106,9 +95,7 @@ template void convert_to_csr(std::shared_ptr exec, const matrix::Diagonal *source, matrix::Csr *result) -{ GKO_NOT_IMPLEMENTED; -} GKO_INSTANTIATE_FOR_EACH_VALUE_AND_INDEX_TYPE( GKO_DECLARE_DIAGONAL_CONVERT_TO_CSR_KERNEL); @@ -117,10 +104,7 @@ GKO_INSTANTIATE_FOR_EACH_VALUE_AND_INDEX_TYPE( template void conj_transpose(std::shared_ptr exec, const matrix::Diagonal *orig, - matrix::Diagonal *trans) -{ - GKO_NOT_IMPLEMENTED; -} + matrix::Diagonal *trans) GKO_NOT_IMPLEMENTED; GKO_INSTANTIATE_FOR_EACH_VALUE_TYPE(GKO_DECLARE_DIAGONAL_CONJ_TRANSPOSE_KERNEL); diff --git a/dpcpp/matrix/ell_kernels.dp.cpp b/dpcpp/matrix/ell_kernels.dp.cpp index d27a19408b4..f0bb1dbf6f6 100644 --- a/dpcpp/matrix/ell_kernels.dp.cpp +++ b/dpcpp/matrix/ell_kernels.dp.cpp @@ -59,10 +59,8 @@ namespace ell { template void spmv(std::shared_ptr exec, const matrix::Ell *a, - const matrix::Dense *b, matrix::Dense *c) -{ - GKO_NOT_IMPLEMENTED; -} + const matrix::Dense *b, + matrix::Dense *c) GKO_NOT_IMPLEMENTED; GKO_INSTANTIATE_FOR_EACH_VALUE_AND_INDEX_TYPE(GKO_DECLARE_ELL_SPMV_KERNEL); @@ -73,10 +71,7 @@ void advanced_spmv(std::shared_ptr exec, const matrix::Ell *a, const matrix::Dense *b, const matrix::Dense *beta, - matrix::Dense *c) -{ - GKO_NOT_IMPLEMENTED; -} + matrix::Dense *c) GKO_NOT_IMPLEMENTED; GKO_INSTANTIATE_FOR_EACH_VALUE_AND_INDEX_TYPE( GKO_DECLARE_ELL_ADVANCED_SPMV_KERNEL); @@ -85,10 +80,7 @@ GKO_INSTANTIATE_FOR_EACH_VALUE_AND_INDEX_TYPE( template void convert_to_dense(std::shared_ptr exec, const matrix::Ell *source, - matrix::Dense *result) -{ - GKO_NOT_IMPLEMENTED; -} + matrix::Dense *result) GKO_NOT_IMPLEMENTED; GKO_INSTANTIATE_FOR_EACH_VALUE_AND_INDEX_TYPE( GKO_DECLARE_ELL_CONVERT_TO_DENSE_KERNEL); @@ -106,10 +98,7 @@ GKO_INSTANTIATE_FOR_EACH_VALUE_AND_INDEX_TYPE( template void count_nonzeros(std::shared_ptr exec, const matrix::Ell *source, - size_type *result) -{ - GKO_NOT_IMPLEMENTED; -} + size_type *result) GKO_NOT_IMPLEMENTED; GKO_INSTANTIATE_FOR_EACH_VALUE_AND_INDEX_TYPE( GKO_DECLARE_ELL_COUNT_NONZEROS_KERNEL); @@ -127,10 +116,7 @@ GKO_INSTANTIATE_FOR_EACH_VALUE_AND_INDEX_TYPE( template void extract_diagonal(std::shared_ptr exec, const matrix::Ell *orig, - matrix::Diagonal *diag) -{ - GKO_NOT_IMPLEMENTED; -} + matrix::Diagonal *diag) GKO_NOT_IMPLEMENTED; GKO_INSTANTIATE_FOR_EACH_VALUE_AND_INDEX_TYPE( GKO_DECLARE_ELL_EXTRACT_DIAGONAL_KERNEL); diff --git a/dpcpp/matrix/hybrid_kernels.dp.cpp b/dpcpp/matrix/hybrid_kernels.dp.cpp index f740541cccc..327903c51ab 100644 --- a/dpcpp/matrix/hybrid_kernels.dp.cpp +++ b/dpcpp/matrix/hybrid_kernels.dp.cpp @@ -60,10 +60,7 @@ namespace hybrid { template void convert_to_dense(std::shared_ptr exec, const matrix::Hybrid *source, - matrix::Dense *result) -{ - GKO_NOT_IMPLEMENTED; -} + matrix::Dense *result) GKO_NOT_IMPLEMENTED; GKO_INSTANTIATE_FOR_EACH_VALUE_AND_INDEX_TYPE( GKO_DECLARE_HYBRID_CONVERT_TO_DENSE_KERNEL); @@ -73,9 +70,7 @@ template void convert_to_csr(std::shared_ptr exec, const matrix::Hybrid *source, matrix::Csr *result) -{ GKO_NOT_IMPLEMENTED; -} GKO_INSTANTIATE_FOR_EACH_VALUE_AND_INDEX_TYPE( GKO_DECLARE_HYBRID_CONVERT_TO_CSR_KERNEL); @@ -84,10 +79,7 @@ GKO_INSTANTIATE_FOR_EACH_VALUE_AND_INDEX_TYPE( template void count_nonzeros(std::shared_ptr exec, const matrix::Hybrid *source, - size_type *result) -{ - GKO_NOT_IMPLEMENTED; -} + size_type *result) GKO_NOT_IMPLEMENTED; GKO_INSTANTIATE_FOR_EACH_VALUE_AND_INDEX_TYPE( GKO_DECLARE_HYBRID_COUNT_NONZEROS_KERNEL); diff --git a/dpcpp/matrix/sellp_kernels.dp.cpp b/dpcpp/matrix/sellp_kernels.dp.cpp index 20f6adcc2a0..310c3706e88 100644 --- a/dpcpp/matrix/sellp_kernels.dp.cpp +++ b/dpcpp/matrix/sellp_kernels.dp.cpp @@ -53,10 +53,8 @@ namespace sellp { template void spmv(std::shared_ptr exec, const matrix::Sellp *a, - const matrix::Dense *b, matrix::Dense *c) -{ - GKO_NOT_IMPLEMENTED; -} + const matrix::Dense *b, + matrix::Dense *c) GKO_NOT_IMPLEMENTED; GKO_INSTANTIATE_FOR_EACH_VALUE_AND_INDEX_TYPE(GKO_DECLARE_SELLP_SPMV_KERNEL); @@ -67,10 +65,7 @@ void advanced_spmv(std::shared_ptr exec, const matrix::Sellp *a, const matrix::Dense *b, const matrix::Dense *beta, - matrix::Dense *c) -{ - GKO_NOT_IMPLEMENTED; -} + matrix::Dense *c) GKO_NOT_IMPLEMENTED; GKO_INSTANTIATE_FOR_EACH_VALUE_AND_INDEX_TYPE( GKO_DECLARE_SELLP_ADVANCED_SPMV_KERNEL); @@ -79,10 +74,7 @@ GKO_INSTANTIATE_FOR_EACH_VALUE_AND_INDEX_TYPE( template void convert_to_dense(std::shared_ptr exec, const matrix::Sellp *source, - matrix::Dense *result) -{ - GKO_NOT_IMPLEMENTED; -} + matrix::Dense *result) GKO_NOT_IMPLEMENTED; GKO_INSTANTIATE_FOR_EACH_VALUE_AND_INDEX_TYPE( GKO_DECLARE_SELLP_CONVERT_TO_DENSE_KERNEL); @@ -110,10 +102,7 @@ GKO_INSTANTIATE_FOR_EACH_VALUE_AND_INDEX_TYPE( template void extract_diagonal(std::shared_ptr exec, const matrix::Sellp *orig, - matrix::Diagonal *diag) -{ - GKO_NOT_IMPLEMENTED; -} + matrix::Diagonal *diag) GKO_NOT_IMPLEMENTED; GKO_INSTANTIATE_FOR_EACH_VALUE_AND_INDEX_TYPE( GKO_DECLARE_SELLP_EXTRACT_DIAGONAL_KERNEL); diff --git a/dpcpp/matrix/sparsity_csr_kernels.dp.cpp b/dpcpp/matrix/sparsity_csr_kernels.dp.cpp index ba682e194e7..cfc21f27727 100644 --- a/dpcpp/matrix/sparsity_csr_kernels.dp.cpp +++ b/dpcpp/matrix/sparsity_csr_kernels.dp.cpp @@ -64,10 +64,8 @@ namespace sparsity_csr { template void spmv(std::shared_ptr exec, const matrix::SparsityCsr *a, - const matrix::Dense *b, matrix::Dense *c) -{ - GKO_NOT_IMPLEMENTED; -} + const matrix::Dense *b, + matrix::Dense *c) GKO_NOT_IMPLEMENTED; GKO_INSTANTIATE_FOR_EACH_VALUE_AND_INDEX_TYPE( GKO_DECLARE_SPARSITY_CSR_SPMV_KERNEL); @@ -79,10 +77,7 @@ void advanced_spmv(std::shared_ptr exec, const matrix::SparsityCsr *a, const matrix::Dense *b, const matrix::Dense *beta, - matrix::Dense *c) -{ - GKO_NOT_IMPLEMENTED; -} + matrix::Dense *c) GKO_NOT_IMPLEMENTED; GKO_INSTANTIATE_FOR_EACH_VALUE_AND_INDEX_TYPE( GKO_DECLARE_SPARSITY_CSR_ADVANCED_SPMV_KERNEL); @@ -92,23 +87,17 @@ template void count_num_diagonal_elements( std::shared_ptr exec, const matrix::SparsityCsr *matrix, - size_type *num_diagonal_elements) -{ - GKO_NOT_IMPLEMENTED; -} + size_type *num_diagonal_elements) GKO_NOT_IMPLEMENTED; GKO_INSTANTIATE_FOR_EACH_VALUE_AND_INDEX_TYPE( GKO_DECLARE_SPARSITY_CSR_COUNT_NUM_DIAGONAL_ELEMENTS_KERNEL); template -void remove_diagonal_elements(std::shared_ptr exec, - const IndexType *row_ptrs, - const IndexType *col_idxs, - matrix::SparsityCsr *matrix) -{ - GKO_NOT_IMPLEMENTED; -} +void remove_diagonal_elements( + std::shared_ptr exec, const IndexType *row_ptrs, + const IndexType *col_idxs, + matrix::SparsityCsr *matrix) GKO_NOT_IMPLEMENTED; GKO_INSTANTIATE_FOR_EACH_VALUE_AND_INDEX_TYPE( GKO_DECLARE_SPARSITY_CSR_REMOVE_DIAGONAL_ELEMENTS_KERNEL); @@ -118,29 +107,22 @@ template inline void convert_sparsity_to_csc(size_type num_rows, const IndexType *row_ptrs, const IndexType *col_idxs, - IndexType *row_idxs, IndexType *col_ptrs) -{ - GKO_NOT_IMPLEMENTED; -} + IndexType *row_idxs, + IndexType *col_ptrs) GKO_NOT_IMPLEMENTED; template void transpose_and_transform( std::shared_ptr exec, matrix::SparsityCsr *trans, - const matrix::SparsityCsr *orig) -{ - GKO_NOT_IMPLEMENTED; -} + const matrix::SparsityCsr *orig) GKO_NOT_IMPLEMENTED; template void transpose(std::shared_ptr exec, const matrix::SparsityCsr *orig, matrix::SparsityCsr *trans) -{ GKO_NOT_IMPLEMENTED; -} GKO_INSTANTIATE_FOR_EACH_VALUE_AND_INDEX_TYPE( GKO_DECLARE_SPARSITY_CSR_TRANSPOSE_KERNEL); @@ -149,9 +131,7 @@ GKO_INSTANTIATE_FOR_EACH_VALUE_AND_INDEX_TYPE( template void sort_by_column_index(std::shared_ptr exec, matrix::SparsityCsr *to_sort) -{ GKO_NOT_IMPLEMENTED; -} GKO_INSTANTIATE_FOR_EACH_VALUE_AND_INDEX_TYPE( GKO_DECLARE_SPARSITY_CSR_SORT_BY_COLUMN_INDEX); @@ -160,10 +140,8 @@ GKO_INSTANTIATE_FOR_EACH_VALUE_AND_INDEX_TYPE( template void is_sorted_by_column_index( std::shared_ptr exec, - const matrix::SparsityCsr *to_check, bool *is_sorted) -{ - GKO_NOT_IMPLEMENTED; -} + const matrix::SparsityCsr *to_check, + bool *is_sorted) GKO_NOT_IMPLEMENTED; GKO_INSTANTIATE_FOR_EACH_VALUE_AND_INDEX_TYPE( GKO_DECLARE_SPARSITY_CSR_IS_SORTED_BY_COLUMN_INDEX); diff --git a/dpcpp/preconditioner/isai_kernels.dp.cpp b/dpcpp/preconditioner/isai_kernels.dp.cpp index 481979eed3b..a8aa955174c 100644 --- a/dpcpp/preconditioner/isai_kernels.dp.cpp +++ b/dpcpp/preconditioner/isai_kernels.dp.cpp @@ -63,10 +63,8 @@ namespace isai { template void forall_matching(const IndexType *fst, IndexType fst_size, - const IndexType *snd, IndexType snd_size, Callback cb) -{ - GKO_NOT_IMPLEMENTED; -} + const IndexType *snd, IndexType snd_size, + Callback cb) GKO_NOT_IMPLEMENTED; template @@ -74,10 +72,7 @@ void generic_generate(std::shared_ptr exec, const matrix::Csr *mtx, matrix::Csr *inverse_mtx, IndexType *excess_rhs_ptrs, IndexType *excess_nz_ptrs, - Callable trs_solve) -{ - GKO_NOT_IMPLEMENTED; -} + Callable trs_solve) GKO_NOT_IMPLEMENTED; template @@ -85,39 +80,30 @@ void generate_tri_inverse(std::shared_ptr exec, const matrix::Csr *mtx, matrix::Csr *inverse_mtx, IndexType *excess_rhs_ptrs, IndexType *excess_nz_ptrs, - bool lower) -{ - GKO_NOT_IMPLEMENTED; -} + bool lower) GKO_NOT_IMPLEMENTED; GKO_INSTANTIATE_FOR_EACH_VALUE_AND_INDEX_TYPE( GKO_DECLARE_ISAI_GENERATE_TRI_INVERSE_KERNEL); template -void generate_excess_system(std::shared_ptr, - const matrix::Csr *input, - const matrix::Csr *inverse, - const IndexType *excess_rhs_ptrs, - const IndexType *excess_nz_ptrs, - matrix::Csr *excess_system, - matrix::Dense *excess_rhs) -{ - GKO_NOT_IMPLEMENTED; -} +void generate_excess_system( + std::shared_ptr, + const matrix::Csr *input, + const matrix::Csr *inverse, + const IndexType *excess_rhs_ptrs, const IndexType *excess_nz_ptrs, + matrix::Csr *excess_system, + matrix::Dense *excess_rhs) GKO_NOT_IMPLEMENTED; GKO_INSTANTIATE_FOR_EACH_VALUE_AND_INDEX_TYPE( GKO_DECLARE_ISAI_GENERATE_EXCESS_SYSTEM_KERNEL); template -void scatter_excess_solution(std::shared_ptr, - const IndexType *excess_block_ptrs, - const matrix::Dense *excess_solution, - matrix::Csr *inverse) -{ - GKO_NOT_IMPLEMENTED; -} +void scatter_excess_solution( + std::shared_ptr, const IndexType *excess_block_ptrs, + const matrix::Dense *excess_solution, + matrix::Csr *inverse) GKO_NOT_IMPLEMENTED; GKO_INSTANTIATE_FOR_EACH_VALUE_AND_INDEX_TYPE( GKO_DECLARE_ISAI_SCATTER_EXCESS_SOLUTION_KERNEL); diff --git a/dpcpp/preconditioner/jacobi_kernels.dp.cpp b/dpcpp/preconditioner/jacobi_kernels.dp.cpp index 06b00835e82..635f08ca87e 100644 --- a/dpcpp/preconditioner/jacobi_kernels.dp.cpp +++ b/dpcpp/preconditioner/jacobi_kernels.dp.cpp @@ -69,38 +69,28 @@ namespace jacobi { void initialize_precisions(std::shared_ptr exec, const Array &source, Array &precisions) -{ GKO_NOT_IMPLEMENTED; -} namespace { template -inline bool has_same_nonzero_pattern(const IndexType *prev_row_ptr, - const IndexType *curr_row_ptr, - const IndexType *next_row_ptr) -{ - GKO_NOT_IMPLEMENTED; -} +inline bool has_same_nonzero_pattern( + const IndexType *prev_row_ptr, const IndexType *curr_row_ptr, + const IndexType *next_row_ptr) GKO_NOT_IMPLEMENTED; template size_type find_natural_blocks(const matrix::Csr *mtx, - uint32 max_block_size, IndexType *block_ptrs) -{ - GKO_NOT_IMPLEMENTED; -} + uint32 max_block_size, + IndexType *block_ptrs) GKO_NOT_IMPLEMENTED; template -inline size_type agglomerate_supervariables(uint32 max_block_size, - size_type num_natural_blocks, - IndexType *block_ptrs) -{ - GKO_NOT_IMPLEMENTED; -} +inline size_type agglomerate_supervariables( + uint32 max_block_size, size_type num_natural_blocks, + IndexType *block_ptrs) GKO_NOT_IMPLEMENTED; } // namespace @@ -110,10 +100,7 @@ template void find_blocks(std::shared_ptr exec, const matrix::Csr *system_matrix, uint32 max_block_size, size_type &num_blocks, - Array &block_pointers) -{ - GKO_NOT_IMPLEMENTED; -} + Array &block_pointers) GKO_NOT_IMPLEMENTED; GKO_INSTANTIATE_FOR_EACH_VALUE_AND_INDEX_TYPE( GKO_DECLARE_JACOBI_FIND_BLOCKS_KERNEL); @@ -125,95 +112,67 @@ namespace { template inline void extract_block(const matrix::Csr *mtx, IndexType block_size, IndexType block_start, - ValueType *block, size_type stride) -{ - GKO_NOT_IMPLEMENTED; -} + ValueType *block, + size_type stride) GKO_NOT_IMPLEMENTED; template inline IndexType choose_pivot(IndexType block_size, const ValueType *block, - size_type stride) -{ - GKO_NOT_IMPLEMENTED; -} + size_type stride) GKO_NOT_IMPLEMENTED; template inline void swap_rows(IndexType row1, IndexType row2, IndexType block_size, - ValueType *block, size_type stride) -{ - GKO_NOT_IMPLEMENTED; -} + ValueType *block, size_type stride) GKO_NOT_IMPLEMENTED; template inline bool apply_gauss_jordan_transform(IndexType row, IndexType col, IndexType block_size, ValueType *block, - size_type stride) -{ - GKO_NOT_IMPLEMENTED; -} + size_type stride) GKO_NOT_IMPLEMENTED; template > -inline void transpose_block(IndexType block_size, const SourceValueType *from, - size_type from_stride, ResultValueType *to, - size_type to_stride, - ValueConverter converter = {}) noexcept -{ - GKO_NOT_IMPLEMENTED; -} +inline void transpose_block( + IndexType block_size, const SourceValueType *from, size_type from_stride, + ResultValueType *to, size_type to_stride, + ValueConverter converter = {}) noexcept GKO_NOT_IMPLEMENTED; template > -inline void conj_transpose_block(IndexType block_size, - const SourceValueType *from, - size_type from_stride, ResultValueType *to, - size_type to_stride, - ValueConverter converter = {}) noexcept -{ - GKO_NOT_IMPLEMENTED; -} +inline void conj_transpose_block( + IndexType block_size, const SourceValueType *from, size_type from_stride, + ResultValueType *to, size_type to_stride, + ValueConverter converter = {}) noexcept GKO_NOT_IMPLEMENTED; template > -inline void permute_and_transpose_block(IndexType block_size, - const IndexType *col_perm, - const SourceValueType *source, - size_type source_stride, - ResultValueType *result, - size_type result_stride, - ValueConverter converter = {}) -{ - GKO_NOT_IMPLEMENTED; -} +inline void permute_and_transpose_block( + IndexType block_size, const IndexType *col_perm, + const SourceValueType *source, size_type source_stride, + ResultValueType *result, size_type result_stride, + ValueConverter converter = {}) GKO_NOT_IMPLEMENTED; template inline bool invert_block(IndexType block_size, IndexType *perm, - ValueType *block, size_type stride) -{ - GKO_NOT_IMPLEMENTED; -} + ValueType *block, + size_type stride) GKO_NOT_IMPLEMENTED; template inline bool validate_precision_reduction_feasibility( std::shared_ptr exec, IndexType block_size, - const ValueType *block, size_type stride) -{ - GKO_NOT_IMPLEMENTED; -} + const ValueType *block, size_type stride) GKO_NOT_IMPLEMENTED; } // namespace @@ -228,10 +187,8 @@ void generate(std::shared_ptr exec, &storage_scheme, Array> &conditioning, Array &block_precisions, - const Array &block_pointers, Array &blocks) -{ - GKO_NOT_IMPLEMENTED; -} + const Array &block_pointers, + Array &blocks) GKO_NOT_IMPLEMENTED; GKO_INSTANTIATE_FOR_EACH_VALUE_AND_INDEX_TYPE( GKO_DECLARE_JACOBI_GENERATE_KERNEL); @@ -247,10 +204,7 @@ inline void apply_block(size_type block_size, size_type num_rhs, const BlockValueType *block, size_type stride, ValueType alpha, const ValueType *b, size_type stride_b, ValueType beta, ValueType *x, size_type stride_x, - ValueConverter converter = {}) -{ - GKO_NOT_IMPLEMENTED; -} + ValueConverter converter = {}) GKO_NOT_IMPLEMENTED; } // namespace @@ -266,10 +220,8 @@ void apply(std::shared_ptr exec, size_type num_blocks, const Array &blocks, const matrix::Dense *alpha, const matrix::Dense *b, - const matrix::Dense *beta, matrix::Dense *x) -{ - GKO_NOT_IMPLEMENTED; -} + const matrix::Dense *beta, + matrix::Dense *x) GKO_NOT_IMPLEMENTED; GKO_INSTANTIATE_FOR_EACH_VALUE_AND_INDEX_TYPE(GKO_DECLARE_JACOBI_APPLY_KERNEL); @@ -282,10 +234,8 @@ void simple_apply( &storage_scheme, const Array &block_precisions, const Array &block_pointers, const Array &blocks, - const matrix::Dense *b, matrix::Dense *x) -{ - GKO_NOT_IMPLEMENTED; -} + const matrix::Dense *b, + matrix::Dense *x) GKO_NOT_IMPLEMENTED; GKO_INSTANTIATE_FOR_EACH_VALUE_AND_INDEX_TYPE( GKO_DECLARE_JACOBI_SIMPLE_APPLY_KERNEL); @@ -298,10 +248,7 @@ void transpose_jacobi( const Array &block_pointers, const Array &blocks, const preconditioner::block_interleaved_storage_scheme &storage_scheme, - Array &out_blocks) -{ - GKO_NOT_IMPLEMENTED; -} + Array &out_blocks) GKO_NOT_IMPLEMENTED; GKO_INSTANTIATE_FOR_EACH_VALUE_AND_INDEX_TYPE( GKO_DECLARE_JACOBI_TRANSPOSE_KERNEL); @@ -314,10 +261,7 @@ void conj_transpose_jacobi( const Array &block_pointers, const Array &blocks, const preconditioner::block_interleaved_storage_scheme &storage_scheme, - Array &out_blocks) -{ - GKO_NOT_IMPLEMENTED; -} + Array &out_blocks) GKO_NOT_IMPLEMENTED; GKO_INSTANTIATE_FOR_EACH_VALUE_AND_INDEX_TYPE( GKO_DECLARE_JACOBI_CONJ_TRANSPOSE_KERNEL); @@ -330,10 +274,7 @@ void convert_to_dense( const Array &block_pointers, const Array &blocks, const preconditioner::block_interleaved_storage_scheme &storage_scheme, - ValueType *result_values, size_type result_stride) -{ - GKO_NOT_IMPLEMENTED; -} + ValueType *result_values, size_type result_stride) GKO_NOT_IMPLEMENTED; GKO_INSTANTIATE_FOR_EACH_VALUE_AND_INDEX_TYPE( GKO_DECLARE_JACOBI_CONVERT_TO_DENSE_KERNEL); diff --git a/dpcpp/solver/bicg_kernels.dp.cpp b/dpcpp/solver/bicg_kernels.dp.cpp index 11a5471002a..bb1b6e0f49d 100644 --- a/dpcpp/solver/bicg_kernels.dp.cpp +++ b/dpcpp/solver/bicg_kernels.dp.cpp @@ -61,10 +61,7 @@ void initialize(std::shared_ptr exec, matrix::Dense *rho, matrix::Dense *r2, matrix::Dense *z2, matrix::Dense *p2, matrix::Dense *q2, - Array *stop_status) -{ - GKO_NOT_IMPLEMENTED; -} + Array *stop_status) GKO_NOT_IMPLEMENTED; GKO_INSTANTIATE_FOR_EACH_VALUE_TYPE(GKO_DECLARE_BICG_INITIALIZE_KERNEL); @@ -75,10 +72,7 @@ void step_1(std::shared_ptr exec, matrix::Dense *p2, const matrix::Dense *z2, const matrix::Dense *rho, const matrix::Dense *prev_rho, - const Array *stop_status) -{ - GKO_NOT_IMPLEMENTED; -} + const Array *stop_status) GKO_NOT_IMPLEMENTED; GKO_INSTANTIATE_FOR_EACH_VALUE_TYPE(GKO_DECLARE_BICG_STEP_1_KERNEL); @@ -91,10 +85,7 @@ void step_2(std::shared_ptr exec, const matrix::Dense *q2, const matrix::Dense *beta, const matrix::Dense *rho, - const Array *stop_status) -{ - GKO_NOT_IMPLEMENTED; -} + const Array *stop_status) GKO_NOT_IMPLEMENTED; GKO_INSTANTIATE_FOR_EACH_VALUE_TYPE(GKO_DECLARE_BICG_STEP_2_KERNEL); diff --git a/dpcpp/solver/bicgstab_kernels.dp.cpp b/dpcpp/solver/bicgstab_kernels.dp.cpp index af4d42289cd..25b133f087e 100644 --- a/dpcpp/solver/bicgstab_kernels.dp.cpp +++ b/dpcpp/solver/bicgstab_kernels.dp.cpp @@ -65,10 +65,7 @@ void initialize(std::shared_ptr exec, matrix::Dense *rho, matrix::Dense *alpha, matrix::Dense *beta, matrix::Dense *gamma, matrix::Dense *omega, - Array *stop_status) -{ - GKO_NOT_IMPLEMENTED; -} + Array *stop_status) GKO_NOT_IMPLEMENTED; GKO_INSTANTIATE_FOR_EACH_VALUE_TYPE(GKO_DECLARE_BICGSTAB_INITIALIZE_KERNEL); @@ -81,10 +78,7 @@ void step_1(std::shared_ptr exec, const matrix::Dense *prev_rho, const matrix::Dense *alpha, const matrix::Dense *omega, - const Array *stop_status) -{ - GKO_NOT_IMPLEMENTED; -} + const Array *stop_status) GKO_NOT_IMPLEMENTED; GKO_INSTANTIATE_FOR_EACH_VALUE_TYPE(GKO_DECLARE_BICGSTAB_STEP_1_KERNEL); @@ -96,10 +90,7 @@ void step_2(std::shared_ptr exec, const matrix::Dense *rho, matrix::Dense *alpha, const matrix::Dense *beta, - const Array *stop_status) -{ - GKO_NOT_IMPLEMENTED; -} + const Array *stop_status) GKO_NOT_IMPLEMENTED; GKO_INSTANTIATE_FOR_EACH_VALUE_TYPE(GKO_DECLARE_BICGSTAB_STEP_2_KERNEL); @@ -111,10 +102,8 @@ void step_3( const matrix::Dense *t, const matrix::Dense *y, const matrix::Dense *z, const matrix::Dense *alpha, const matrix::Dense *beta, const matrix::Dense *gamma, - matrix::Dense *omega, const Array *stop_status) -{ - GKO_NOT_IMPLEMENTED; -} + matrix::Dense *omega, + const Array *stop_status) GKO_NOT_IMPLEMENTED; GKO_INSTANTIATE_FOR_EACH_VALUE_TYPE(GKO_DECLARE_BICGSTAB_STEP_3_KERNEL); @@ -123,10 +112,7 @@ template void finalize(std::shared_ptr exec, matrix::Dense *x, const matrix::Dense *y, const matrix::Dense *alpha, - Array *stop_status) -{ - GKO_NOT_IMPLEMENTED; -} + Array *stop_status) GKO_NOT_IMPLEMENTED; GKO_INSTANTIATE_FOR_EACH_VALUE_TYPE(GKO_DECLARE_BICGSTAB_FINALIZE_KERNEL); diff --git a/dpcpp/solver/cg_kernels.dp.cpp b/dpcpp/solver/cg_kernels.dp.cpp index 9cb2e0f30da..2e5bf4bc47b 100644 --- a/dpcpp/solver/cg_kernels.dp.cpp +++ b/dpcpp/solver/cg_kernels.dp.cpp @@ -59,10 +59,7 @@ void initialize(std::shared_ptr exec, matrix::Dense *z, matrix::Dense *p, matrix::Dense *q, matrix::Dense *prev_rho, matrix::Dense *rho, - Array *stop_status) -{ - GKO_NOT_IMPLEMENTED; -} + Array *stop_status) GKO_NOT_IMPLEMENTED; GKO_INSTANTIATE_FOR_EACH_VALUE_TYPE(GKO_DECLARE_CG_INITIALIZE_KERNEL); @@ -72,10 +69,7 @@ void step_1(std::shared_ptr exec, matrix::Dense *p, const matrix::Dense *z, const matrix::Dense *rho, const matrix::Dense *prev_rho, - const Array *stop_status) -{ - GKO_NOT_IMPLEMENTED; -} + const Array *stop_status) GKO_NOT_IMPLEMENTED; GKO_INSTANTIATE_FOR_EACH_VALUE_TYPE(GKO_DECLARE_CG_STEP_1_KERNEL); @@ -87,10 +81,7 @@ void step_2(std::shared_ptr exec, const matrix::Dense *q, const matrix::Dense *beta, const matrix::Dense *rho, - const Array *stop_status) -{ - GKO_NOT_IMPLEMENTED; -} + const Array *stop_status) GKO_NOT_IMPLEMENTED; GKO_INSTANTIATE_FOR_EACH_VALUE_TYPE(GKO_DECLARE_CG_STEP_2_KERNEL); diff --git a/dpcpp/solver/cgs_kernels.dp.cpp b/dpcpp/solver/cgs_kernels.dp.cpp index 04d81b682c5..42e1ff9b37e 100644 --- a/dpcpp/solver/cgs_kernels.dp.cpp +++ b/dpcpp/solver/cgs_kernels.dp.cpp @@ -64,10 +64,7 @@ void initialize(std::shared_ptr exec, matrix::Dense *gamma, matrix::Dense *prev_rho, matrix::Dense *rho, - Array *stop_status) -{ - GKO_NOT_IMPLEMENTED; -} + Array *stop_status) GKO_NOT_IMPLEMENTED; GKO_INSTANTIATE_FOR_EACH_VALUE_TYPE(GKO_DECLARE_CGS_INITIALIZE_KERNEL); @@ -78,10 +75,7 @@ void step_1(std::shared_ptr exec, matrix::Dense *p, const matrix::Dense *q, matrix::Dense *beta, const matrix::Dense *rho, const matrix::Dense *rho_prev, - const Array *stop_status) -{ - GKO_NOT_IMPLEMENTED; -} + const Array *stop_status) GKO_NOT_IMPLEMENTED; GKO_INSTANTIATE_FOR_EACH_VALUE_TYPE(GKO_DECLARE_CGS_STEP_1_KERNEL); @@ -93,10 +87,7 @@ void step_2(std::shared_ptr exec, matrix::Dense *t, matrix::Dense *alpha, const matrix::Dense *rho, const matrix::Dense *gamma, - const Array *stop_status) -{ - GKO_NOT_IMPLEMENTED; -} + const Array *stop_status) GKO_NOT_IMPLEMENTED; GKO_INSTANTIATE_FOR_EACH_VALUE_TYPE(GKO_DECLARE_CGS_STEP_2_KERNEL); @@ -105,10 +96,7 @@ void step_3(std::shared_ptr exec, const matrix::Dense *t, const matrix::Dense *u_hat, matrix::Dense *r, matrix::Dense *x, const matrix::Dense *alpha, - const Array *stop_status) -{ - GKO_NOT_IMPLEMENTED; -} + const Array *stop_status) GKO_NOT_IMPLEMENTED; GKO_INSTANTIATE_FOR_EACH_VALUE_TYPE(GKO_DECLARE_CGS_STEP_3_KERNEL); diff --git a/dpcpp/solver/fcg_kernels.dp.cpp b/dpcpp/solver/fcg_kernels.dp.cpp index 79422ed1960..70bff10d329 100644 --- a/dpcpp/solver/fcg_kernels.dp.cpp +++ b/dpcpp/solver/fcg_kernels.dp.cpp @@ -59,10 +59,7 @@ void initialize(std::shared_ptr exec, matrix::Dense *q, matrix::Dense *t, matrix::Dense *prev_rho, matrix::Dense *rho, matrix::Dense *rho_t, - Array *stop_status) -{ - GKO_NOT_IMPLEMENTED; -} + Array *stop_status) GKO_NOT_IMPLEMENTED; GKO_INSTANTIATE_FOR_EACH_VALUE_TYPE(GKO_DECLARE_FCG_INITIALIZE_KERNEL); @@ -72,10 +69,7 @@ void step_1(std::shared_ptr exec, matrix::Dense *p, const matrix::Dense *z, const matrix::Dense *rho_t, const matrix::Dense *prev_rho, - const Array *stop_status) -{ - GKO_NOT_IMPLEMENTED; -} + const Array *stop_status) GKO_NOT_IMPLEMENTED; GKO_INSTANTIATE_FOR_EACH_VALUE_TYPE(GKO_DECLARE_FCG_STEP_1_KERNEL); @@ -87,10 +81,7 @@ void step_2(std::shared_ptr exec, const matrix::Dense *q, const matrix::Dense *beta, const matrix::Dense *rho, - const Array *stop_status) -{ - GKO_NOT_IMPLEMENTED; -} + const Array *stop_status) GKO_NOT_IMPLEMENTED; GKO_INSTANTIATE_FOR_EACH_VALUE_TYPE(GKO_DECLARE_FCG_STEP_2_KERNEL); diff --git a/dpcpp/solver/gmres_kernels.dp.cpp b/dpcpp/solver/gmres_kernels.dp.cpp index c8924b0b67f..8a174cb22f8 100644 --- a/dpcpp/solver/gmres_kernels.dp.cpp +++ b/dpcpp/solver/gmres_kernels.dp.cpp @@ -60,30 +60,22 @@ namespace { template void finish_arnoldi(size_type num_rows, matrix::Dense *krylov_bases, matrix::Dense *hessenberg_iter, size_type iter, - const stopping_status *stop_status) -{ - GKO_NOT_IMPLEMENTED; -} + const stopping_status *stop_status) GKO_NOT_IMPLEMENTED; template void calculate_sin_and_cos(matrix::Dense *givens_sin, matrix::Dense *givens_cos, matrix::Dense *hessenberg_iter, - size_type iter, const size_type rhs) -{ - GKO_NOT_IMPLEMENTED; -} + size_type iter, + const size_type rhs) GKO_NOT_IMPLEMENTED; template void givens_rotation(matrix::Dense *givens_sin, matrix::Dense *givens_cos, matrix::Dense *hessenberg_iter, size_type iter, - const stopping_status *stop_status) -{ - GKO_NOT_IMPLEMENTED; -} + const stopping_status *stop_status) GKO_NOT_IMPLEMENTED; template @@ -91,30 +83,21 @@ void calculate_next_residual_norm( matrix::Dense *givens_sin, matrix::Dense *givens_cos, matrix::Dense> *residual_norm, matrix::Dense *residual_norm_collection, size_type iter, - const stopping_status *stop_status) -{ - GKO_NOT_IMPLEMENTED; -} + const stopping_status *stop_status) GKO_NOT_IMPLEMENTED; template void solve_upper_triangular( const matrix::Dense *residual_norm_collection, const matrix::Dense *hessenberg, matrix::Dense *y, - const size_type *final_iter_nums) -{ - GKO_NOT_IMPLEMENTED; -} + const size_type *final_iter_nums) GKO_NOT_IMPLEMENTED; template void calculate_qy(const matrix::Dense *krylov_bases, const matrix::Dense *y, matrix::Dense *before_preconditioner, - const size_type *final_iter_nums) -{ - GKO_NOT_IMPLEMENTED; -} + const size_type *final_iter_nums) GKO_NOT_IMPLEMENTED; } // namespace @@ -126,10 +109,8 @@ void initialize_1(std::shared_ptr exec, matrix::Dense *residual, matrix::Dense *givens_sin, matrix::Dense *givens_cos, - Array *stop_status, size_type krylov_dim) -{ - GKO_NOT_IMPLEMENTED; -} + Array *stop_status, + size_type krylov_dim) GKO_NOT_IMPLEMENTED; GKO_INSTANTIATE_FOR_EACH_VALUE_TYPE(GKO_DECLARE_GMRES_INITIALIZE_1_KERNEL); @@ -140,10 +121,8 @@ void initialize_2(std::shared_ptr exec, matrix::Dense> *residual_norm, matrix::Dense *residual_norm_collection, matrix::Dense *krylov_bases, - Array *final_iter_nums, size_type krylov_dim) -{ - GKO_NOT_IMPLEMENTED; -} + Array *final_iter_nums, + size_type krylov_dim) GKO_NOT_IMPLEMENTED; GKO_INSTANTIATE_FOR_EACH_VALUE_TYPE(GKO_DECLARE_GMRES_INITIALIZE_2_KERNEL); @@ -157,10 +136,7 @@ void step_1(std::shared_ptr exec, size_type num_rows, matrix::Dense *krylov_bases, matrix::Dense *hessenberg_iter, size_type iter, Array *final_iter_nums, - const Array *stop_status) -{ - GKO_NOT_IMPLEMENTED; -} + const Array *stop_status) GKO_NOT_IMPLEMENTED; GKO_INSTANTIATE_FOR_EACH_VALUE_TYPE(GKO_DECLARE_GMRES_STEP_1_KERNEL); @@ -172,10 +148,7 @@ void step_2(std::shared_ptr exec, const matrix::Dense *hessenberg, matrix::Dense *y, matrix::Dense *before_preconditioner, - const Array *final_iter_nums) -{ - GKO_NOT_IMPLEMENTED; -} + const Array *final_iter_nums) GKO_NOT_IMPLEMENTED; GKO_INSTANTIATE_FOR_EACH_VALUE_TYPE(GKO_DECLARE_GMRES_STEP_2_KERNEL); diff --git a/dpcpp/solver/ir_kernels.dp.cpp b/dpcpp/solver/ir_kernels.dp.cpp index bd3c6a98bc8..03f1150e004 100644 --- a/dpcpp/solver/ir_kernels.dp.cpp +++ b/dpcpp/solver/ir_kernels.dp.cpp @@ -48,10 +48,7 @@ namespace ir { void initialize(std::shared_ptr exec, - Array *stop_status) -{ - GKO_NOT_IMPLEMENTED; -} + Array *stop_status) GKO_NOT_IMPLEMENTED; } // namespace ir diff --git a/dpcpp/solver/lower_trs_kernels.dp.cpp b/dpcpp/solver/lower_trs_kernels.dp.cpp index 6f0546ed6a4..6b6d71b30fd 100644 --- a/dpcpp/solver/lower_trs_kernels.dp.cpp +++ b/dpcpp/solver/lower_trs_kernels.dp.cpp @@ -60,10 +60,7 @@ namespace lower_trs { void should_perform_transpose(std::shared_ptr exec, - bool &do_transpose) -{ - GKO_NOT_IMPLEMENTED; -} + bool &do_transpose) GKO_NOT_IMPLEMENTED; void init_struct(std::shared_ptr exec, @@ -97,10 +94,8 @@ void solve(std::shared_ptr exec, const matrix::Csr *matrix, const solver::SolveStruct *solve_struct, matrix::Dense *trans_b, matrix::Dense *trans_x, - const matrix::Dense *b, matrix::Dense *x) -{ - GKO_NOT_IMPLEMENTED; -} + const matrix::Dense *b, + matrix::Dense *x) GKO_NOT_IMPLEMENTED; GKO_INSTANTIATE_FOR_EACH_VALUE_AND_INDEX_TYPE( GKO_DECLARE_LOWER_TRS_SOLVE_KERNEL); diff --git a/dpcpp/solver/upper_trs_kernels.dp.cpp b/dpcpp/solver/upper_trs_kernels.dp.cpp index 67efc9896f6..2d7afa5dbc8 100644 --- a/dpcpp/solver/upper_trs_kernels.dp.cpp +++ b/dpcpp/solver/upper_trs_kernels.dp.cpp @@ -97,10 +97,8 @@ void solve(std::shared_ptr exec, const matrix::Csr *matrix, const solver::SolveStruct *solve_struct, matrix::Dense *trans_b, matrix::Dense *trans_x, - const matrix::Dense *b, matrix::Dense *x) -{ - GKO_NOT_IMPLEMENTED; -} + const matrix::Dense *b, + matrix::Dense *x) GKO_NOT_IMPLEMENTED; GKO_INSTANTIATE_FOR_EACH_VALUE_AND_INDEX_TYPE( GKO_DECLARE_UPPER_TRS_SOLVE_KERNEL); diff --git a/dpcpp/stop/criterion_kernels.dp.cpp b/dpcpp/stop/criterion_kernels.dp.cpp index dce2c8ae84c..bc59cc1cae8 100644 --- a/dpcpp/stop/criterion_kernels.dp.cpp +++ b/dpcpp/stop/criterion_kernels.dp.cpp @@ -58,7 +58,7 @@ void set_all_statuses(std::shared_ptr exec, stopping_status *__restrict__ stop_status_ptr = stop_status->get_data(); exec->get_queue()->submit([&](sycl::handler &cgh) { cgh.parallel_for(sycl::range<1>{size}, [=](sycl::id<1> idx_id) { - const int idx = idx_id[0]; + const auto idx = idx_id[0]; stop_status_ptr[idx].stop(stoppingId, setFinalized); }); }); diff --git a/dpcpp/stop/residual_norm_kernels.dp.cpp b/dpcpp/stop/residual_norm_kernels.dp.cpp index f0f205d438d..e42861c38bc 100644 --- a/dpcpp/stop/residual_norm_kernels.dp.cpp +++ b/dpcpp/stop/residual_norm_kernels.dp.cpp @@ -62,10 +62,7 @@ void residual_norm(std::shared_ptr exec, ValueType rel_residual_goal, uint8 stoppingId, bool setFinalized, Array *stop_status, Array *device_storage, bool *all_converged, - bool *one_changed) -{ - GKO_NOT_IMPLEMENTED; -} + bool *one_changed) GKO_NOT_IMPLEMENTED; GKO_INSTANTIATE_FOR_EACH_NON_COMPLEX_VALUE_TYPE( GKO_DECLARE_RESIDUAL_NORM_KERNEL); diff --git a/include/ginkgo/core/base/math.hpp b/include/ginkgo/core/base/math.hpp index 9217e00f800..c210718f13c 100644 --- a/include/ginkgo/core/base/math.hpp +++ b/include/ginkgo/core/base/math.hpp @@ -914,6 +914,11 @@ GKO_INLINE GKO_ATTRIBUTES constexpr xstd::enable_if_t::value, abs(const T &x) { #ifdef CL_SYCL_LANGUAGE_VERSION + // FIXME: This implementation is due to two DPC++ issues + // 1) plain `sqrt` call evaluates to `std::sqrt` which fails on GPUs + // 2) DPC++ GPUs require the OpenCL backend (`SYCL_BE=OPENCL`) to support + // complex multiplication, on top of extra binary objects. For now, we + // avoid this issue with the current implementation. return cl::sycl::sqrt(real(x) * real(x) + imag(x) * imag(x)); #else return sqrt(squared_norm(x)); diff --git a/include/ginkgo/core/reorder/reordering_base.hpp b/include/ginkgo/core/reorder/reordering_base.hpp index e4ee5c9e8f1..7b93d575813 100644 --- a/include/ginkgo/core/reorder/reordering_base.hpp +++ b/include/ginkgo/core/reorder/reordering_base.hpp @@ -1,5 +1,5 @@ /************************************************************* -Copyright (c) 2017-2019, the Ginkgo authors +Copyright (c) 2017-2020, the Ginkgo authors All rights reserved. Redistribution and use in source and binary forms, with or without diff --git a/omp/get_info.cmake b/omp/get_info.cmake index 4fe24965649..3dc866e13fc 100644 --- a/omp/get_info.cmake +++ b/omp/get_info.cmake @@ -4,4 +4,6 @@ ginkgo_print_variable(${detailed_log} "OpenMP_CXX_LIB_NAMES") ginkgo_print_variable(${detailed_log} "OpenMP_CXX_LIBRARIES") ginkgo_print_module_footer(${detailed_log} "OMP variables:") ginkgo_print_variable(${detailed_log} "GINKGO_COMPILER_FLAGS") +ginkgo_print_module_footer(${detailed_log} "OMP environment variables:") +ginkgo_print_env_variable(${detailed_log} "OMP_NUM_THREADS") ginkgo_print_module_footer(${detailed_log} "") diff --git a/omp/test/reorder/rcm_kernels.cpp b/omp/test/reorder/rcm_kernels.cpp index 8fcad63a4f4..9f7a891cb89 100644 --- a/omp/test/reorder/rcm_kernels.cpp +++ b/omp/test/reorder/rcm_kernels.cpp @@ -1,5 +1,5 @@ /************************************************************* -Copyright (c) 2017-2019, the Ginkgo authors +Copyright (c) 2017-2020, the Ginkgo authors All rights reserved. Redistribution and use in source and binary forms, with or without diff --git a/reference/test/reorder/rcm.cpp b/reference/test/reorder/rcm.cpp index b8358f4edd3..c81045c8fd6 100644 --- a/reference/test/reorder/rcm.cpp +++ b/reference/test/reorder/rcm.cpp @@ -1,5 +1,5 @@ /************************************************************* -Copyright (c) 2017-2019, the Ginkgo authors +Copyright (c) 2017-2020, the Ginkgo authors All rights reserved. Redistribution and use in source and binary forms, with or without