QUDA  0.9.0
covdev_reference.cpp
Go to the documentation of this file.
1 #include <stdio.h>
2 #include <stdlib.h>
3 #include <math.h>
4 #include <string.h>
5 
6 #include <test_util.h>
7 #include <quda_internal.h>
8 #include <quda.h>
9 #include <util_quda.h>
10 #include <covdev_reference.h>
11 #include "misc.h"
12 #include <blas_quda.h>
13 
14 #include <blas_reference.h>
15 
16 extern void *memset(void *s, int c, size_t n);
17 
18 #include <dslash_util.h>
19 
20 //
21 // covdevReference()
22 //
23 // if oddBit is zero: calculate even parity spinor elements (using odd parity spinor)
24 // if oddBit is one: calculate odd parity spinor elements
25 //
26 // if daggerBit is zero: perform ordinary covariant derivative operator
27 // if daggerBit is one: perform hermitian covariant derivative operator
28 //
29 template<typename Float>
30 void display_link_internal(Float* link)
31 {
32  int i, j;
33 
34  for (i = 0;i < 3; i++){
35  for(j=0;j < 3; j++){
36  printf("(%10f,%10f) \t", link[i*3*2 + j*2], link[i*3*2 + j*2 + 1]);
37  }
38  printf("\n");
39  }
40  printf("\n");
41  return;
42 }
43 
44 
45 template <typename sFloat, typename gFloat>
46 void covdevReference(sFloat *res, gFloat **link, sFloat *spinorField,
47  int oddBit, int daggerBit, int mu)
48 {
49  for (int i=0; i<Vh*mySpinorSiteSize; i++) res[i] = 0.0;
50 
51  gFloat *linkEven[4], *linkOdd[4];
52 
53  for (int dir = 0; dir < 4; dir++) {
54  linkEven[dir] = link[dir];
55  linkOdd[dir] = link[dir] + Vh*gaugeSiteSize;
56  }
57 
58  for (int sid = 0; sid < Vh; sid++) {
60 
61  sFloat gaugedSpinor[mySpinorSiteSize];
62 
63  gFloat *lnk = gaugeLink(sid, mu, oddBit, linkEven, linkOdd, 1);
64  sFloat *spinor = spinorNeighbor(sid, mu, oddBit, spinorField, 1);
65 
66  if (daggerBit) {
67  for (int s = 0; s < 4; s++)
68  su3Tmul(&gaugedSpinor[s*6], lnk, &spinor[s*6]);
69  } else {
70  for (int s = 0; s < 4; s++)
71  su3Mul (&gaugedSpinor[s*6], lnk, &spinor[s*6]);
72  }
73 
74  sum(&res[offset], &res[offset], gaugedSpinor, mySpinorSiteSize);
75  } // 4-d volume
76 }
77 
78 
79 
80 
81 void covdev_dslash(void *res, void **link, void *spinorField, int oddBit, int daggerBit, int mu,
82  QudaPrecision sPrecision, QudaPrecision gPrecision) {
83 
84  if (sPrecision == QUDA_DOUBLE_PRECISION) {
85  if (gPrecision == QUDA_DOUBLE_PRECISION){
86  covdevReference((double*)res, (double**)link, (double*)spinorField, oddBit, daggerBit, mu);
87  } else {
88  covdevReference((double*)res, (float**) link, (double*)spinorField, oddBit, daggerBit, mu);
89  }
90  }
91  else {
92  if (gPrecision == QUDA_DOUBLE_PRECISION){
93  covdevReference((float*)res, (double**)link, (float*)spinorField, oddBit, daggerBit, mu);
94  } else {
95  covdevReference((float*)res, (float**) link, (float*)spinorField, oddBit, daggerBit, mu);
96  }
97  }
98 }
99 
100 
101 
102 
103 template <typename sFloat, typename gFloat>
104 void Mat(sFloat *out, gFloat **link, sFloat *in, int daggerBit, int mu)
105 {
106  sFloat *inEven = in;
107  sFloat *inOdd = in + Vh*mySpinorSiteSize;
108  sFloat *outEven = out;
109  sFloat *outOdd = out + Vh*mySpinorSiteSize;
110 
111  // full dslash operator
112  covdevReference(outOdd, link, inEven, 1, daggerBit, mu);
113  covdevReference(outEven, link, inOdd, 0, daggerBit, mu);
114 }
115 
116 
117 void mat(void *out, void **link, void *in, int dagger_bit, int mu,
118  QudaPrecision sPrecision, QudaPrecision gPrecision)
119 {
120 
121  if (sPrecision == QUDA_DOUBLE_PRECISION){
122  if (gPrecision == QUDA_DOUBLE_PRECISION) {
123  Mat((double*)out, (double**)link, (double*)in, dagger_bit, mu);
124  } else {
125  Mat((double*)out, (float**) link, (double*)in, dagger_bit, mu);
126  }
127  } else {
128  if (gPrecision == QUDA_DOUBLE_PRECISION){
129  Mat((float*)out, (double**)link, (float*)in, dagger_bit, mu);
130  } else {
131  Mat((float*)out, (float**) link, (float*)in, dagger_bit, mu);
132  }
133  }
134 }
135 
136 
137 
138 template <typename sFloat, typename gFloat>
139 void Matdagmat(sFloat *out, gFloat **link, sFloat *in, int daggerBit, int mu, sFloat* tmp, QudaParity parity)
140 {
141  switch(parity){
142  case QUDA_EVEN_PARITY:
143  {
144  sFloat *inEven = in;
145  sFloat *outEven = out;
146  covdevReference(tmp, link, inEven, 1, daggerBit, mu);
147  covdevReference(outEven, link, tmp, 0, daggerBit, mu);
148  break;
149  }
150  case QUDA_ODD_PARITY:
151  {
152  sFloat *inOdd = in;
153  sFloat *outOdd = out;
154  covdevReference(tmp, link, inOdd, 0, daggerBit, mu);
155  covdevReference(outOdd, link, tmp, 1, daggerBit, mu);
156  break;
157  }
158 
159  default:
160  fprintf(stderr, "ERROR: invalid parity in %s,line %d\n", __FUNCTION__, __LINE__);
161  break;
162  }
163 
164 }
165 
166 
167 
168 void matdagmat(void *out, void **link, void *in, int dagger_bit, int mu,
169  QudaPrecision sPrecision, QudaPrecision gPrecision, void *tmp, QudaParity parity)
170 {
171  if (sPrecision == QUDA_DOUBLE_PRECISION) {
172  if (gPrecision == QUDA_DOUBLE_PRECISION) {
173  Matdagmat((double*)out, (double**)link, (double*)in, dagger_bit, mu, (double*)tmp, parity);
174  } else {
175  Matdagmat((double*)out, (float**) link, (double*)in, dagger_bit, mu, (double*)tmp, parity);
176  }
177  } else {
178  if (gPrecision == QUDA_DOUBLE_PRECISION){
179  Matdagmat((float*)out, (double**)link, (float*)in, dagger_bit, mu, (float*)tmp, parity);
180  } else {
181  Matdagmat((float*)out, (float**) link, (float*)in, dagger_bit, mu, (float*)tmp, parity);
182  }
183  }
184 }
185 
186 #ifdef MULTI_GPU
187 
188 template <typename sFloat, typename gFloat>
189 void covdevReference_mg4dir(sFloat *res, gFloat **link, gFloat **ghostLink, sFloat *spinorField,
190  sFloat **fwd_nbr_spinor, sFloat **back_nbr_spinor,
191  int oddBit, int daggerBit, int mu)
192 {
193  for (int i=0; i<Vh*mySpinorSiteSize; i++) res[i] = 0.0;
194 
195  gFloat *linkEven[4], *linkOdd[4];
196  gFloat *ghostLinkEven[4], *ghostLinkOdd[4];
197 
198  for (int dir = 0; dir < 4; dir++) {
199  linkEven[dir] = link[dir];
200  linkOdd[dir] = link[dir] + Vh*gaugeSiteSize;
201 
202  ghostLinkEven[dir] = ghostLink[dir];
203  ghostLinkOdd[dir] = ghostLink[dir] + (faceVolume[dir]/2)*gaugeSiteSize;
204  }
205 
206  for (int sid = 0; sid < Vh; sid++) {
208 
209  gFloat *lnk = gaugeLink_mg4dir(sid, mu, oddBit, linkEven, linkOdd, ghostLinkEven, ghostLinkOdd, 1, 1);
210  sFloat *spinor = spinorNeighbor_mg4dir(sid, mu, oddBit, spinorField, fwd_nbr_spinor, back_nbr_spinor, 1, 1);
211 
212  sFloat gaugedSpinor[mySpinorSiteSize];
213 
214  if (daggerBit) {
215  for (int s = 0; s < 4; s++)
216  su3Tmul(&gaugedSpinor[s*6], lnk, &spinor[s*6]);
217  } else {
218  for (int s = 0; s < 4; s++)
219  su3Mul (&gaugedSpinor[s*6], lnk, &spinor[s*6]);
220  }
221  sum(&res[offset], &res[offset], gaugedSpinor, mySpinorSiteSize);
222  } // 4-d volume
223 }
224 
225 void covdev_dslash_mg4dir(cpuColorSpinorField* out, void **link, void** ghostLink,
226  cpuColorSpinorField* in, int oddBit, int daggerBit, int mu,
227  QudaPrecision sPrecision, QudaPrecision gPrecision)
228 {
229  QudaParity otherparity = QUDA_INVALID_PARITY;
230  if (oddBit == QUDA_EVEN_PARITY) {
231  otherparity = QUDA_ODD_PARITY;
232  } else if (oddBit == QUDA_ODD_PARITY) {
233  otherparity = QUDA_EVEN_PARITY;
234  } else {
235  errorQuda("ERROR: full parity not supported in function %s", __FUNCTION__);
236  }
237  const int nFace = 1;
238 
239  in->exchangeGhost(otherparity, nFace, daggerBit);
240 
241  void** fwd_nbr_spinor = in->fwdGhostFaceBuffer;
242  void** back_nbr_spinor = in->backGhostFaceBuffer;
243 
244  if (sPrecision == QUDA_DOUBLE_PRECISION) {
245  if (gPrecision == QUDA_DOUBLE_PRECISION) {
246  covdevReference_mg4dir((double*)out->V(), (double**)link, (double**)ghostLink, (double*)in->V(),
247  (double**)fwd_nbr_spinor, (double**)back_nbr_spinor, oddBit, daggerBit, mu);
248  } else {
249  covdevReference_mg4dir((double*)out->V(), (float**) link, (float**) ghostLink, (double*)in->V(),
250  (double**)fwd_nbr_spinor, (double**)back_nbr_spinor, oddBit, daggerBit, mu);
251  }
252  } else {
253  if (gPrecision == QUDA_DOUBLE_PRECISION) {
254  covdevReference_mg4dir((float*)out->V(), (double**)link, (double**)ghostLink, (float*)in->V(),
255  (float**)fwd_nbr_spinor, (float**)back_nbr_spinor, oddBit, daggerBit, mu);
256  } else {
257  covdevReference_mg4dir((float*)out->V(), (float**)link, (float**)ghostLink, (float*)in->V(),
258  (float**)fwd_nbr_spinor, (float**)back_nbr_spinor, oddBit, daggerBit, mu);
259  }
260  }
261 
262 }
263 
264 template <typename sFloat, typename gFloat>
265 void Mat_mg4dir(cpuColorSpinorField *out, gFloat **link, gFloat **ghostLink, cpuColorSpinorField *in, int daggerBit, int mu)
266 {
267  const int nFace = 1;
268  {
269  cpuColorSpinorField &inEven = static_cast<cpuColorSpinorField&>(in->Even());
270  cpuColorSpinorField &outOdd = static_cast<cpuColorSpinorField&>(out->Odd());
271 
272  inEven.exchangeGhost(QUDA_EVEN_PARITY, nFace, daggerBit);
273 
274  covdevReference_mg4dir(reinterpret_cast<sFloat*>(outOdd.V()), link, ghostLink,
275  reinterpret_cast<sFloat*>(inEven.V()),
276  reinterpret_cast<sFloat**>(inEven.fwdGhostFaceBuffer),
277  reinterpret_cast<sFloat**>(inEven.backGhostFaceBuffer),
278  1, daggerBit, mu);
279  }
280 
281  {
282  cpuColorSpinorField &inOdd = static_cast<cpuColorSpinorField&>(in->Odd());
283  cpuColorSpinorField &outEven = static_cast<cpuColorSpinorField&>(out->Even());
284 
285  inOdd.exchangeGhost(QUDA_ODD_PARITY, nFace, daggerBit);
286 
287  covdevReference_mg4dir(reinterpret_cast<sFloat*>(outEven.V()), link, ghostLink,
288  reinterpret_cast<sFloat*>(inOdd.V()),
289  reinterpret_cast<sFloat**>(inOdd.fwdGhostFaceBuffer),
290  reinterpret_cast<sFloat**>(inOdd.backGhostFaceBuffer),
291  0, daggerBit, mu);
292  }
293 }
294 
295 
296 void mat_mg4dir(cpuColorSpinorField *out, void **link, void **ghostLink, cpuColorSpinorField *in, int dagger_bit, int mu,
297  QudaPrecision sPrecision, QudaPrecision gPrecision)
298 {
299 
300  if (sPrecision == QUDA_DOUBLE_PRECISION){
301  if (gPrecision == QUDA_DOUBLE_PRECISION) {
302  Mat_mg4dir<double, double>(out, (double**)link, (double**) ghostLink, in, dagger_bit, mu);
303  } else {
304  Mat_mg4dir<double, float> (out, (float**) link, (float**) ghostLink, in, dagger_bit, mu);
305  }
306  } else {
307  if (gPrecision == QUDA_DOUBLE_PRECISION){
308  Mat_mg4dir<float, double> (out, (double**)link, (double**) ghostLink, in, dagger_bit, mu);
309  } else {
310  Mat_mg4dir<float, float> (out, (float**) link, (float**) ghostLink, in, dagger_bit, mu);
311  }
312  }
313 }
314 
315 
316 
317 void matdagmat_mg4dir(cpuColorSpinorField* out, void **link, void** ghostLink, cpuColorSpinorField* in,
318  int dagger_bit, int mu, QudaPrecision sPrecision, QudaPrecision gPrecision,
320 {
321  //assert sPrecision and gPrecision must be the same
322  if (sPrecision != gPrecision){
323  errorQuda("Spinor precision and gPrecison is not the same");
324  }
325 
326  QudaParity otherparity = QUDA_INVALID_PARITY;
327  if (parity == QUDA_EVEN_PARITY){
328  otherparity = QUDA_ODD_PARITY;
329  } else if (parity == QUDA_ODD_PARITY) {
330  otherparity = QUDA_EVEN_PARITY;
331  } else {
332  errorQuda("ERROR: full parity not supported in function %s\n", __FUNCTION__);
333  }
334 
335  covdev_dslash_mg4dir(tmp, link, ghostLink, in, otherparity, dagger_bit, mu, sPrecision, gPrecision);
336 
337  covdev_dslash_mg4dir(out, link, ghostLink, tmp, parity, dagger_bit, mu, sPrecision, gPrecision);
338 }
339 
340 #endif
341 
double mu
Definition: test_util.cpp:1643
enum QudaPrecision_s QudaPrecision
void covdevReference(sFloat *res, gFloat **link, sFloat *spinorField, int oddBit, int daggerBit, int mu)
#define errorQuda(...)
Definition: util_quda.h:90
void covdev_dslash(void *res, void **link, void *spinorField, int oddBit, int daggerBit, int mu, QudaPrecision sPrecision, QudaPrecision gPrecision)
cudaColorSpinorField * tmp
Definition: covdev_test.cpp:44
const ColorSpinorField & Even() const
const ColorSpinorField & Odd() const
void matdagmat_mg4dir(cpuColorSpinorField *out, void **link, void **ghostLink, cpuColorSpinorField *in, int dagger_bit, int mu, QudaPrecision sPrecision, QudaPrecision gPrecision, cpuColorSpinorField *tmp, QudaParity parity)
void Mat(sFloat *out, gFloat **link, sFloat *in, int daggerBit, int mu)
void Matdagmat(sFloat *out, gFloat **link, sFloat *in, int daggerBit, int mu, sFloat *tmp, QudaParity parity)
void exchangeGhost(QudaParity parity, int nFace, int dagger, const MemoryLocation *pack_destination=nullptr, const MemoryLocation *halo_location=nullptr, bool gdr_send=false, bool gdr_recv=false) const
This is a unified ghost exchange function for doing a complete halo exchange regardless of the type o...
static Float * spinorNeighbor(int i, int dir, int oddBit, Float *spinorField, int neighbor_distance)
Definition: dslash_util.h:127
size_t size_t offset
int printf(const char *,...) __attribute__((__format__(__printf__
void display_link_internal(Float *link)
__host__ __device__ void sum(double &a, double &b)
cpuColorSpinorField * in
#define mySpinorSiteSize
#define gaugeSiteSize
Definition: test_util.h:6
void matdagmat(void *out, void **link, void *in, int dagger_bit, int mu, QudaPrecision sPrecision, QudaPrecision gPrecision, void *tmp, QudaParity parity)
static void * backGhostFaceBuffer[QUDA_MAX_DIM]
enum QudaParity_s QudaParity
int fprintf(FILE *, const char *,...) __attribute__((__format__(__printf__
static void * fwdGhostFaceBuffer[QUDA_MAX_DIM]
void * memset(void *s, int c, size_t n)
static Float * gaugeLink(int i, int dir, int oddBit, Float **gaugeEven, Float **gaugeOdd, int nbr_distance)
Definition: dslash_util.h:104
cpuColorSpinorField * out
void covdev_dslash_mg4dir(cpuColorSpinorField *out, void **link, void **ghostLink, cpuColorSpinorField *in, int oddBit, int daggerBit, int mu, QudaPrecision sPrecision, QudaPrecision gPrecision)
Main header file for the QUDA library.
int Vh
Definition: test_util.cpp:29
int faceVolume[4]
Definition: test_util.cpp:32
const void * c
static void su3Mul(sFloat *res, gFloat *mat, sFloat *vec)
Definition: dslash_util.h:80
static void su3Tmul(sFloat *res, gFloat *mat, sFloat *vec)
Definition: dslash_util.h:85
void mat(void *out, void **link, void *in, int dagger_bit, int mu, QudaPrecision sPrecision, QudaPrecision gPrecision)
void mat_mg4dir(cpuColorSpinorField *out, void **link, void **ghostLink, cpuColorSpinorField *in, int daggerBit, int mu, QudaPrecision sPrecision, QudaPrecision gPrecision)
QudaParity parity
Definition: covdev_test.cpp:53
cpuColorSpinorField * spinor
Definition: covdev_test.cpp:41