QUDA  v1.1.0
A library for QCD on GPUs
solver.cpp
Go to the documentation of this file.
1 #include <quda_internal.h>
2 #include <invert_quda.h>
3 #include <multigrid.h>
4 #include <eigensolve_quda.h>
5 #include <cmath>
6 #include <limits>
7 
8 namespace quda {
9 
10  static void report(const char *type) {
11  if (getVerbosity() >= QUDA_VERBOSE) printfQuda("Creating a %s solver\n", type);
12  }
13 
14  Solver::Solver(const DiracMatrix &mat, const DiracMatrix &matSloppy, const DiracMatrix &matPrecon,
15  const DiracMatrix &matEig, SolverParam &param, TimeProfile &profile) :
16  mat(mat),
17  matSloppy(matSloppy),
18  matPrecon(matPrecon),
19  matEig(matEig),
20  param(param),
21  profile(profile),
22  node_parity(0),
23  eig_solve(nullptr),
24  deflate_init(false),
25  deflate_compute(true),
26  recompute_evals(!param.eig_param.preserve_evals)
27  {
28  // compute parity of the node
29  for (int i=0; i<4; i++) node_parity += commCoords(i);
31  }
32 
34  {
35  if (eig_solve) {
36  delete eig_solve;
37  eig_solve = nullptr;
38  }
39  }
40 
41  // solver factory
43  const DiracMatrix &matPrecon, const DiracMatrix &matEig, TimeProfile &profile)
44  {
45  Solver *solver = nullptr;
46 
48  errorQuda("Explicit preconditoner not supported for %d solver", param.inv_type);
49 
51  errorQuda("Explicit preconditoner not supported for %d preconditioner", param.inv_type_precondition);
52 
53  switch (param.inv_type) {
54  case QUDA_CG_INVERTER:
55  report("CG");
56  solver = new CG(mat, matSloppy, matPrecon, matEig, param, profile);
57  break;
59  report("BiCGstab");
60  solver = new BiCGstab(mat, matSloppy, matPrecon, matEig, param, profile);
61  break;
62  case QUDA_GCR_INVERTER:
63  report("GCR");
64  if (param.preconditioner) {
65  Solver *mg = param.mg_instance ? static_cast<MG*>(param.preconditioner) : static_cast<multigrid_solver*>(param.preconditioner)->mg;
66  // FIXME dirty hack to ensure that preconditioner precision set in interface isn't used in the outer GCR-MG solver
68  solver = new GCR(mat, *(mg), matSloppy, matPrecon, matEig, param, profile);
69  } else {
70  solver = new GCR(mat, matSloppy, matPrecon, matEig, param, profile);
71  }
72  break;
74  report("CA-CG");
75  solver = new CACG(mat, matSloppy, matPrecon, matEig, param, profile);
76  break;
78  report("CA-CGNE");
79  solver = new CACGNE(mat, matSloppy, matPrecon, matEig, param, profile);
80  break;
82  report("CA-CGNR");
83  solver = new CACGNR(mat, matSloppy, matPrecon, matEig, param, profile);
84  break;
86  report("CA-GCR");
87  solver = new CAGCR(mat, matSloppy, matPrecon, matEig, param, profile);
88  break;
89  case QUDA_MR_INVERTER:
90  report("MR");
91  solver = new MR(mat, matSloppy, param, profile);
92  break;
93  case QUDA_SD_INVERTER:
94  report("SD");
95  solver = new SD(mat, param, profile);
96  break;
97  case QUDA_XSD_INVERTER:
98 #ifdef MULTI_GPU
99  report("XSD");
100  solver = new XSD(mat, param, profile);
101 #else
102  errorQuda("Extended Steepest Descent is multi-gpu only");
103 #endif
104  break;
105  case QUDA_PCG_INVERTER:
106  report("PCG");
107  solver = new PreconCG(mat, matSloppy, matPrecon, matEig, param, profile);
108  break;
109  case QUDA_MPCG_INVERTER:
110  report("MPCG");
111  solver = new MPCG(mat, param, profile);
112  break;
114  report("MPBICGSTAB");
115  solver = new MPBiCGstab(mat, param, profile);
116  break;
118  report("BICGSTABL");
119  solver = new BiCGstabL(mat, matSloppy, param, profile);
120  break;
121  case QUDA_EIGCG_INVERTER:
122  report("EIGCG");
123  solver = new IncEigCG(mat, matSloppy, matPrecon, param, profile);
124  break;
126  report("INC EIGCG");
127  solver = new IncEigCG(mat, matSloppy, matPrecon, param, profile);
128  break;
130  report("GMRESDR");
131  if (param.preconditioner) {
133  // FIXME dirty hack to ensure that preconditioner precision set in interface isn't used in the outer GCR-MG solver
135  solver = new GMResDR(mat, *(mg->mg), matSloppy, matPrecon, param, profile);
136  } else {
137  solver = new GMResDR(mat, matSloppy, matPrecon, param, profile);
138  }
139  break;
140  case QUDA_CGNE_INVERTER:
141  report("CGNE");
142  solver = new CGNE(mat, matSloppy, matPrecon, matEig, param, profile);
143  break;
144  case QUDA_CGNR_INVERTER:
145  report("CGNR");
146  solver = new CGNR(mat, matSloppy, matPrecon, matEig, param, profile);
147  break;
148  case QUDA_CG3_INVERTER:
149  report("CG3");
150  solver = new CG3(mat, matSloppy, matPrecon, param, profile);
151  break;
152  case QUDA_CG3NE_INVERTER:
153  report("CG3NE");
154  solver = new CG3NE(mat, matSloppy, matPrecon, param, profile);
155  break;
156  case QUDA_CG3NR_INVERTER:
157  report("CG3NR");
158  solver = new CG3NR(mat, matSloppy, matPrecon, param, profile);
159  break;
160  default:
161  errorQuda("Invalid solver type %d", param.inv_type);
162  }
163 
164  if (!mat.hermitian() && solver->hermitian()) errorQuda("Cannot solve non-Hermitian system with Hermitian solver");
165  return solver;
166  }
167 
169  {
170  if (deflate_init) return;
171 
172  // Deflation requested + first instance of solver
173  bool profile_running = profile.isRunning(QUDA_PROFILE_INIT);
174  if (!param.is_preconditioner && !profile_running) profile.TPSTART(QUDA_PROFILE_INIT);
175 
177 
178  // Clone from an existing vector
181  // This is the vector precision used by matEig
183 
184  if (deflate_compute) {
185  evecs.reserve(param.eig_param.n_conv);
186  evals.reserve(param.eig_param.n_conv);
187 
189 
190  if (space && space->evecs.size() != 0) {
191  if (getVerbosity() >= QUDA_VERBOSE) printfQuda("Restoring deflation space of size %lu\n", space->evecs.size());
192 
193  if ((!space->svd && param.eig_param.n_conv != (int)space->evecs.size())
194  || (space->svd && 2 * param.eig_param.n_conv != (int)space->evecs.size()))
195  errorQuda("Preserved deflation space size %lu does not match expected %d", space->evecs.size(),
197 
198  // move vectors from preserved space to local space
199  for (auto &vec : space->evecs) evecs.push_back(vec);
200 
201  if (param.eig_param.n_conv != (int)space->evals.size())
202  errorQuda("Preserved eigenvalues %lu does not match expected %lu", space->evals.size(), evals.size());
203 
204  // move vectors from preserved space to local space
205  for (auto &val : space->evals) evals.push_back(val);
206 
207  space->evecs.resize(0);
208  space->evals.resize(0);
209 
210  delete space;
212 
213  // we successfully got the deflation space so disable any subsequent recalculation
214  deflate_compute = false;
215  } else {
216  // Computing the deflation space, rather than transferring, so we create space.
217  for (int i = 0; i < param.eig_param.n_conv; i++) evecs.push_back(ColorSpinorField::Create(csParam));
218 
219  evals.resize(param.eig_param.n_conv);
220  for (int i = 0; i < param.eig_param.n_conv; i++) evals[i] = 0.0;
221  }
222  }
223 
224  deflate_init = true;
225 
226  if (!param.is_preconditioner && !profile_running) profile.TPSTOP(QUDA_PROFILE_INIT);
227  }
228 
230  {
231  if (deflate_init) {
233  if (getVerbosity() >= QUDA_VERBOSE) printfQuda("Preserving deflation space of size %lu\n", evecs.size());
234 
237  // first ensure that any existing space is freed
238  for (auto &vec : space->evecs)
239  if (vec) delete vec;
240  space->evecs.resize(0);
241  delete space;
242  }
243 
244  deflation_space *space = new deflation_space;
245 
246  // if evecs size = 2x evals size then we are doing an SVD deflation
247  space->svd = (evecs.size() == 2 * evals.size()) ? true : false;
248 
249  space->evecs.reserve(evecs.size());
250  for (auto &vec : evecs) space->evecs.push_back(vec);
251 
252  space->evals.reserve(evals.size());
253  for (auto &val : evals) space->evals.push_back(val);
254 
256  } else {
257  for (auto &vec : evecs)
258  if (vec) delete vec;
259  }
260 
261  evecs.resize(0);
262  deflate_init = false;
263  }
264  }
265 
266  void Solver::injectDeflationSpace(std::vector<ColorSpinorField *> &defl_space)
267  {
268  if (!evecs.empty()) errorQuda("Solver deflation space should be empty, instead size=%lu\n", defl_space.size());
269  // Create space for the eigenvalues
270  evals.resize(defl_space.size());
271  // Create space for the eigenvectors, destroy defl_space
272  for (auto &e : defl_space) { evecs.push_back(e); }
273  defl_space.resize(0);
274  }
275 
276  void Solver::extractDeflationSpace(std::vector<ColorSpinorField *> &defl_space)
277  {
278  if (!defl_space.empty())
279  errorQuda("Container deflation space should be empty, instead size=%lu\n", defl_space.size());
280  // We do not care about the eigenvalues, they will be recomputed.
281  evals.resize(0);
282  // Create space for the eigenvectors, destroy evecs
283  for (auto &e : evecs) { defl_space.push_back(e); }
284  evecs.resize(0);
285  }
286 
288  {
289  if (!deflate_init) errorQuda("Deflation space for this solver not computed");
290 
291  // Double the size deflation space to accomodate for the extra singular vectors
292  // Clone from an existing vector
295  // This is the vector precision used by matResidual
297  for (int i = param.eig_param.n_conv; i < 2 * param.eig_param.n_conv; i++) {
299  }
300  }
301 
303  {
304  for (int i = 0; i < param.num_src; i++) {
305  (*this)(out.Component(i), in.Component(i));
308  }
309  }
310 
311  double Solver::stopping(double tol, double b2, QudaResidualType residual_type)
312  {
313  double stop=0.0;
314  if ( (residual_type & QUDA_L2_ABSOLUTE_RESIDUAL) &&
315  (residual_type & QUDA_L2_RELATIVE_RESIDUAL) ) {
316  // use the most stringent stopping condition
317  double lowest = (b2 < 1.0) ? b2 : 1.0;
318  stop = lowest*tol*tol;
319  } else if (residual_type & QUDA_L2_ABSOLUTE_RESIDUAL) {
320  stop = tol*tol;
321  } else {
322  stop = b2*tol*tol;
323  }
324 
325  return stop;
326  }
327 
328  bool Solver::convergence(double r2, double hq2, double r2_tol, double hq_tol) {
329 
330  // check the heavy quark residual norm if necessary
332  if (std::isnan(hq2) || std::isinf(hq2))
333  errorQuda("Solver appears to have diverged with heavy quark residual %9.6e", hq2);
334 
335  if (hq2 > hq_tol) return false;
336  }
337 
338  // check the L2 relative residual norm if necessary
340  if (std::isnan(r2) || std::isinf(r2)) errorQuda("Solver appears to have diverged with residual %9.6e", r2);
341 
342  if (r2 > r2_tol) return false;
343  }
344 
345  return true;
346  }
347 
348  bool Solver::convergenceHQ(double r2, double hq2, double r2_tol, double hq_tol) {
349 
350  // check the heavy quark residual norm if necessary
352  if (std::isnan(hq2) || std::isinf(hq2))
353  errorQuda("Solver appears to have diverged with heavy quark residual %9.6e", hq2);
354 
355  if (hq2 > hq_tol) return false;
356  }
357 
358  return true;
359  }
360 
361  bool Solver::convergenceL2(double r2, double hq2, double r2_tol, double hq_tol) {
362 
363  // check the L2 relative residual norm if necessary
365  if (std::isnan(r2) || std::isinf(r2)) errorQuda("Solver appears to have diverged with residual %9.6e", r2);
366 
367  if (r2 > r2_tol) return false;
368  }
369 
370  return true;
371  }
372 
373  void Solver::PrintStats(const char* name, int k, double r2, double b2, double hq2) {
374  if (getVerbosity() >= QUDA_VERBOSE) {
376  printfQuda("%s: %5d iterations, <r,r> = %9.6e, |r|/|b| = %9.6e, heavy-quark residual = %9.6e\n", name, k, r2,
377  sqrt(r2 / b2), hq2);
378  } else {
379  printfQuda("%s: %5d iterations, <r,r> = %9.6e, |r|/|b| = %9.6e\n", name, k, r2, sqrt(r2 / b2));
380  }
381  }
382 
383  if (std::isnan(r2) || std::isinf(r2)) errorQuda("Solver appears to have diverged");
384  }
385 
386  void Solver::PrintSummary(const char *name, int k, double r2, double b2,
387  double r2_tol, double hq_tol) {
388  if (getVerbosity() >= QUDA_SUMMARIZE) {
389  if (param.compute_true_res) {
391  printfQuda("%s: Convergence at %d iterations, L2 relative residual: iterated = %9.6e, true = %9.6e "
392  "(requested = %9.6e), heavy-quark residual = %9.6e (requested = %9.6e)\n",
393  name, k, sqrt(r2 / b2), param.true_res, sqrt(r2_tol / b2), param.true_res_hq, hq_tol);
394  } else {
395  printfQuda("%s: Convergence at %d iterations, L2 relative residual: iterated = %9.6e, true = %9.6e "
396  "(requested = %9.6e)\n",
397  name, k, sqrt(r2 / b2), param.true_res, sqrt(r2_tol / b2));
398  }
399  } else {
401  printfQuda("%s: Convergence at %d iterations, L2 relative residual: iterated = %9.6e "
402  "(requested = %9.6e), heavy-quark residual = %9.6e (requested = %9.6e)\n",
403  name, k, sqrt(r2 / b2), sqrt(r2_tol / b2), param.true_res_hq, hq_tol);
404  } else {
405  printfQuda("%s: Convergence at %d iterations, L2 relative residual: iterated = %9.6e (requested = %9.6e)\n",
406  name, k, sqrt(r2 / b2), sqrt(r2_tol / b2));
407  }
408  }
409  }
410  }
411 
413  {
414  double eps = 0.;
416 
417  switch (prec) {
420  case QUDA_HALF_PRECISION: eps = pow(2., -13); break;
421  case QUDA_QUARTER_PRECISION: eps = pow(2., -6); break;
422  default: errorQuda("Invalid precision %d", param.precision); break;
423  }
424  return eps;
425  }
426 
427  bool MultiShiftSolver::convergence(const double *r2, const double *r2_tol, int n) const {
428 
429  // check the L2 relative residual norm if necessary
431  for (int i = 0; i < n; i++) {
432  if (std::isnan(r2[i]) || std::isinf(r2[i]))
433  errorQuda("Multishift solver appears to have diverged on shift %d with residual %9.6e", i, r2[i]);
434 
435  if (r2[i] > r2_tol[i] && r2_tol[i] != 0.0) return false;
436  }
437  }
438 
439  return true;
440  }
441 
442 } // namespace quda
Communication-avoiding CG solver. This solver does un-preconditioned CG, running in steps of n_krylov...
Definition: invert_quda.h:989
Communication-avoiding GCR solver. This solver does un-preconditioned GCR, first building up a polyno...
Definition: invert_quda.h:1099
Conjugate-Gradient Solver.
Definition: invert_quda.h:639
static ColorSpinorField * Create(const ColorSpinorParam &param)
ColorSpinorField & Component(const int idx) const
virtual bool hermitian() const
Definition: dirac_quda.h:1962
static EigenSolver * create(QudaEigParam *eig_param, const DiracMatrix &mat, TimeProfile &profile)
Creates the eigensolver using the parameters given and the matrix.
bool convergence(const double *r2, const double *r2_tol, int n) const
Definition: solver.cpp:427
bool deflate_init
Definition: invert_quda.h:474
double precisionEpsilon(QudaPrecision prec=QUDA_INVALID_PRECISION) const
Returns the epsilon tolerance for a given precision, by default returns the solver precision.
Definition: solver.cpp:412
void extractDeflationSpace(std::vector< ColorSpinorField * > &defl_space)
Extracts the deflation space from the solver to the vector argument. Note the solver deflation space ...
Definition: solver.cpp:276
bool deflate_compute
Definition: invert_quda.h:475
TimeProfile & profile
Definition: invert_quda.h:471
Solver(const DiracMatrix &mat, const DiracMatrix &matSloppy, const DiracMatrix &matPrecon, const DiracMatrix &matEig, SolverParam &param, TimeProfile &profile)
Definition: solver.cpp:14
bool convergenceL2(double r2, double hq2, double r2_tol, double hq_tol)
Test for L2 solver convergence – ignore HQ residual.
Definition: solver.cpp:361
virtual void blocksolve(ColorSpinorField &out, ColorSpinorField &in)
Definition: solver.cpp:302
const DiracMatrix & mat
Definition: invert_quda.h:465
virtual ~Solver()
Definition: solver.cpp:33
bool convergence(double r2, double hq2, double r2_tol, double hq_tol)
Definition: solver.cpp:328
std::vector< ColorSpinorField * > evecs
Definition: invert_quda.h:477
bool convergenceHQ(double r2, double hq2, double r2_tol, double hq_tol)
Test for HQ solver convergence – ignore L2 residual.
Definition: solver.cpp:348
void destroyDeflationSpace()
Destroy the allocated deflation space.
Definition: solver.cpp:229
void PrintSummary(const char *name, int k, double r2, double b2, double r2_tol, double hq_tol)
Prints out the summary of the solver convergence (requires a verbosity of QUDA_SUMMARIZE)....
Definition: solver.cpp:386
const DiracMatrix & matEig
Definition: invert_quda.h:468
void injectDeflationSpace(std::vector< ColorSpinorField * > &defl_space)
Injects a deflation space into the solver from the vector argument. Note the input space is reduced t...
Definition: solver.cpp:266
SolverParam & param
Definition: invert_quda.h:470
static double stopping(double tol, double b2, QudaResidualType residual_type)
Set the solver L2 stopping condition.
Definition: solver.cpp:311
void extendSVDDeflationSpace()
Extends the deflation space to twice its size for SVD deflation.
Definition: solver.cpp:287
std::vector< Complex > evals
Definition: invert_quda.h:478
virtual bool hermitian()=0
EigenSolver * eig_solve
Definition: invert_quda.h:473
void PrintStats(const char *name, int k, double r2, double b2, double hq2)
Prints out the running statistics of the solver (requires a verbosity of QUDA_VERBOSE)
Definition: solver.cpp:373
void constructDeflationSpace(const ColorSpinorField &meta, const DiracMatrix &mat)
Constructs the deflation space and eigensolver.
Definition: solver.cpp:168
const DiracMatrix & matPrecon
Definition: invert_quda.h:467
static Solver * create(SolverParam &param, const DiracMatrix &mat, const DiracMatrix &matSloppy, const DiracMatrix &matPrecon, const DiracMatrix &matEig, TimeProfile &profile)
Solver factory.
Definition: solver.cpp:42
const DiracMatrix & matSloppy
Definition: invert_quda.h:466
bool isRunning(QudaProfileType idx)
Definition: timer.h:260
int commCoords(int)
double tol
double epsilon
QudaPrecision prec
void mat(void *out, void **link, void *in, int dagger_bit, int mu, QudaPrecision sPrecision, QudaPrecision gPrecision)
enum QudaPrecision_s QudaPrecision
@ QUDA_SUMMARIZE
Definition: enum_quda.h:266
@ QUDA_VERBOSE
Definition: enum_quda.h:267
@ QUDA_HEAVY_QUARK_RESIDUAL
Definition: enum_quda.h:195
@ QUDA_L2_ABSOLUTE_RESIDUAL
Definition: enum_quda.h:194
@ QUDA_L2_RELATIVE_RESIDUAL
Definition: enum_quda.h:193
enum QudaResidualType_s QudaResidualType
@ QUDA_CA_CGNE_INVERTER
Definition: enum_quda.h:130
@ QUDA_GCR_INVERTER
Definition: enum_quda.h:109
@ QUDA_CGNE_INVERTER
Definition: enum_quda.h:124
@ QUDA_INC_EIGCG_INVERTER
Definition: enum_quda.h:117
@ QUDA_MR_INVERTER
Definition: enum_quda.h:110
@ QUDA_CA_CG_INVERTER
Definition: enum_quda.h:129
@ QUDA_MPCG_INVERTER
Definition: enum_quda.h:115
@ QUDA_CG3NE_INVERTER
Definition: enum_quda.h:127
@ QUDA_BICGSTABL_INVERTER
Definition: enum_quda.h:123
@ QUDA_CGNR_INVERTER
Definition: enum_quda.h:125
@ QUDA_CG3NR_INVERTER
Definition: enum_quda.h:128
@ QUDA_PCG_INVERTER
Definition: enum_quda.h:114
@ QUDA_CA_GCR_INVERTER
Definition: enum_quda.h:132
@ QUDA_SD_INVERTER
Definition: enum_quda.h:112
@ QUDA_MPBICGSTAB_INVERTER
Definition: enum_quda.h:111
@ QUDA_CG_INVERTER
Definition: enum_quda.h:107
@ QUDA_CA_CGNR_INVERTER
Definition: enum_quda.h:131
@ QUDA_CG3_INVERTER
Definition: enum_quda.h:126
@ QUDA_EIGCG_INVERTER
Definition: enum_quda.h:116
@ QUDA_XSD_INVERTER
Definition: enum_quda.h:113
@ QUDA_MG_INVERTER
Definition: enum_quda.h:122
@ QUDA_BICGSTAB_INVERTER
Definition: enum_quda.h:108
@ QUDA_GMRESDR_INVERTER
Definition: enum_quda.h:118
@ QUDA_DOUBLE_PRECISION
Definition: enum_quda.h:65
@ QUDA_SINGLE_PRECISION
Definition: enum_quda.h:64
@ QUDA_INVALID_PRECISION
Definition: enum_quda.h:66
@ QUDA_QUARTER_PRECISION
Definition: enum_quda.h:62
@ QUDA_HALF_PRECISION
Definition: enum_quda.h:63
@ QUDA_ZERO_FIELD_CREATE
Definition: enum_quda.h:361
void stop()
Stop profiling.
Definition: device.cpp:228
__host__ __device__ ValueType sqrt(ValueType x)
Definition: complex_quda.h:120
@ QUDA_PROFILE_INIT
Definition: timer.h:106
__host__ __device__ ValueType pow(ValueType x, ExponentType e)
Definition: complex_quda.h:111
ColorSpinorParam csParam
Definition: pack_test.cpp:25
QudaGaugeParam param
Definition: pack_test.cpp:18
void * preserve_deflation_space
Definition: quda.h:435
QudaBoolean preserve_deflation
Definition: quda.h:431
int n_conv
Definition: quda.h:475
QudaPrecision precision
Definition: invert_quda.h:136
bool is_preconditioner
verbosity to use for preconditioner
Definition: invert_quda.h:238
QudaResidualType residual_type
Definition: invert_quda.h:49
QudaPrecision precision_precondition
Definition: invert_quda.h:145
QudaPrecision precision_sloppy
Definition: invert_quda.h:139
bool mg_instance
whether to use a global or local (node) reduction for this solver
Definition: invert_quda.h:245
double true_res_offset[QUDA_MAX_MULTI_SHIFT]
Definition: invert_quda.h:178
void * preconditioner
Definition: invert_quda.h:33
QudaPrecision precision_eigensolver
Definition: invert_quda.h:148
QudaInverterType inv_type
Definition: invert_quda.h:22
QudaEigParam eig_param
Definition: invert_quda.h:55
double true_res_hq_offset[QUDA_MAX_MULTI_SHIFT]
Definition: invert_quda.h:184
QudaInverterType inv_type_precondition
Definition: invert_quda.h:28
This is an object that captures the state required for a deflated solver.
Definition: invert_quda.h:1457
std::vector< ColorSpinorField * > evecs
Definition: invert_quda.h:1459
std::vector< Complex > evals
Definition: invert_quda.h:1460
#define printfQuda(...)
Definition: util_quda.h:114
QudaVerbosity getVerbosity()
Definition: util_quda.cpp:21
#define errorQuda(...)
Definition: util_quda.h:120