QUDA  v0.7.0
A library for QCD on GPUs
All Classes Namespaces Files Functions Variables Typedefs Enumerations Enumerator Friends Macros Pages
color_spinor_field.h
Go to the documentation of this file.
1 #ifndef _COLOR_SPINOR_FIELD_H
2 #define _COLOR_SPINOR_FIELD_H
3 
4 #include <quda_internal.h>
5 #include <quda.h>
6 
7 #include <iostream>
8 
9 #include <lattice_field.h>
10 
11 namespace quda {
12  struct FullClover;
13 
15  public:
16  int nColor; // Number of colors of the field
17  int nSpin; // =1 for staggered, =2 for coarse Dslash, =4 for 4d spinor
18 
19  QudaTwistFlavorType twistFlavor; // used by twisted mass
20 
21  QudaSiteOrder siteOrder; // defined for full fields
22 
23  QudaFieldOrder fieldOrder; // Float, Float2, Float4 etc.
26 
27  QudaDWFPCType PCtype; // used to select preconditioning method in DWF
28 
29  void *v; // pointer to field
30  void *norm;
31 
33  int eigv_dim; //number of eigenvectors
34  int eigv_id; //eigenvector index
35 
37 
42  PCtype(QUDA_PC_INVALID), eigv_dim(0), eigv_id(-1) { ; }
43 
44  // used to create cpu params
45  ColorSpinorParam(void *V, QudaInvertParam &inv_param, const int *X, const bool pc_solution)
46  : LatticeFieldParam(4, X, 0, inv_param.cpu_prec), nColor(3),
47  nSpin( (inv_param.dslash_type == QUDA_ASQTAD_DSLASH ||
48  inv_param.dslash_type == QUDA_STAGGERED_DSLASH) ? 1 : 4),
50  fieldOrder(QUDA_INVALID_FIELD_ORDER), gammaBasis(inv_param.gamma_basis),
52  v(V), eigv_dim(0), eigv_id(-1) {
53 
54  if (nDim > QUDA_MAX_DIM) errorQuda("Number of dimensions too great");
55  for (int d=0; d<nDim; d++) x[d] = X[d];
56 
57  if (!pc_solution) {
59  } else {
60  x[0] /= 2; // X defined the full lattice dimensions
62  }
63 
64  if (inv_param.dslash_type == QUDA_DOMAIN_WALL_DSLASH ||
66  inv_param.dslash_type == QUDA_MOBIUS_DWF_DSLASH) {
67  nDim++;
68  x[4] = inv_param.Ls;
69  }
71  nDim++;
72  x[4] = 2;//for two flavors
73  }
74 
75  if (inv_param.dirac_order == QUDA_INTERNAL_DIRAC_ORDER) {
79  } else if (inv_param.dirac_order == QUDA_CPS_WILSON_DIRAC_ORDER) {
82  } else if (inv_param.dirac_order == QUDA_QDP_DIRAC_ORDER) {
85  } else if (inv_param.dirac_order == QUDA_DIRAC_ORDER) {
88  } else if (inv_param.dirac_order == QUDA_QDPJIT_DIRAC_ORDER) {
91  } else {
92  errorQuda("Dirac order %d not supported", inv_param.dirac_order);
93  }
94  }
95 
96  // used to create cuda param from a cpu param
98  : LatticeFieldParam(cpuParam.nDim, cpuParam.x, inv_param.sp_pad, inv_param.cuda_prec),
99  nColor(cpuParam.nColor), nSpin(cpuParam.nSpin), twistFlavor(cpuParam.twistFlavor),
102  create(QUDA_COPY_FIELD_CREATE), PCtype(cpuParam.PCtype), v(0), eigv_dim(cpuParam.eigv_dim), eigv_id(-1)
103  {
104  siteSubset = cpuParam.siteSubset;
107  }
108 
110  this->precision = precision;
111  fieldOrder = (precision == QUDA_DOUBLE_PRECISION || nSpin == 1) ?
113  }
114 
115  void print() {
116  printfQuda("nColor = %d\n", nColor);
117  printfQuda("nSpin = %d\n", nSpin);
118  printfQuda("twistFlavor = %d\n", twistFlavor);
119  printfQuda("nDim = %d\n", nDim);
120  for (int d=0; d<nDim; d++) printfQuda("x[%d] = %d\n", d, x[d]);
121  printfQuda("precision = %d\n", precision);
122  printfQuda("pad = %d\n", pad);
123  printfQuda("siteSubset = %d\n", siteSubset);
124  printfQuda("siteOrder = %d\n", siteOrder);
125  printfQuda("fieldOrder = %d\n", fieldOrder);
126  printfQuda("gammaBasis = %d\n", gammaBasis);
127  printfQuda("create = %d\n", create);
128  printfQuda("v = %lx\n", (unsigned long)v);
129  printfQuda("norm = %lx\n", (unsigned long)norm);
131  if(eigv_dim != 0) printfQuda("nEv = %d\n", eigv_dim);
132  }
133 
134  virtual ~ColorSpinorParam() {
135  }
136  };
137 
138  class cpuColorSpinorField;
139  class cudaColorSpinorField;
140 
142 
143  private:
144  void create(int nDim, const int *x, int Nc, int Ns, QudaTwistFlavorType Twistflavor,
147  int nev = 0, int evid = -1);
148  void destroy();
149 
150  protected:
151  bool init;
153 
154  int nColor;
155  int nSpin;
156 
157  int nDim;
158  int x[QUDA_MAX_DIM];
159 
160  int volume;
161  int volumeCB;
162  int pad;
163  int stride;
164 
166 
167  QudaDWFPCType PCtype; // used to select preconditioning method in DWF
168 
169  int real_length; // physical length only
170  int length; // length including pads, but not ghost zone - used for BLAS
171 
172  void *v; // the field elements
173  void *norm; // the normalization field
175  int eigv_dim;
176  int eigv_id;
177  int eigv_volume; // volume of a single eigenvector
178  int eigv_stride; // stride of a single eigenvector
179  int eigv_real_length; // physical length of a single eigenvector
180  int eigv_length; // length including pads (but not ghost zones)
181 
182  // multi-GPU parameters
183  void* ghost[QUDA_MAX_DIM]; // pointers to the ghost regions - NULL by default
184  void* ghostNorm[QUDA_MAX_DIM]; // pointers to ghost norms - NULL by default
185 
186  int ghostFace[QUDA_MAX_DIM];// the size of each face
187  int ghostOffset[QUDA_MAX_DIM]; // offsets to each ghost zone
188  int ghostNormOffset[QUDA_MAX_DIM]; // offsets to each ghost zone for norm field
189 
190  int ghost_length; // length of ghost zone
191  int ghost_norm_length; // length of ghost zone for norm
192  int total_length; // total length of spinor (physical + pad + ghost)
193  int total_norm_length; // total length of norm
194 
195  size_t bytes; // size in bytes of spinor field
196  size_t norm_bytes; // size in bytes of norm field
197 
198  /*Warning: we need copies of the above params for eigenvectors*/
199  //multi_GPU parameters:
200 
201  //ghost pointers are always for single eigenvector..
206 
207  size_t eigv_bytes; // size in bytes of spinor field
208  size_t eigv_norm_bytes; // makes no sense but let's keep it...
209 
214 
215  // in the case of full fields, these are references to the even / odd sublattices
218 
220  std::vector<ColorSpinorField*> eigenvectors;
221 
222  void createGhostZone();
223 
224  // resets the above attributes based on contents of param
225  void reset(const ColorSpinorParam &);
226  void fill(ColorSpinorParam &) const;
227  static void checkField(const ColorSpinorField &, const ColorSpinorField &);
228  void clearGhostPointers();
229 
230  char aux_string[TuneKey::aux_n]; // used as a label in the autotuner
231  void setTuningString(); // set the vol_string and aux_string for use in tuning
232 
233  public:
234  //ColorSpinorField();
237 
238  virtual ~ColorSpinorField();
239 
240  virtual ColorSpinorField& operator=(const ColorSpinorField &);
241 
242  QudaPrecision Precision() const { return precision; }
243  int Ncolor() const { return nColor; }
244  int Nspin() const { return nSpin; }
246  int Ndim() const { return nDim; }
247  const int* X() const { return x; }
248  int X(int d) const { return x[d]; }
249  int RealLength() const { return real_length; }
250  int Length() const { return length; }
251  int TotalLength() const { return total_length; }
252  int Stride() const { return stride; }
253  int Volume() const { return volume; }
254  int VolumeCB() const { return siteSubset == QUDA_PARITY_SITE_SUBSET ? volume : volume / 2; }
255  int Pad() const { return pad; }
256  size_t Bytes() const { return bytes; }
257  size_t NormBytes() const { return norm_bytes; }
258  void PrintDims() const { printfQuda("dimensions=%d %d %d %d\n", x[0], x[1], x[2], x[3]); }
259 
260  const char *AuxString() const { return aux_string; }
261 
262  void* V() {return v;}
263  const void* V() const {return v;}
264  void* Norm(){return norm;}
265  const void* Norm() const {return norm;}
266 
268  int EigvDim() const { return eigv_dim; }
269  int EigvId() const { return eigv_id; }
270  int EigvVolume() const { return eigv_volume; }
271  int EigvStride() const { return eigv_stride; }
272  int EigvLength() const { return eigv_length; }
273  int EigvRealLength() const { return eigv_real_length; }
274  int EigvTotalLength() const { return eigv_total_length; }
275 
276  size_t EigvBytes() const { return eigv_bytes; }
277  size_t EigvNormBytes() const { return eigv_norm_bytes; }
278  int EigvGhostLength() const { return eigv_ghost_length; }
279 
280  QudaDWFPCType DWFPCtype() const { return PCtype; }
281 
282  virtual QudaFieldLocation Location() const = 0;
284  QudaSiteOrder SiteOrder() const { return siteOrder; }
287 
288  int GhostLength() const { return ghost_length; }
289  const int *GhostFace() const { return ghostFace; }
290  int GhostOffset(const int i) const { return ghostOffset[i]; }
291  int GhostNormOffset(const int i ) const { return ghostNormOffset[i]; }
292  void* Ghost(const int i);
293  const void* Ghost(const int i) const;
294  void* GhostNorm(const int i);
295  const void* GhostNorm(const int i) const;
296 
297  friend std::ostream& operator<<(std::ostream &out, const ColorSpinorField &);
298  friend class ColorSpinorParam;
299  };
300 
301  // CUDA implementation
303 
304  friend class cpuColorSpinorField;
305 
306  private:
307  //void *v; // the field elements
308  //void *norm; // the normalization field
309  bool alloc; // whether we allocated memory
310  bool init;
311 
312  bool texInit; // whether a texture object has been created or not
313 #ifdef USE_TEXTURE_OBJECTS
314  cudaTextureObject_t tex;
315  cudaTextureObject_t texNorm;
316  void createTexObject();
317  void destroyTexObject();
318 #endif
319 
320 #ifdef GPU_COMMS // This is a hack for half precision.
321  // Since the ghost data and ghost norm data are not contiguous,
322  // separate MPI calls are needed when using GPUDirect RDMA.
323  void *my_fwd_norm_face[2][QUDA_MAX_DIM];
324  void *my_back_norm_face[2][QUDA_MAX_DIM];
325  void *from_fwd_norm_face[2][QUDA_MAX_DIM];
326  void *from_back_norm_face[2][QUDA_MAX_DIM];
327 
328  MsgHandle ***mh_recv_norm_fwd[2];
329  MsgHandle ***mh_recv_norm_back[2];
330  MsgHandle ***mh_send_norm_fwd[2];
331  MsgHandle ***mh_send_norm_back[2];
332 #endif
333 
334  bool reference; // whether the field is a reference or not
335 
336  static size_t ghostFaceBytes;
337  static void* ghostFaceBuffer[2]; // gpu memory
338  static void* fwdGhostFaceBuffer[2][QUDA_MAX_DIM]; // pointers to ghostFaceBuffer
339  static void* backGhostFaceBuffer[2][QUDA_MAX_DIM]; // pointers to ghostFaceBuffer
340  static int initGhostFaceBuffer;
341 
342  void create(const QudaFieldCreate);
343  void destroy();
344  void copy(const cudaColorSpinorField &);
345 
346  void zeroPad();
347 
352  void copySpinorField(const ColorSpinorField &src);
353 
354  void loadSpinorField(const ColorSpinorField &src);
355  void saveSpinorField (ColorSpinorField &src) const;
356 
358  bool initComms;
359 
361  size_t bufferMessageHandler;
362 
364  int nFaceComms; //FIXME - currently can only support one nFace in a field at once
365 
366 
367  public:
368 
369  static int bufferIndex;
370 
371  //cudaColorSpinorField();
376  virtual ~cudaColorSpinorField();
377 
381 
382  void switchBufferPinned();
383 
385  void createComms(int nFace);
387  void destroyComms();
388  void allocateGhostBuffer(int nFace);
389  static void freeGhostBuffer(void);
390 
404  void packGhost(const int nFace, const QudaParity parity, const int dim, const QudaDirection dir, const int dagger,
405  cudaStream_t* stream, void *buffer=0, double a=0, double b=0);
406 
407 
408  void packGhostExtended(const int nFace, const int R[], const QudaParity parity, const int dim, const QudaDirection dir,
409  const int dagger,cudaStream_t* stream, void *buffer=0);
410 
411 
412  void packGhost(FullClover &clov, FullClover &clovInv, const int nFace, const QudaParity parity, const int dim,
413  const QudaDirection dir, const int dagger, cudaStream_t* stream, void *buffer=0, double a=0);
414 
424  void sendGhost(void *ghost_spinor, const int nFace, const int dim, const QudaDirection dir,
425  const int dagger, cudaStream_t *stream);
426 
436  void unpackGhost(const void* ghost_spinor, const int nFace, const int dim,
437  const QudaDirection dir, const int dagger, cudaStream_t* stream);
438 
449  void unpackGhostExtended(const void* ghost_spinor, const int nFace, const QudaParity parity,
450  const int dim, const QudaDirection dir, const int dagger, cudaStream_t* stream);
451 
452 
453  void streamInit(cudaStream_t *stream_p);
454 
455  void pack(int nFace, int parity, int dagger, int stream_idx, bool zeroCopyPack,
456  double a=0, double b=0);
457 
458  void pack(FullClover &clov, FullClover &clovInv, int nFace, int parity, int dagger,
459  int stream_idx, bool zeroCopyPack, double a=0);
460 
461  void pack(int nFace, int parity, int dagger, cudaStream_t *stream_p, bool zeroCopyPack,
462  double a=0, double b=0);
463 
464  void pack(FullClover &clov, FullClover &clovInv, int nFace, int parity, int dagger,
465  cudaStream_t *stream_p, bool zeroCopyPack, double a=0);
466 
467  void packExtended(const int nFace, const int R[], const int parity, const int dagger,
468  const int dim, cudaStream_t *stream_p, const bool zeroCopyPack=false);
469 
470  void gather(int nFace, int dagger, int dir, cudaStream_t *stream_p=NULL);
471 
472  void recvStart(int nFace, int dir, int dagger=0);
473  void sendStart(int nFace, int dir, int dagger=0);
474  void commsStart(int nFace, int dir, int dagger=0);
475  int commsQuery(int nFace, int dir, int dagger=0);
476  void scatter(int nFace, int dagger, int dir, cudaStream_t *stream_p);
477  void scatter(int nFace, int dagger, int dir);
478 
479  void scatterExtended(int nFace, int parity, int dagger, int dir);
480 
481 
482 
483 #ifdef USE_TEXTURE_OBJECTS
484  const cudaTextureObject_t& Tex() const { return tex; }
485  const cudaTextureObject_t& TexNorm() const { return texNorm; }
486 #endif
487 
488  cudaColorSpinorField& Even() const;
489  cudaColorSpinorField& Odd() const;
490 
491  cudaColorSpinorField& Eigenvec(const int idx) const;
492  void CopyEigenvecSubset(cudaColorSpinorField& dst, const int range, const int first_element=0) const;
493 
494  void zero();
495 
496  QudaFieldLocation Location() const;
497 
503  bool isNative() const;
504 
505  friend std::ostream& operator<<(std::ostream &out, const cudaColorSpinorField &);
506 
507  void getTexObjectInfo() const;
508  };
509 
510  // Forward declaration of accessor functors
511  template <typename Float> class ColorSpinorFieldOrder;
512  template <typename Float> class SpaceColorSpinOrder;
513  template <typename Float> class SpaceSpinColorOrder;
514  template <typename Float> class QOPDomainWallOrder;
515 
516  // CPU implementation
518 
519  friend class cudaColorSpinorField;
520 
521  template <typename Float> friend class SpaceColorSpinOrder;
522  template <typename Float> friend class SpaceSpinColorOrder;
523  template <typename Float> friend class QOPDomainWallOrder;
524 
525  public:
526  static void* fwdGhostFaceBuffer[QUDA_MAX_DIM]; //cpu memory
527  static void* backGhostFaceBuffer[QUDA_MAX_DIM]; //cpu memory
528  static void* fwdGhostFaceSendBuffer[QUDA_MAX_DIM]; //cpu memory
529  static void* backGhostFaceSendBuffer[QUDA_MAX_DIM]; //cpu memory
531 
532  private:
533  //void *v; // the field elements
534  //void *norm; // the normalization field
535  bool init;
536  bool reference; // whether the field is a reference or not
537 
538  void create(const QudaFieldCreate);
539  void destroy();
540 
541  public:
542  //cpuColorSpinorField();
546  virtual ~cpuColorSpinorField();
547 
551 
552  //cpuColorSpinorField& Even() const;
553  //cpuColorSpinorField& Odd() const;
554 
555  void Source(const QudaSourceType sourceType, const int st=0, const int s=0, const int c=0);
556  static int Compare(const cpuColorSpinorField &a, const cpuColorSpinorField &b, const int resolution=1);
557  void PrintVector(unsigned int x);
558 
559  void allocateGhostBuffer(void);
560  static void freeGhostBuffer(void);
561 
562  void packGhost(void* ghost_spinor, const int dim,
563  const QudaDirection dir, const QudaParity parity, const int dagger);
564  void unpackGhost(void* ghost_spinor, const int dim,
565  const QudaDirection dir, const int dagger);
566 
567  void copy(const cpuColorSpinorField&);
568  void zero();
569 
570  QudaFieldLocation Location() const;
571  };
572 
574  QudaFieldLocation location, void *Dst=0, void *Src=0,
575  void *dstNorm=0, void*srcNorm=0);
576  void genericSource(cpuColorSpinorField &a, QudaSourceType sourceType, int x, int s, int c);
577  int genericCompare(const cpuColorSpinorField &a, const cpuColorSpinorField &b, int tol);
578  void genericPrintVector(cpuColorSpinorField &a, unsigned int x);
579 
580  void exchangeExtendedGhost(cudaColorSpinorField* spinor, int R[], int parity, cudaStream_t *stream_p);
581 
583  QudaFieldLocation location, const int parity, void *Dst, void *Src, void *dstNorm, void *srcNorm);
584 
585 
586 } // namespace quda
587 
588 #endif // _COLOR_SPINOR_FIELD_H
QudaDslashType dslash_type
Definition: test_util.cpp:1560
QudaDiracFieldOrder dirac_order
Definition: quda.h:156
int genericCompare(const cpuColorSpinorField &a, const cpuColorSpinorField &b, int tol)
void setPrecision(QudaPrecision precision)
void copy(const cpuColorSpinorField &)
friend std::ostream & operator<<(std::ostream &out, const ColorSpinorField &)
void unpackGhost(void *ghost_spinor, const int dim, const QudaDirection dir, const int dagger)
enum QudaPrecision_s QudaPrecision
int V
Definition: test_util.cpp:29
void streamInit(cudaStream_t *stream_p)
void packGhost(const int nFace, const QudaParity parity, const int dim, const QudaDirection dir, const int dagger, cudaStream_t *stream, void *buffer=0, double a=0, double b=0)
ColorSpinorParam(ColorSpinorParam &cpuParam, QudaInvertParam &inv_param)
int ghostFace[QUDA_MAX_DIM]
void * ghost[QUDA_MAX_DIM]
void * ghostNorm[QUDA_MAX_DIM]
const int * X() const
void unpackGhost(const void *ghost_spinor, const int nFace, const int dim, const QudaDirection dir, const int dagger, cudaStream_t *stream)
void gather(int nFace, int dagger, int dir, cudaStream_t *stream_p=NULL)
#define errorQuda(...)
Definition: util_quda.h:73
QudaDslashType dslash_type
Definition: quda.h:85
enum QudaFieldOrder_s QudaFieldOrder
void sendGhost(void *ghost_spinor, const int nFace, const int dim, const QudaDirection dir, const int dagger, cudaStream_t *stream)
int commsQuery(int nFace, int dir, int dagger=0)
void fill(ColorSpinorParam &) const
void pack(int nFace, int parity, int dagger, int stream_idx, bool zeroCopyPack, double a=0, double b=0)
int ghostOffset[QUDA_MAX_DIM]
cpuColorSpinorField(const cpuColorSpinorField &)
cudaStream_t * stream
void Source(const QudaSourceType sourceType, const int st=0, const int s=0, const int c=0)
enum QudaSiteOrder_s QudaSiteOrder
QudaDWFPCType DWFPCtype() const
__host__ __device__ void copy(T1 &a, const T2 &b)
QudaPrecision precision
Definition: lattice_field.h:41
int GhostNormOffset(const int i) const
void copyGenericColorSpinor(ColorSpinorField &dst, const ColorSpinorField &src, QudaFieldLocation location, void *Dst=0, void *Src=0, void *dstNorm=0, void *srcNorm=0)
void unpackGhostExtended(const void *ghost_spinor, const int nFace, const QudaParity parity, const int dim, const QudaDirection dir, const int dagger, cudaStream_t *stream)
int eigv_dim
used for eigcg:
void scatterExtended(int nFace, int parity, int dagger, int dir)
cpuColorSpinorField * spinor
Definition: dslash_test.cpp:40
void packGhost(void *ghost_spinor, const int dim, const QudaDirection dir, const QudaParity parity, const int dagger)
QudaPrecision cpu_prec
Definition: dslash_test.cpp:34
void sendStart(int nFace, int dir, int dagger=0)
QudaSiteSubset siteSubset
Definition: lattice_field.h:42
ColorSpinorField(const ColorSpinorField &)
QudaDagType dagger
Definition: test_util.cpp:1558
enum QudaSourceType_s QudaSourceType
cudaColorSpinorField & Odd() const
int x[QUDA_MAX_DIM]
Definition: lattice_field.h:38
enum QudaDirection_s QudaDirection
static void * backGhostFaceSendBuffer[QUDA_MAX_DIM]
enum QudaDWFPCType_s QudaDWFPCType
ColorSpinorField * odd
const char * AuxString() const
const QudaFieldLocation location
Definition: pack_test.cpp:46
std::vector< ColorSpinorField * > eigenvectors
for eigcg:
static void checkField(const ColorSpinorField &, const ColorSpinorField &)
void recvStart(int nFace, int dir, int dagger=0)
void reset(const ColorSpinorParam &)
void CopyEigenvecSubset(cudaColorSpinorField &dst, const int range, const int first_element=0) const
friend std::ostream & operator<<(std::ostream &out, const cudaColorSpinorField &)
void * GhostNorm(const int i)
void packExtended(const int nFace, const int R[], const int parity, const int dagger, const int dim, cudaStream_t *stream_p, const bool zeroCopyPack=false)
void exchangeExtendedGhost(cudaColorSpinorField *spinor, int R[], int parity, cudaStream_t *stream_p)
int EigvDim() const
for eigcg only:
ColorSpinorField & operator=(const ColorSpinorField &)
static void * backGhostFaceBuffer[QUDA_MAX_DIM]
int ghostNormOffset[QUDA_MAX_DIM]
enum QudaParity_s QudaParity
void genericSource(cpuColorSpinorField &a, QudaSourceType sourceType, int x, int s, int c)
static void * fwdGhostFaceSendBuffer[QUDA_MAX_DIM]
QudaTwistFlavorType twist_flavor
Definition: test_util.cpp:1570
QudaSiteOrder SiteOrder() const
QudaTwistFlavorType twistFlavor
int x[4]
static void * fwdGhostFaceBuffer[QUDA_MAX_DIM]
static int Compare(const cpuColorSpinorField &a, const cpuColorSpinorField &b, const int resolution=1)
QudaFieldOrder FieldOrder() const
void genericPrintVector(cpuColorSpinorField &a, unsigned int x)
const void * V() const
virtual ColorSpinorField & operator=(const ColorSpinorField &)
void * Ghost(const int i)
enum QudaSiteSubset_s QudaSiteSubset
QudaPrecision cuda_prec
Definition: dslash_test.cpp:35
size_t EigvNormBytes() const
QudaFieldLocation Location() const
virtual QudaFieldLocation Location() const =0
enum QudaFieldLocation_s QudaFieldLocation
char aux_string[TuneKey::aux_n]
QudaInvertParam inv_param
Definition: dslash_test.cpp:38
cpuColorSpinorField * out
const void * Norm() const
void copyExtendedColorSpinor(ColorSpinorField &dst, const ColorSpinorField &src, QudaFieldLocation location, const int parity, void *Dst, void *Src, void *dstNorm, void *srcNorm)
enum QudaGammaBasis_s QudaGammaBasis
Main header file for the QUDA library.
QudaPrecision Precision() const
static const int aux_n
Definition: tune_key.h:12
#define printfQuda(...)
Definition: util_quda.h:67
QudaGammaBasis GammaBasis() const
QudaTwistFlavorType twistFlavor
QudaTwistFlavorType TwistFlavor() const
void packGhostExtended(const int nFace, const int R[], const QudaParity parity, const int dim, const QudaDirection dir, const int dagger, cudaStream_t *stream, void *buffer=0)
enum QudaFieldCreate_s QudaFieldCreate
void init(int argc, char **argv)
Definition: dslash_test.cpp:79
ColorSpinorParam(void *V, QudaInvertParam &inv_param, const int *X, const bool pc_solution)
cudaColorSpinorField(const cudaColorSpinorField &)
#define QUDA_MAX_DIM
Maximum number of dimensions supported by QUDA. In practice, no routines make use of more than 5...
QudaFieldLocation Location() const
void commsStart(int nFace, int dir, int dagger=0)
VOLATILE spinorFloat * s
void scatter(int nFace, int dagger, int dir, cudaStream_t *stream_p)
int GhostOffset(const int i) const
void initComms(int argc, char **argv, const int *commDims)
Definition: test_util.cpp:48
QudaSiteSubset SiteSubset() const
const QudaParity parity
Definition: dslash_test.cpp:29
ColorSpinorField * even
cudaColorSpinorField & Even() const
const int * GhostFace() const
ColorSpinorField & operator=(const ColorSpinorField &)
enum QudaTwistFlavorType_s QudaTwistFlavorType
cudaColorSpinorField & Eigenvec(const int idx) const
for deflated solvers: