Skip to content

Add CUDA matrix-free smoother#217

Open
Rombur wants to merge 1 commit intoORNL-CEES:masterfrom
Rombur:cuda_smoother
Open

Add CUDA matrix-free smoother#217
Rombur wants to merge 1 commit intoORNL-CEES:masterfrom
Rombur:cuda_smoother

Conversation

@Rombur
Copy link
Copy Markdown
Collaborator

@Rombur Rombur commented Jul 29, 2019

No description provided.

@codecov-io
Copy link
Copy Markdown

Codecov Report

Merging #217 into master will not change coverage.
The diff coverage is n/a.

Impacted file tree graph

@@           Coverage Diff           @@
##           master     #217   +/-   ##
=======================================
  Coverage   89.78%   89.78%           
=======================================
  Files          57       57           
  Lines        3299     3299           
=======================================
  Hits         2962     2962           
  Misses        337      337

Continue to review full report at Codecov.

Legend - Click here to learn more
Δ = absolute <relative> (impact), ø = not affected, ? = missing data
Powered by Codecov. Last update b7c66df...9af44b1. Read the comment docs.

std::transform(prec_type.begin(), prec_type.end(), prec_type.begin(),
tolower);

ASSERT_THROW(prec_type == "jacobi", "Only Jacobi smoother is implemented.");
Copy link
Copy Markdown
Collaborator

Choose a reason for hiding this comment

The reason will be displayed to describe this comment to others. Learn more.

We should soon be able to use PreconditionChebyshev as well.

}

template <typename ScalarType>
__global__ void extract_inv_diag(ScalarType const *const matrix_value,
Copy link
Copy Markdown
Collaborator

Choose a reason for hiding this comment

The reason will be displayed to describe this comment to others. Learn more.

Looks like this is never used, is it?

SparseMatrixDevice<typename VectorType::value_type> const &smoother,
VectorType const &b, VectorType &x)
{
ASSERT_THROW_NOT_IMPLEMENTED();
Copy link
Copy Markdown
Collaborator

Choose a reason for hiding this comment

The reason will be displayed to describe this comment to others. Learn more.

Do we actually need to instantiate the operator for any type different from dealii::LinearAlgebra::distributed::Vector< double, dealii::MemorySpace::CUDA>>? If not, we could just use a static_assert instead.

dealii::LinearAlgebra::distributed::Vector<double, dealii::MemorySpace::CUDA>
tmp(x);
smoother.vmult(tmp, r);
x.add(-1., tmp);
Copy link
Copy Markdown
Collaborator

Choose a reason for hiding this comment

The reason will be displayed to describe this comment to others. Learn more.

There is nothing special about dealii::LinearAlgebra::distributed::Vector<double, dealii::MemorySpace::CUDA> here, is it? If we can indeed use static_assert, this would look much nicer with VectorType instead.

namespace
{
template <typename VectorType>
struct SmootherOperator
Copy link
Copy Markdown
Collaborator

Choose a reason for hiding this comment

The reason will be displayed to describe this comment to others. Learn more.

So all this is just for being able to specialize for VectorType. Do we actually need that if only allow for one type?

Sign up for free to join this conversation on GitHub. Already have an account? Sign in to comment

Labels

None yet

Projects

None yet

Development

Successfully merging this pull request may close these issues.

3 participants