16 extern void *
memset(
void *s,
int c,
size_t n);
26 template<
typename Float>
31 for (i = 0;i < 3; i++){
33 printf(
"(%10f,%10f) \t", link[i*3*2 + j*2], link[i*3*2 + j*2 + 1]);
42 template <
typename sFloat,
typename gFloat>
44 int oddBit,
int daggerBit,
int mu)
48 gFloat *linkEven[4], *linkOdd[4];
50 for (
int dir = 0; dir < 4; dir++) {
51 linkEven[dir] = link[dir];
55 for (
int sid = 0; sid <
Vh; sid++) {
60 gFloat *lnk = gaugeLink(sid,
mu, oddBit, linkEven, linkOdd, 1);
61 sFloat *
spinor = spinorNeighbor(sid,
mu, oddBit, spinorField, 1);
64 for (
int s = 0; s < 4; s++)
65 su3Tmul(&gaugedSpinor[s*6], lnk, &
spinor[s*6]);
67 for (
int s = 0; s < 4; s++)
68 su3Mul (&gaugedSpinor[s*6], lnk, &
spinor[s*6]);
75 void covdev_dslash(
void *res,
void **link,
void *spinorField,
int oddBit,
int daggerBit,
int mu,
81 covdevReference((
double*)res, (
double**)link, (
double*)spinorField, oddBit, daggerBit,
mu);
83 covdevReference((
double*)res, (
float**) link, (
double*)spinorField, oddBit, daggerBit,
mu);
88 covdevReference((
float*)res, (
double**)link, (
float*)spinorField, oddBit, daggerBit,
mu);
90 covdevReference((
float*)res, (
float**) link, (
float*)spinorField, oddBit, daggerBit,
mu);
95 template <
typename sFloat,
typename gFloat>
96 void Mat(sFloat *out, gFloat **link, sFloat *in,
int daggerBit,
int mu)
100 sFloat *outEven = out;
109 void mat(
void *out,
void **link,
void *in,
int dagger_bit,
int mu,
115 Mat((
double*)out, (
double**)link, (
double*)in, dagger_bit,
mu);
117 Mat((
double*)out, (
float**) link, (
double*)in, dagger_bit,
mu);
121 Mat((
float*)out, (
double**)link, (
float*)in, dagger_bit,
mu);
123 Mat((
float*)out, (
float**) link, (
float*)in, dagger_bit,
mu);
130 template <
typename sFloat,
typename gFloat>
137 sFloat *outEven = out;
145 sFloat *outOdd = out;
152 fprintf(stderr,
"ERROR: invalid parity in %s,line %d\n", __FUNCTION__, __LINE__);
160 void matdagmat(
void *out,
void **link,
void *in,
int dagger_bit,
int mu,
180 template <
typename sFloat,
typename gFloat>
181 void covdevReference_mg4dir(sFloat *res, gFloat **link, gFloat **
ghostLink, sFloat *spinorField,
182 sFloat **fwd_nbr_spinor, sFloat **back_nbr_spinor,
int oddBit,
int daggerBit,
int mu)
186 gFloat *linkEven[4], *linkOdd[4];
187 gFloat *ghostLinkEven[4], *ghostLinkOdd[4];
189 for (
int dir = 0; dir < 4; dir++) {
190 linkEven[dir] = link[dir];
197 for (
int sid = 0; sid <
Vh; sid++) {
200 gFloat *lnk = gaugeLink_mg4dir(sid,
mu, oddBit, linkEven, linkOdd, ghostLinkEven, ghostLinkOdd, 1, 1);
201 sFloat *
spinor = spinorNeighbor_mg4dir(sid,
mu, oddBit, spinorField, fwd_nbr_spinor, back_nbr_spinor, 1, 1);
206 for (
int s = 0; s < 4; s++)
207 su3Tmul(&gaugedSpinor[s*6], lnk, &
spinor[s*6]);
209 for (
int s = 0; s < 4; s++)
210 su3Mul (&gaugedSpinor[s*6], lnk, &
spinor[s*6]);
226 errorQuda(
"ERROR: full parity not supported in function %s", __FUNCTION__);
237 covdevReference_mg4dir((
double*)out->
V(), (
double**)link, (
double**)
ghostLink, (
double*)in->
V(),
238 (
double**)fwd_nbr_spinor, (
double**)back_nbr_spinor, oddBit, daggerBit,
mu);
240 covdevReference_mg4dir((
double*)out->
V(), (
float**) link, (
float**)
ghostLink, (
double*)in->
V(),
241 (
double**)fwd_nbr_spinor, (
double**)back_nbr_spinor, oddBit, daggerBit,
mu);
245 covdevReference_mg4dir((
float*)out->
V(), (
double**)link, (
double**)
ghostLink, (
float*)in->
V(),
246 (
float**)fwd_nbr_spinor, (
float**)back_nbr_spinor, oddBit, daggerBit,
mu);
248 covdevReference_mg4dir((
float*)out->
V(), (
float**)link, (
float**)
ghostLink, (
float*)in->
V(),
249 (
float**)fwd_nbr_spinor, (
float**)back_nbr_spinor, oddBit, daggerBit,
mu);
255 template <
typename sFloat,
typename gFloat>
265 covdevReference_mg4dir(
reinterpret_cast<sFloat*
>(outOdd.
V()), link,
ghostLink,
266 reinterpret_cast<sFloat*
>(inEven.
V()),
278 covdevReference_mg4dir(
reinterpret_cast<sFloat*
>(outEven.
V()), link,
ghostLink,
279 reinterpret_cast<sFloat*
>(inOdd.
V()),
293 Mat_mg4dir<double, double>(out, (
double**)link, (
double**)
ghostLink, in, dagger_bit,
mu);
295 Mat_mg4dir<double, float> (out, (
float**) link, (
float**)
ghostLink, in, dagger_bit,
mu);
299 Mat_mg4dir<float, double> (out, (
double**)link, (
double**)
ghostLink, in, dagger_bit,
mu);
301 Mat_mg4dir<float, float> (out, (
float**) link, (
float**)
ghostLink, in, dagger_bit,
mu);
313 if (sPrecision != gPrecision){
314 errorQuda(
"Spinor precision and gPrecison is not the same");
323 errorQuda(
"ERROR: full parity not supported in function %s\n", __FUNCTION__);
const ColorSpinorField & Odd() const
const ColorSpinorField & Even() const
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, QudaPrecision ghost_precision=QUDA_INVALID_PRECISION) const
This is a unified ghost exchange function for doing a complete halo exchange regardless of the type o...
static void * fwdGhostFaceBuffer[QUDA_MAX_DIM]
static void * backGhostFaceBuffer[QUDA_MAX_DIM]
void covdevReference(sFloat *res, gFloat **link, sFloat *spinorField, int oddBit, int daggerBit, int mu)
void Mat(sFloat *out, gFloat **link, sFloat *in, int daggerBit, int mu)
void display_link_internal(Float *link)
void Matdagmat(sFloat *out, gFloat **link, sFloat *in, int daggerBit, int mu, sFloat *tmp, QudaParity parity)
void matdagmat(void *out, void **link, void *in, int dagger_bit, int mu, QudaPrecision sPrecision, QudaPrecision gPrecision, void *tmp, QudaParity parity)
void covdev_dslash(void *res, void **link, void *spinorField, int oddBit, int daggerBit, int mu, QudaPrecision sPrecision, QudaPrecision gPrecision)
void mat(void *out, void **link, void *in, int dagger_bit, int mu, QudaPrecision sPrecision, QudaPrecision gPrecision)
void * memset(void *s, int c, size_t n)
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 covdev_dslash_mg4dir(cpuColorSpinorField *out, void **link, void **ghostLink, cpuColorSpinorField *in, int oddBit, int daggerBit, 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)
cudaColorSpinorField * tmp
cpuColorSpinorField * spinor
enum QudaPrecision_s QudaPrecision
enum QudaParity_s QudaParity
FloatingPoint< float > Float
__host__ __device__ T sum(const array< T, s > &a)
Main header file for the QUDA library.