diff --git a/common/cuda_hip/base/blas_bindings.hpp b/common/cuda_hip/base/blas_bindings.hpp index 5df6282adee..facc1a2eb56 100644 --- a/common/cuda_hip/base/blas_bindings.hpp +++ b/common/cuda_hip/base/blas_bindings.hpp @@ -13,7 +13,7 @@ #define BLAS_OP_T HIPBLAS_OP_T #define BLAS_OP_C HIPBLAS_OP_C #else // GKO_COMPILING_CUDA -#include "cuda/base/cublas_bindings.cuh" +#include "cuda/base/cublas_bindings.hpp" #define BLAS_OP_N CUBLAS_OP_N #define BLAS_OP_T CUBLAS_OP_T diff --git a/common/cuda_hip/base/kernel_launch.hpp b/common/cuda_hip/base/kernel_launch.hpp index 3d59e145a86..cdfb42cdd29 100644 --- a/common/cuda_hip/base/kernel_launch.hpp +++ b/common/cuda_hip/base/kernel_launch.hpp @@ -8,7 +8,7 @@ #endif -#include +#include #include diff --git a/common/cuda_hip/base/pointer_mode_guard.hpp b/common/cuda_hip/base/pointer_mode_guard.hpp index 4b2f0e5bbd1..382fd85ca76 100644 --- a/common/cuda_hip/base/pointer_mode_guard.hpp +++ b/common/cuda_hip/base/pointer_mode_guard.hpp @@ -12,7 +12,7 @@ #ifdef GKO_COMPILING_HIP #include "hip/base/pointer_mode_guard.hip.hpp" #else // GKO_COMPILING_CUDA -#include "cuda/base/pointer_mode_guard.cuh" +#include "common/cuda_hip/base/pointer_mode_guard.hpp" #endif diff --git a/common/cuda_hip/base/randlib_bindings.hpp b/common/cuda_hip/base/randlib_bindings.hpp index 3deb43e2721..6207723b806 100644 --- a/common/cuda_hip/base/randlib_bindings.hpp +++ b/common/cuda_hip/base/randlib_bindings.hpp @@ -11,7 +11,7 @@ #define RANDLIB_RNG_PSEUDO_DEFAULT HIPRAND_RNG_PSEUDO_DEFAULT #else // GKO_COMPILING_CUDA -#include "cuda/base/curand_bindings.cuh" +#include "common/cuda_hip/base/curand_bindings.hpp" #define RANDLIB_RNG_PSEUDO_DEFAULT CURAND_RNG_PSEUDO_DEFAULT #endif diff --git a/common/cuda_hip/base/sparselib_bindings.hpp b/common/cuda_hip/base/sparselib_bindings.hpp index a700bfa419c..b2d6cf2d398 100644 --- a/common/cuda_hip/base/sparselib_bindings.hpp +++ b/common/cuda_hip/base/sparselib_bindings.hpp @@ -9,7 +9,7 @@ #ifdef GKO_COMPILING_HIP #include "hip/base/hipsparse_bindings.hip.hpp" #else // GKO_COMPILING_CUDA -#include "cuda/base/cusparse_bindings.cuh" +#include "common/cuda_hip/base/cusparse_bindings.hpp" #endif diff --git a/common/cuda_hip/components/format_conversion.hpp b/common/cuda_hip/components/format_conversion.hpp index a16d09b2e3a..f0a78a68d45 100644 --- a/common/cuda_hip/components/format_conversion.hpp +++ b/common/cuda_hip/components/format_conversion.hpp @@ -9,7 +9,7 @@ #ifdef GKO_COMPILING_HIP #include "hip/components/format_conversion.hip.hpp" #else // GKO_COMPILING_CUDA -#include "cuda/components/format_conversion.cuh" +#include "common/cuda_hip/components/format_conversion.hpp" #endif diff --git a/common/cuda_hip/components/memory.hpp b/common/cuda_hip/components/memory.hpp index 974431e2fb8..d1d855f461d 100644 --- a/common/cuda_hip/components/memory.hpp +++ b/common/cuda_hip/components/memory.hpp @@ -9,7 +9,7 @@ #ifdef GKO_COMPILING_HIP #include "hip/components/memory.hip.hpp" #else // GKO_COMPILING_CUDA -#include "cuda/components/memory.cuh" +#include "common/cuda_hip/components/memory.hpp" #endif diff --git a/common/cuda_hip/matrix/ell_kernels.cpp b/common/cuda_hip/matrix/ell_kernels.cpp index 96e9dac9d78..8526df515b0 100644 --- a/common/cuda_hip/matrix/ell_kernels.cpp +++ b/common/cuda_hip/matrix/ell_kernels.cpp @@ -2,7 +2,7 @@ // // SPDX-License-Identifier: BSD-3-Clause -#include +#include #include diff --git a/cuda/matrix/fbcsr_kernels.instantiate.cu b/common/cuda_hip/matrix/fbcsr_kernels.instantiate.cpp similarity index 90% rename from cuda/matrix/fbcsr_kernels.instantiate.cu rename to common/cuda_hip/matrix/fbcsr_kernels.instantiate.cpp index f6165ac5e5c..a3beaac4a85 100644 --- a/cuda/matrix/fbcsr_kernels.instantiate.cu +++ b/common/cuda_hip/matrix/fbcsr_kernels.instantiate.cpp @@ -2,12 +2,12 @@ // // SPDX-License-Identifier: BSD-3-Clause -#include "cuda/matrix/fbcsr_kernels.template.cu" +#include "common/cuda_hip/matrix/fbcsr_kernels.template.cpp" namespace gko { namespace kernels { -namespace cuda { +namespace GKO_DEVICE_NAMESPACE { /** * @brief The fixed-size block compressed sparse row matrix format namespace. * @@ -42,6 +42,6 @@ GKO_INSTANTIATE_FOR_EACH_VALUE_AND_INDEX_TYPE( } // namespace fbcsr -} // namespace cuda +} // namespace GKO_DEVICE_NAMESPACE } // namespace kernels } // namespace gko diff --git a/common/cuda_hip/matrix/fbcsr_kernels.cpp b/common/cuda_hip/matrix/fbcsr_kernels.template.cpp similarity index 100% rename from common/cuda_hip/matrix/fbcsr_kernels.cpp rename to common/cuda_hip/matrix/fbcsr_kernels.template.cpp diff --git a/common/cuda_hip/matrix/sparsity_csr_kernels.cpp b/common/cuda_hip/matrix/sparsity_csr_kernels.cpp index 540722d843c..869925480b8 100644 --- a/common/cuda_hip/matrix/sparsity_csr_kernels.cpp +++ b/common/cuda_hip/matrix/sparsity_csr_kernels.cpp @@ -2,7 +2,7 @@ // // SPDX-License-Identifier: BSD-3-Clause -#include +#include #include diff --git a/common/cuda_hip/solver/cb_gmres_kernels.cpp b/common/cuda_hip/solver/cb_gmres_kernels.cpp index 9be99c094fc..64fb0248954 100644 --- a/common/cuda_hip/solver/cb_gmres_kernels.cpp +++ b/common/cuda_hip/solver/cb_gmres_kernels.cpp @@ -2,7 +2,7 @@ // // SPDX-License-Identifier: BSD-3-Clause -#include +#include #include diff --git a/common/unified/base/kernel_launch.hpp b/common/unified/base/kernel_launch.hpp index b32572546f0..04db86c27e5 100644 --- a/common/unified/base/kernel_launch.hpp +++ b/common/unified/base/kernel_launch.hpp @@ -19,7 +19,7 @@ #define GKO_DEVICE_NAMESPACE cuda #define GKO_KERNEL __device__ -#include "cuda/base/types.hpp" +#include "common/cuda_hip/base/types.hpp" namespace gko { @@ -271,7 +271,7 @@ typename to_device_type_impl::type map_to_device(T&& param) #if defined(GKO_COMPILING_CUDA) -#include "cuda/base/kernel_launch.cuh" +#include "common/cuda_hip/base/kernel_launch.hpp" #elif defined(GKO_COMPILING_HIP) #include "hip/base/kernel_launch.hip.hpp" #elif defined(GKO_COMPILING_DPCPP) diff --git a/common/unified/base/kernel_launch_reduction.hpp b/common/unified/base/kernel_launch_reduction.hpp index c3158d35a1c..a9b738f2ca9 100644 --- a/common/unified/base/kernel_launch_reduction.hpp +++ b/common/unified/base/kernel_launch_reduction.hpp @@ -20,7 +20,7 @@ #if defined(GKO_COMPILING_CUDA) -#include "cuda/base/kernel_launch_reduction.cuh" +#include "common/cuda_hip/base/kernel_launch_reduction.hpp" #elif defined(GKO_COMPILING_HIP) #include "hip/base/kernel_launch_reduction.hip.hpp" #elif defined(GKO_COMPILING_DPCPP) diff --git a/common/unified/base/kernel_launch_solver.hpp b/common/unified/base/kernel_launch_solver.hpp index f4240805c64..11f282e5c09 100644 --- a/common/unified/base/kernel_launch_solver.hpp +++ b/common/unified/base/kernel_launch_solver.hpp @@ -108,7 +108,7 @@ const device_type* row_vector(const matrix::Dense* mtx) #if defined(GKO_COMPILING_CUDA) -#include "cuda/base/kernel_launch_solver.cuh" +#include "common/cuda_hip/base/kernel_launch_solver.hpp" #elif defined(GKO_COMPILING_HIP) #include "hip/base/kernel_launch_solver.hip.hpp" #elif defined(GKO_COMPILING_DPCPP) diff --git a/cuda/base/config.hpp b/cuda/base/config.hpp index 44c304bde5d..8de05a2b93e 100644 --- a/cuda/base/config.hpp +++ b/cuda/base/config.hpp @@ -9,7 +9,7 @@ #include -#include "cuda/base/math.hpp" +#include "common/cuda_hip/base/math.hpp" namespace gko { diff --git a/cuda/base/cublas_bindings.hpp b/cuda/base/cublas_bindings.hpp index b1f22702cd3..b9ede7715f2 100644 --- a/cuda/base/cublas_bindings.hpp +++ b/cuda/base/cublas_bindings.hpp @@ -12,7 +12,7 @@ #include -#include "cuda/base/math.hpp" +#include "common/cuda_hip/base/math.hpp" #include "cuda/base/types.hpp" diff --git a/cuda/base/curand_bindings.hpp b/cuda/base/curand_bindings.hpp index a37611b2a7a..cb74033e257 100644 --- a/cuda/base/curand_bindings.hpp +++ b/cuda/base/curand_bindings.hpp @@ -12,7 +12,7 @@ #include -#include "cuda/base/math.hpp" +#include "common/cuda_hip/base/math.hpp" #include "cuda/base/types.hpp" diff --git a/cuda/components/format_conversion.cuh b/cuda/components/format_conversion.cuh index bccc927c9cd..00e7f7b2942 100644 --- a/cuda/components/format_conversion.cuh +++ b/cuda/components/format_conversion.cuh @@ -11,7 +11,7 @@ #include "cuda/components/cooperative_groups.cuh" -#include "cuda/components/thread_ids.cuh" +#include "common/cuda_hip/components/thread_ids.hpp" #ifdef GINKGO_BENCHMARK_ENABLE_TUNING diff --git a/cuda/matrix/csr_kernels.template.cu b/cuda/matrix/csr_kernels.template.cu index 8676e1438bc..45d0257e508 100644 --- a/cuda/matrix/csr_kernels.template.cu +++ b/cuda/matrix/csr_kernels.template.cu @@ -40,20 +40,20 @@ #include "core/synthesizer/implementation_selection.hpp" #include "cuda/base/config.hpp" #include "cuda/base/cusparse_bindings.hpp" -#include "cuda/base/math.hpp" +#include "common/cuda_hip/base/math.hpp" #include "cuda/base/pointer_mode_guard.hpp" #include "cuda/base/thrust.cuh" #include "cuda/base/types.hpp" -#include "cuda/components/atomic.cuh" +#include "common/cuda_hip/components/atomic.hpp" #include "cuda/components/cooperative_groups.cuh" #include "cuda/components/format_conversion.cuh" -#include "cuda/components/intrinsics.cuh" -#include "cuda/components/merging.cuh" -#include "cuda/components/prefix_sum.cuh" -#include "cuda/components/reduction.cuh" -#include "cuda/components/segment_scan.cuh" -#include "cuda/components/thread_ids.cuh" -#include "cuda/components/uninitialized_array.hpp" +#include "common/cuda_hip/components/intrinsics.hpp" +#include "common/cuda_hip/components/merging.hpp" +#include "common/cuda_hip/components/prefix_sum.hpp" +#include "common/cuda_hip/components/reduction.hpp" +#include "common/cuda_hip/components/segment_scan.hpp" +#include "common/cuda_hip/components/thread_ids.hpp" +#include "common/cuda_hip/components/uninitialized_array.hpp" namespace gko { diff --git a/cuda/solver/common_trs_kernels.cuh b/cuda/solver/common_trs_kernels.cuh index dc20927d9b5..1e33343559e 100644 --- a/cuda/solver/common_trs_kernels.cuh +++ b/cuda/solver/common_trs_kernels.cuh @@ -24,13 +24,13 @@ #include "core/matrix/dense_kernels.hpp" #include "core/synthesizer/implementation_selection.hpp" #include "cuda/base/cusparse_bindings.hpp" -#include "cuda/base/math.hpp" +#include "common/cuda_hip/base/math.hpp" #include "cuda/base/pointer_mode_guard.hpp" #include "cuda/base/types.hpp" -#include "cuda/components/atomic.cuh" +#include "common/cuda_hip/components/atomic.hpp" #include "cuda/components/memory.cuh" -#include "cuda/components/thread_ids.cuh" -#include "cuda/components/uninitialized_array.hpp" +#include "common/cuda_hip/components/thread_ids.hpp" +#include "common/cuda_hip/components/uninitialized_array.hpp" namespace gko { diff --git a/cuda/solver/lower_trs_kernels.cu b/cuda/solver/lower_trs_kernels.cu index 46b4cb4c2e4..c569086e01c 100644 --- a/cuda/solver/lower_trs_kernels.cu +++ b/cuda/solver/lower_trs_kernels.cu @@ -18,7 +18,7 @@ #include "cuda/base/cusparse_bindings.hpp" -#include "cuda/base/math.hpp" +#include "common/cuda_hip/base/math.hpp" #include "cuda/base/types.hpp" #include "cuda/solver/common_trs_kernels.cuh" diff --git a/cuda/solver/upper_trs_kernels.cu b/cuda/solver/upper_trs_kernels.cu index a8ee5f77cca..71017204d7b 100644 --- a/cuda/solver/upper_trs_kernels.cu +++ b/cuda/solver/upper_trs_kernels.cu @@ -18,7 +18,7 @@ #include "cuda/base/cusparse_bindings.hpp" -#include "cuda/base/math.hpp" +#include "common/cuda_hip/base/math.hpp" #include "cuda/base/types.hpp" #include "cuda/solver/common_trs_kernels.cuh" diff --git a/cuda/stop/criterion_kernels.cu b/cuda/stop/criterion_kernels.cu index 17bcbbc1567..61ea4e04baa 100644 --- a/cuda/stop/criterion_kernels.cu +++ b/cuda/stop/criterion_kernels.cu @@ -10,9 +10,9 @@ #include -#include "cuda/base/math.hpp" +#include "common/cuda_hip/base/math.hpp" #include "cuda/base/types.hpp" -#include "cuda/components/thread_ids.cuh" +#include "common/cuda_hip/components/thread_ids.hpp" namespace gko { diff --git a/cuda/stop/residual_norm_kernels.cu b/cuda/stop/residual_norm_kernels.cu index 18102d91ec5..2c15cd94b05 100644 --- a/cuda/stop/residual_norm_kernels.cu +++ b/cuda/stop/residual_norm_kernels.cu @@ -11,9 +11,9 @@ #include "core/base/array_access.hpp" -#include "cuda/base/math.hpp" +#include "common/cuda_hip/base/math.hpp" #include "cuda/base/types.hpp" -#include "cuda/components/thread_ids.cuh" +#include "common/cuda_hip/components/thread_ids.hpp" namespace gko { diff --git a/cuda/test/base/math.cu b/cuda/test/base/math.cu index c7f70fe3011..f5f994a17c2 100644 --- a/cuda/test/base/math.cu +++ b/cuda/test/base/math.cu @@ -17,7 +17,7 @@ #include -#include "cuda/base/math.hpp" +#include "common/cuda_hip/base/math.hpp" #include "cuda/base/types.hpp" #include "cuda/test/utils.hpp" diff --git a/cuda/test/components/merging.cu b/cuda/test/components/merging.cu index 6ef7d3ab3c4..7fadc2cc4bf 100644 --- a/cuda/test/components/merging.cu +++ b/cuda/test/components/merging.cu @@ -2,7 +2,7 @@ // // SPDX-License-Identifier: BSD-3-Clause -#include "cuda/components/merging.cuh" +#include "common/cuda_hip/components/merging.hpp" #include diff --git a/cuda/test/components/searching.cu b/cuda/test/components/searching.cu index 0eeb383c05c..3633fce55e7 100644 --- a/cuda/test/components/searching.cu +++ b/cuda/test/components/searching.cu @@ -2,7 +2,7 @@ // // SPDX-License-Identifier: BSD-3-Clause -#include "cuda/components/searching.cuh" +#include "common/cuda_hip/components/searching.hpp" #include diff --git a/cuda/test/components/sorting.cu b/cuda/test/components/sorting.cu index 19c7daab782..d7a48fdab04 100644 --- a/cuda/test/components/sorting.cu +++ b/cuda/test/components/sorting.cu @@ -2,7 +2,7 @@ // // SPDX-License-Identifier: BSD-3-Clause -#include "cuda/components/sorting.cuh" +#include "common/cuda_hip/components/sorting.hpp" #include