From ab2d9a8c76492a8225135c863b13406513dcd46b Mon Sep 17 00:00:00 2001 From: "T.D. (Tongdong) Qiu" Date: Fri, 20 Mar 2020 10:04:58 +0100 Subject: [PATCH] Changed int idx to indicate which matrix is added to WellContributions to an enum --- opm/simulators/linalg/bda/WellContributions.cu | 15 +++++++-------- opm/simulators/linalg/bda/WellContributions.hpp | 12 ++++++++++-- opm/simulators/wells/StandardWell_impl.hpp | 6 +++--- 3 files changed, 20 insertions(+), 13 deletions(-) diff --git a/opm/simulators/linalg/bda/WellContributions.cu b/opm/simulators/linalg/bda/WellContributions.cu index 79e5e5e4e..ac79bb21c 100644 --- a/opm/simulators/linalg/bda/WellContributions.cu +++ b/opm/simulators/linalg/bda/WellContributions.cu @@ -161,20 +161,21 @@ namespace Opm } - void WellContributions::addMatrix(int idx, int *colIndices, double *values, unsigned int val_size) + void WellContributions::addMatrix(MatrixType type, int *colIndices, double *values, unsigned int val_size) { if (!allocated) { OPM_THROW(std::logic_error,"Error cannot add wellcontribution before allocating memory in WellContributions"); } - switch (idx) { - case 0: + + switch (type) { + case MatrixType::C: cudaMemcpy(d_Cnnzs + num_blocks_so_far * dim * dim_wells, values, sizeof(double) * val_size * dim * dim_wells, cudaMemcpyHostToDevice); cudaMemcpy(d_Ccols + num_blocks_so_far, colIndices, sizeof(int) * val_size, cudaMemcpyHostToDevice); break; - case 1: + case MatrixType::D: cudaMemcpy(d_Dnnzs + num_wells_so_far * dim_wells * dim_wells, values, sizeof(double) * dim_wells * dim_wells, cudaMemcpyHostToDevice); break; - case 2: + case MatrixType::B: cudaMemcpy(d_Bnnzs + num_blocks_so_far * dim * dim_wells, values, sizeof(double) * val_size * dim * dim_wells, cudaMemcpyHostToDevice); cudaMemcpy(d_Bcols + num_blocks_so_far, colIndices, sizeof(int) * val_size, cudaMemcpyHostToDevice); val_pointers[num_wells_so_far] = num_blocks_so_far; @@ -187,10 +188,8 @@ namespace Opm OPM_THROW(std::logic_error,"Error unsupported matrix ID for WellContributions::addMatrix()"); } cudaCheckLastError("WellContributions::addMatrix() failed"); - if(idx == 2){ + if (MatrixType::B == type) { num_blocks_so_far += val_size; - } - if(idx == 2){ num_wells_so_far++; } } diff --git a/opm/simulators/linalg/bda/WellContributions.hpp b/opm/simulators/linalg/bda/WellContributions.hpp index e495bdcc9..74f7efc09 100644 --- a/opm/simulators/linalg/bda/WellContributions.hpp +++ b/opm/simulators/linalg/bda/WellContributions.hpp @@ -68,6 +68,14 @@ namespace Opm unsigned int *d_val_pointers = nullptr; cudaStream_t stream; public: + + /// StandardWell has C, D and B matrices that need to be copied + enum class MatrixType { + C, + D, + B + }; + /// Set a cudaStream to be used /// \param[in] stream the cudaStream that is used to launch the kernel in void setCudaStream(cudaStream_t stream); @@ -97,11 +105,11 @@ namespace Opm void addNumBlocks(unsigned int numBlocks); /// Store a matrix in this object, in blocked csr format, can only be called after alloc() is called - /// \param[in] idx indicate if C, D or B is sent + /// \param[in] type indicate if C, D or B is sent /// \param[in] colIndices columnindices of blocks in C or B, ignored for D /// \param[in] values array of nonzeroes /// \param[in] val_size number of blocks in C or B, ignored for D - void addMatrix(int idx, int *colIndices, double *values, unsigned int val_size); + void addMatrix(MatrixType type, int *colIndices, double *values, unsigned int val_size); /// Return the number of wells added to this object /// \return the number of wells added to this object diff --git a/opm/simulators/wells/StandardWell_impl.hpp b/opm/simulators/wells/StandardWell_impl.hpp index 57da28f5a..366778267 100644 --- a/opm/simulators/wells/StandardWell_impl.hpp +++ b/opm/simulators/wells/StandardWell_impl.hpp @@ -2756,7 +2756,7 @@ namespace Opm } } } - wellContribs.addMatrix(0, colIndices.data(), nnzValues.data(), duneC_.nonzeroes()); + wellContribs.addMatrix(WellContributions::MatrixType::C, colIndices.data(), nnzValues.data(), duneC_.nonzeroes()); // invDuneD colIndices.clear(); @@ -2768,7 +2768,7 @@ namespace Opm nnzValues.emplace_back(invDuneD_[0][0][i][j]); } } - wellContribs.addMatrix(1, colIndices.data(), nnzValues.data(), 1); + wellContribs.addMatrix(WellContributions::MatrixType::D, colIndices.data(), nnzValues.data(), 1); // duneB colIndices.clear(); @@ -2782,7 +2782,7 @@ namespace Opm } } } - wellContribs.addMatrix(2, colIndices.data(), nnzValues.data(), duneB_.nonzeroes()); + wellContribs.addMatrix(WellContributions::MatrixType::B, colIndices.data(), nnzValues.data(), duneB_.nonzeroes()); }