From 0779e7945f354bef3081f14e706b637705607c23 Mon Sep 17 00:00:00 2001 From: Daniel Arndt Date: Fri, 2 Nov 2018 20:33:25 +0100 Subject: [PATCH] Let ArrayView have a MemorySpace template parameter --- include/deal.II/base/partitioner.h | 35 ++++---- include/deal.II/base/partitioner.templates.h | 34 ++++---- .../lac/la_parallel_vector.templates.h | 81 +++++++++++-------- source/base/partitioner.cuda.inst.in | 31 +++---- source/base/partitioner.inst.in | 23 +++--- tests/cuda/parallel_partitioner_06.cu | 28 ++++--- tests/cuda/parallel_partitioner_07.cu | 40 ++++----- tests/cuda/parallel_partitioner_08.cu | 18 +++-- 8 files changed, 156 insertions(+), 134 deletions(-) diff --git a/include/deal.II/base/partitioner.h b/include/deal.II/base/partitioner.h index ec65709bb5..440ada1941 100644 --- a/include/deal.II/base/partitioner.h +++ b/include/deal.II/base/partitioner.h @@ -423,11 +423,11 @@ namespace Utilities template void export_to_ghosted_array_start( - const unsigned int communication_channel, - const ArrayView &locally_owned_array, - const ArrayView & temporary_storage, - const ArrayView & ghost_array, - std::vector & requests) const; + const unsigned int communication_channel, + const ArrayView &locally_owned_array, + const ArrayView & temporary_storage, + const ArrayView & ghost_array, + std::vector & requests) const; /** * Finish the exports of the data in a locally owned array to the range @@ -448,8 +448,9 @@ namespace Utilities */ template void - export_to_ghosted_array_finish(const ArrayView & ghost_array, - std::vector &requests) const; + export_to_ghosted_array_finish( + const ArrayView &ghost_array, + std::vector & requests) const; /** * Start importing the data on an array indexed by the ghost indices of @@ -490,11 +491,11 @@ namespace Utilities template void import_from_ghosted_array_start( - const VectorOperation::values vector_operation, - const unsigned int communication_channel, - const ArrayView & ghost_array, - const ArrayView & temporary_storage, - std::vector & requests) const; + const VectorOperation::values vector_operation, + const unsigned int communication_channel, + const ArrayView &ghost_array, + const ArrayView &temporary_storage, + std::vector & requests) const; /** * Finish importing the data from an array indexed by the ghost @@ -533,11 +534,11 @@ namespace Utilities template void import_from_ghosted_array_finish( - const VectorOperation::values vector_operation, - const ArrayView &temporary_storage, - const ArrayView & locally_owned_storage, - const ArrayView & ghost_array, - std::vector & requests) const; + const VectorOperation::values vector_operation, + const ArrayView &temporary_storage, + const ArrayView & locally_owned_storage, + const ArrayView & ghost_array, + std::vector & requests) const; #endif /** diff --git a/include/deal.II/base/partitioner.templates.h b/include/deal.II/base/partitioner.templates.h index 5e9d5cc8e4..f21e4e65c2 100644 --- a/include/deal.II/base/partitioner.templates.h +++ b/include/deal.II/base/partitioner.templates.h @@ -40,11 +40,11 @@ namespace Utilities template void Partitioner::export_to_ghosted_array_start( - const unsigned int communication_channel, - const ArrayView &locally_owned_array, - const ArrayView & temporary_storage, - const ArrayView & ghost_array, - std::vector & requests) const + const unsigned int communication_channel, + const ArrayView &locally_owned_array, + const ArrayView & temporary_storage, + const ArrayView & ghost_array, + std::vector & requests) const { AssertDimension(temporary_storage.size(), n_import_indices()); Assert(ghost_array.size() == n_ghost_indices() || @@ -153,8 +153,8 @@ namespace Utilities template void Partitioner::export_to_ghosted_array_finish( - const ArrayView & ghost_array, - std::vector &requests) const + const ArrayView &ghost_array, + std::vector & requests) const { Assert(ghost_array.size() == n_ghost_indices() || ghost_array.size() == n_ghost_indices_in_larger_set, @@ -238,11 +238,11 @@ namespace Utilities template void Partitioner::import_from_ghosted_array_start( - const VectorOperation::values vector_operation, - const unsigned int communication_channel, - const ArrayView & ghost_array, - const ArrayView & temporary_storage, - std::vector & requests) const + const VectorOperation::values vector_operation, + const unsigned int communication_channel, + const ArrayView &ghost_array, + const ArrayView &temporary_storage, + std::vector & requests) const { AssertDimension(temporary_storage.size(), n_import_indices()); Assert(ghost_array.size() == n_ghost_indices() || @@ -464,11 +464,11 @@ namespace Utilities template void Partitioner::import_from_ghosted_array_finish( - const VectorOperation::values vector_operation, - const ArrayView &temporary_storage, - const ArrayView & locally_owned_array, - const ArrayView & ghost_array, - std::vector & requests) const + const VectorOperation::values vector_operation, + const ArrayView &temporary_storage, + const ArrayView & locally_owned_array, + const ArrayView & ghost_array, + std::vector & requests) const { AssertDimension(temporary_storage.size(), n_import_indices()); Assert(ghost_array.size() == n_ghost_indices() || diff --git a/include/deal.II/lac/la_parallel_vector.templates.h b/include/deal.II/lac/la_parallel_vector.templates.h index d6d7295389..84eb0b5f92 100644 --- a/include/deal.II/lac/la_parallel_vector.templates.h +++ b/include/deal.II/lac/la_parallel_vector.templates.h @@ -944,19 +944,21 @@ namespace LinearAlgebra partitioner->import_from_ghosted_array_start( operation, counter, - ArrayView(data.values.get() + partitioner->local_size(), - partitioner->n_ghost_indices()), - ArrayView(import_data.values.get(), - partitioner->n_import_indices()), + ArrayView(data.values.get() + + partitioner->local_size(), + partitioner->n_ghost_indices()), + ArrayView(import_data.values.get(), + partitioner->n_import_indices()), compress_requests); # else partitioner->import_from_ghosted_array_start( operation, counter, - ArrayView(data.values_dev.get() + partitioner->local_size(), - partitioner->n_ghost_indices()), - ArrayView(import_data.values_dev.get(), - partitioner->n_import_indices()), + ArrayView(data.values_dev.get() + + partitioner->local_size(), + partitioner->n_ghost_indices()), + ArrayView(import_data.values_dev.get(), + partitioner->n_import_indices()), compress_requests); # endif #endif @@ -985,11 +987,13 @@ namespace LinearAlgebra ExcNotInitialized()); partitioner->import_from_ghosted_array_finish( operation, - ArrayView(import_data.values.get(), - partitioner->n_import_indices()), - ArrayView(data.values.get(), partitioner->local_size()), - ArrayView(data.values.get() + partitioner->local_size(), - partitioner->n_ghost_indices()), + ArrayView( + import_data.values.get(), partitioner->n_import_indices()), + ArrayView(data.values.get(), + partitioner->local_size()), + ArrayView(data.values.get() + + partitioner->local_size(), + partitioner->n_ghost_indices()), compress_requests); # else Assert(partitioner->n_import_indices() == 0 || @@ -997,11 +1001,13 @@ namespace LinearAlgebra ExcNotInitialized()); partitioner->import_from_ghosted_array_finish( operation, - ArrayView(import_data.values_dev.get(), - partitioner->n_import_indices()), - ArrayView(data.values_dev.get(), partitioner->local_size()), - ArrayView(data.values_dev.get() + partitioner->local_size(), - partitioner->n_ghost_indices()), + ArrayView( + import_data.values_dev.get(), partitioner->n_import_indices()), + ArrayView(data.values_dev.get(), + partitioner->local_size()), + ArrayView(data.values_dev.get() + + partitioner->local_size(), + partitioner->n_ghost_indices()), compress_requests); # endif @@ -1095,23 +1101,26 @@ namespace LinearAlgebra # if !(defined(DEAL_II_COMPILER_CUDA_AWARE) && \ defined(DEAL_II_WITH_CUDA_AWARE_MPI)) - partitioner->export_to_ghosted_array_start( + partitioner->export_to_ghosted_array_start( counter, - ArrayView(data.values.get(), partitioner->local_size()), - ArrayView(import_data.values.get(), - partitioner->n_import_indices()), - ArrayView(data.values.get() + partitioner->local_size(), - partitioner->n_ghost_indices()), + ArrayView(data.values.get(), + partitioner->local_size()), + ArrayView(import_data.values.get(), + partitioner->n_import_indices()), + ArrayView(data.values.get() + + partitioner->local_size(), + partitioner->n_ghost_indices()), update_ghost_values_requests); # else - partitioner->export_to_ghosted_array_start( + partitioner->export_to_ghosted_array_start( counter, - ArrayView(data.values_dev.get(), - partitioner->local_size()), - ArrayView(import_data.values_dev.get(), - partitioner->n_import_indices()), - ArrayView(data.values_dev.get() + partitioner->local_size(), - partitioner->n_ghost_indices()), + ArrayView(data.values_dev.get(), + partitioner->local_size()), + ArrayView(import_data.values_dev.get(), + partitioner->n_import_indices()), + ArrayView(data.values_dev.get() + + partitioner->local_size(), + partitioner->n_ghost_indices()), update_ghost_values_requests); # endif @@ -1140,13 +1149,15 @@ namespace LinearAlgebra # if !(defined(DEAL_II_COMPILER_CUDA_AWARE) && \ defined(DEAL_II_WITH_CUDA_AWARE_MPI)) partitioner->export_to_ghosted_array_finish( - ArrayView(data.values.get() + partitioner->local_size(), - partitioner->n_ghost_indices()), + ArrayView( + data.values.get() + partitioner->local_size(), + partitioner->n_ghost_indices()), update_ghost_values_requests); # else partitioner->export_to_ghosted_array_finish( - ArrayView(data.values_dev.get() + partitioner->local_size(), - partitioner->n_ghost_indices()), + ArrayView( + data.values_dev.get() + partitioner->local_size(), + partitioner->n_ghost_indices()), update_ghost_values_requests); # endif } diff --git a/source/base/partitioner.cuda.inst.in b/source/base/partitioner.cuda.inst.in index d6b6ab7f6c..f674beac2b 100644 --- a/source/base/partitioner.cuda.inst.in +++ b/source/base/partitioner.cuda.inst.in @@ -18,32 +18,33 @@ for (SCALAR : MPI_SCALARS) { #ifdef DEAL_II_WITH_MPI - template void Utilities::MPI::Partitioner::export_to_ghosted_array_start< - SCALAR, - dealii::MemorySpace::CUDA>(const unsigned int, - const ArrayView &, - const ArrayView &, - const ArrayView &, - std::vector &) const; - template void Utilities::MPI::Partitioner:: - export_to_ghosted_array_finish( - const ArrayView &, std::vector &) const; + export_to_ghosted_array_start( + const unsigned int, + const ArrayView &, + const ArrayView &, + const ArrayView &, + std::vector &) const; + + template void Utilities::MPI::Partitioner::export_to_ghosted_array_finish< + SCALAR, + MemorySpace::CUDA>(const ArrayView &, + std::vector &) const; template void Utilities::MPI::Partitioner::import_from_ghosted_array_start< SCALAR, MemorySpace::CUDA>(const VectorOperation::values, const unsigned int, - const ArrayView &, - const ArrayView &, + const ArrayView &, + const ArrayView &, std::vector &) const; template void Utilities::MPI::Partitioner::import_from_ghosted_array_finish< SCALAR, MemorySpace::CUDA>(const VectorOperation::values, - const ArrayView &, - const ArrayView &, - const ArrayView &, + const ArrayView &, + const ArrayView &, + const ArrayView &, std::vector &) const; #endif } diff --git a/source/base/partitioner.inst.in b/source/base/partitioner.inst.in index c1b1696302..84bd8252cc 100644 --- a/source/base/partitioner.inst.in +++ b/source/base/partitioner.inst.in @@ -21,26 +21,27 @@ for (SCALAR : MPI_SCALARS) template void Utilities::MPI::Partitioner::export_to_ghosted_array_start< SCALAR, MemorySpace::Host>(const unsigned int, - const ArrayView &, - const ArrayView &, - const ArrayView &, + const ArrayView &, + const ArrayView &, + const ArrayView &, + std::vector &) const; + template void Utilities::MPI::Partitioner::export_to_ghosted_array_finish< + SCALAR, + MemorySpace::Host>(const ArrayView &, std::vector &) const; - template void Utilities::MPI::Partitioner:: - export_to_ghosted_array_finish( - const ArrayView &, std::vector &) const; template void Utilities::MPI::Partitioner::import_from_ghosted_array_start< SCALAR, MemorySpace::Host>(const VectorOperation::values, const unsigned int, - const ArrayView &, - const ArrayView &, + const ArrayView &, + const ArrayView &, std::vector &) const; template void Utilities::MPI::Partitioner::import_from_ghosted_array_finish< SCALAR, MemorySpace::Host>(const VectorOperation::values, - const ArrayView &, - const ArrayView &, - const ArrayView &, + const ArrayView &, + const ArrayView &, + const ArrayView &, std::vector &) const; #endif } diff --git a/tests/cuda/parallel_partitioner_06.cu b/tests/cuda/parallel_partitioner_06.cu index 14c1096346..c787091f57 100644 --- a/tests/cuda/parallel_partitioner_06.cu +++ b/tests/cuda/parallel_partitioner_06.cu @@ -28,7 +28,7 @@ template void -print_cuda_view(const ArrayView cuda_view) +print_cuda_view(const ArrayView cuda_view) { std::vector cpu_values(cuda_view.size()); Utilities::CUDA::copy_to_host(cuda_view.data(), cpu_values); @@ -110,8 +110,8 @@ test() nullptr, Utilities::CUDA::delete_device_data); locally_owned_data.reset( Utilities::CUDA::allocate_device_data(local_size)); - ArrayView locally_owned_data_view(locally_owned_data.get(), - local_size); + ArrayView locally_owned_data_view( + locally_owned_data.get(), local_size); Utilities::CUDA::copy_to_dev(cpu_locally_owned_data, locally_owned_data.get()); @@ -120,14 +120,15 @@ test() nullptr, Utilities::CUDA::delete_device_data); ghosts.reset( Utilities::CUDA::allocate_device_data(v.n_ghost_indices())); - ArrayView ghosts_view(ghosts.get(), v.n_ghost_indices()); + ArrayView ghosts_view(ghosts.get(), + v.n_ghost_indices()); std::unique_ptr temp_array( nullptr, Utilities::CUDA::delete_device_data); temp_array.reset( Utilities::CUDA::allocate_device_data(v.n_import_indices())); - ArrayView temp_array_view(temp_array.get(), - v.n_import_indices()); + ArrayView temp_array_view( + temp_array.get(), v.n_import_indices()); std::vector requests; @@ -147,8 +148,8 @@ test() AssertCuda(cuda_error); Assert(temp_array_view.size() >= w.n_import_indices(), ExcInternalError()); - ArrayView temp_array_view_w(temp_array_view.data(), - w.n_import_indices()); + ArrayView temp_array_view_w( + temp_array_view.data(), w.n_import_indices()); w.export_to_ghosted_array_start( 3, locally_owned_data_view, temp_array_view_w, ghosts_view, requests); @@ -158,14 +159,15 @@ test() nullptr, Utilities::CUDA::delete_device_data); temp_array2.reset( Utilities::CUDA::allocate_device_data(x.n_import_indices())); - ArrayView temp_array2_view(temp_array2.get(), - x.n_import_indices()); + ArrayView temp_array2_view( + temp_array2.get(), x.n_import_indices()); std::unique_ptr ghosts2( nullptr, Utilities::CUDA::delete_device_data); ghosts2.reset( Utilities::CUDA::allocate_device_data(x.n_ghost_indices())); - ArrayView ghosts2_view(ghosts2.get(), x.n_ghost_indices()); + ArrayView ghosts2_view(ghosts2.get(), + x.n_ghost_indices()); std::vector requests2; x.export_to_ghosted_array_start( @@ -182,8 +184,8 @@ test() AssertCuda(cuda_error); Assert(temp_array_view.size() >= x.n_import_indices(), ExcInternalError()); - ArrayView temp_array_view_x(temp_array_view.data(), - x.n_import_indices()); + ArrayView temp_array_view_x( + temp_array_view.data(), x.n_import_indices()); x.export_to_ghosted_array_start( 3, locally_owned_data_view, temp_array_view_x, ghosts_view, requests); x.export_to_ghosted_array_finish(ghosts_view, diff --git a/tests/cuda/parallel_partitioner_07.cu b/tests/cuda/parallel_partitioner_07.cu index 1372d1f4c6..b2df4ae436 100644 --- a/tests/cuda/parallel_partitioner_07.cu +++ b/tests/cuda/parallel_partitioner_07.cu @@ -28,7 +28,7 @@ template void -print_cuda_view(const ArrayView cuda_view) +print_cuda_view(const ArrayView cuda_view) { std::vector cpu_values(cuda_view.size()); Utilities::CUDA::copy_to_host(cuda_view.data(), cpu_values); @@ -104,22 +104,22 @@ test() std::unique_ptr ghost_array( Utilities::CUDA::allocate_device_data(cpu_ghost_array.size()), Utilities::CUDA::delete_device_data); - ArrayView ghost_array_view(ghost_array.get(), - cpu_ghost_array.size()); + ArrayView ghost_array_view( + ghost_array.get(), cpu_ghost_array.size()); Utilities::CUDA::copy_to_dev(cpu_ghost_array, ghost_array.get()); // set up other arrays std::unique_ptr locally_owned_array( Utilities::CUDA::allocate_device_data(local_size), Utilities::CUDA::delete_device_data); - ArrayView locally_owned_array_view(locally_owned_array.get(), - local_size); + ArrayView locally_owned_array_view( + locally_owned_array.get(), local_size); std::unique_ptr temp_array( Utilities::CUDA::allocate_device_data(v.n_import_indices()), Utilities::CUDA::delete_device_data); - ArrayView temp_array_view(temp_array.get(), - v.n_import_indices()); + ArrayView temp_array_view( + temp_array.get(), v.n_import_indices()); std::vector requests; @@ -129,8 +129,9 @@ test() Utilities::CUDA::allocate_device_data( ghost_array_view.size()), Utilities::CUDA::delete_device_data); - ArrayView ghosts_view(ghosts.get(), ghost_array_view.size()); - const cudaError_t cuda_error = + ArrayView ghosts_view( + ghosts.get(), ghost_array_view.size()); + const cudaError_t cuda_error = cudaMemcpy(ghosts.get(), ghost_array_view.data(), ghost_array_view.size() * sizeof(unsigned int), @@ -160,15 +161,16 @@ test() locally_owned_array_view.size() * sizeof(unsigned int)); AssertCuda(cuda_error); Assert(temp_array_view.size() >= w.n_import_indices(), ExcInternalError()); - ArrayView temp_array_view_w(temp_array_view.data(), - w.n_import_indices()); + ArrayView temp_array_view_w( + temp_array_view.data(), w.n_import_indices()); { std::unique_ptr ghosts( Utilities::CUDA::allocate_device_data( ghost_array_view.size()), Utilities::CUDA::delete_device_data); - ArrayView ghosts_view(ghosts.get(), ghost_array_view.size()); - const cudaError_t cuda_error = + ArrayView ghosts_view( + ghosts.get(), ghost_array_view.size()); + const cudaError_t cuda_error = cudaMemcpy(ghosts.get(), ghost_array_view.data(), ghost_array_view.size() * sizeof(unsigned int), @@ -199,15 +201,16 @@ test() locally_owned_array_view.size() * sizeof(unsigned int)); AssertCuda(cuda_error); Assert(temp_array_view.size() >= x.n_import_indices(), ExcInternalError()); - ArrayView temp_array_view_x(temp_array_view.data(), - x.n_import_indices()); + ArrayView temp_array_view_x( + temp_array_view.data(), x.n_import_indices()); { std::unique_ptr ghosts( Utilities::CUDA::allocate_device_data( ghost_array_view.size()), Utilities::CUDA::delete_device_data); - ArrayView ghosts_view(ghosts.get(), ghost_array_view.size()); - const cudaError_t cuda_error = + ArrayView ghosts_view( + ghosts.get(), ghost_array_view.size()); + const cudaError_t cuda_error = cudaMemcpy(ghosts.get(), ghost_array_view.data(), ghost_array_view.size() * sizeof(unsigned int), @@ -236,7 +239,8 @@ test() std::unique_ptr ghosts( Utilities::CUDA::allocate_device_data(cpu_ghosts.size()), Utilities::CUDA::delete_device_data); - ArrayView ghosts_view(ghosts.get(), cpu_ghosts.size()); + ArrayView ghosts_view(ghosts.get(), + cpu_ghosts.size()); Utilities::CUDA::copy_to_dev(cpu_ghosts, ghosts.get()); x.import_from_ghosted_array_start( diff --git a/tests/cuda/parallel_partitioner_08.cu b/tests/cuda/parallel_partitioner_08.cu index 65a0fbd4f2..a6a84564ba 100644 --- a/tests/cuda/parallel_partitioner_08.cu +++ b/tests/cuda/parallel_partitioner_08.cu @@ -25,7 +25,7 @@ template void -print_cuda_view(const ArrayView cuda_view) +print_cuda_view(const ArrayView cuda_view) { std::vector cpu_values(cuda_view.size()); Utilities::CUDA::copy_to_host(cuda_view.data(), cpu_values); @@ -95,19 +95,21 @@ test() // create vector std::vector cpu_owned(rank == 0 ? 8 : 0); - for (int i = 0; i < cpu_owned.size(); ++i) + for (unsigned int i = 0; i < cpu_owned.size(); ++i) cpu_owned[i] = i; std::unique_ptr owned( Utilities::CUDA::allocate_device_data(cpu_owned.size()), Utilities::CUDA::delete_device_data); - ArrayView owned_view(owned.get(), cpu_owned.size()); + ArrayView owned_view(owned.get(), + cpu_owned.size()); Utilities::CUDA::copy_to_dev(cpu_owned, owned.get()); std::vector cpu_ghost(4, 0); std::unique_ptr ghost( Utilities::CUDA::allocate_device_data(cpu_ghost.size()), Utilities::CUDA::delete_device_data); - ArrayView ghost_view(ghost.get(), cpu_ghost.size()); + ArrayView ghost_view(ghost.get(), + cpu_ghost.size()); Utilities::CUDA::copy_to_dev(cpu_ghost, ghost.get()); // update ghost values @@ -120,8 +122,8 @@ test() Utilities::CUDA::allocate_device_data( tight_partitioner->n_import_indices()), Utilities::CUDA::delete_device_data); - ArrayView tmp_data_view(tmp_data.get(), - tight_partitioner->n_import_indices()); + ArrayView tmp_data_view( + tmp_data.get(), tight_partitioner->n_import_indices()); // begin exchange, and ... tight_partitioner->export_to_ghosted_array_start( @@ -145,8 +147,8 @@ test() Utilities::CUDA::allocate_device_data( tight_partitioner->n_import_indices()), Utilities::CUDA::delete_device_data); - ArrayView import_data_view(tmp_data.get(), - tight_partitioner->n_import_indices()); + ArrayView import_data_view( + tmp_data.get(), tight_partitioner->n_import_indices()); // now do insert: auto compress = [&](VectorOperation::values operation) { -- 2.39.5