30 location(a.Location())
36 for (
int dir = 0; dir <
nDim; ++dir)
x[dir] = a.
X()[dir];
41 clover(0),
norm(0), cloverInv(0), invNorm(0), csw(
param.csw), coeff(
param.coeff),
47 errorQuda(
"QDPJIT ordered clover fields only supported for reference fields");
109 if (!
param.inverse) {
114 evenInvNorm = evenNorm;
115 oddInvNorm = oddNorm;
144 evenNorm = evenInvNorm;
145 oddNorm = oddInvNorm;
149 if (!
param.inverse) {
154 evenInvNorm = evenNorm;
155 oddInvNorm = oddNorm;
184 static_cast<char *
>(packClover) +
bytes :
205 static_cast<char *
>(packClover) + src.
Bytes() :
243 if (
V(
false) && cpu.
V(
false)) {
248 }
else if((
V(
false) && !cpu.
V(
false)) || (!
V(
false) && cpu.
V(
false))) {
249 errorQuda(
"Mismatch between Clover field GPU V(false) and CPU.V(false)");
253 if (
V(
true) && cpu.
V(
true)) {
258 }
else if ((
V(
true) && !cpu.
V(
true)) || (!
V(
true) && cpu.
V(
true))) {
259 errorQuda(
"Mismatch between Clover field GPU V(true) and CPU.V(true)");
270 size_t buffer_offset = 0;
290 size_t buffer_offset = 0;
316 auto clover_parity =
clover;
317 auto norm_parity =
norm;
320 auto bytes_parity =
bytes;
324 norm_bytes_parity /= 2;
326 clover_parity = even;
327 norm_parity = evenNorm;
328 cloverInv_parity = evenInv;
329 invNorm_parity = evenInvNorm;
332 norm_parity = oddNorm;
333 cloverInv_parity = oddInv;
334 invNorm_parity = oddInvNorm;
342 if (clover_parity != cloverInv_parity) {
355 default:
errorQuda(
"Invalid CloverPrefetchType.");
407 size_t buffer_offset = 0;
409 std::memcpy(
static_cast<char *
>(buffer),
clover,
bytes);
415 std::memcpy(
static_cast<char *
>(buffer) + buffer_offset,
cloverInv,
bytes);
425 size_t buffer_offset = 0;
427 std::memcpy(
clover,
static_cast<char *
>(buffer),
bytes);
433 std::memcpy(
cloverInv,
static_cast<char *
>(buffer) + buffer_offset,
bytes);
443 output << static_cast<const LatticeFieldParam&>(
param);
444 output <<
"direct = " <<
param.direct << std::endl;
445 output <<
"inverse = " <<
param.inverse << std::endl;
446 output <<
"clover = " <<
param.clover << std::endl;
447 output <<
"norm = " <<
param.norm << std::endl;
448 output <<
"cloverInv = " <<
param.cloverInv << std::endl;
449 output <<
"invNorm = " <<
param.invNorm << std::endl;
450 output <<
"csw = " <<
param.csw << std::endl;
451 output <<
"coeff = " <<
param.coeff << std::endl;
452 output <<
"twisted = " <<
param.twisted << std::endl;
453 output <<
"mu2 = " <<
param.mu2 << std::endl;
454 output <<
"rho = " <<
param.rho << std::endl;
455 output <<
"order = " <<
param.order << std::endl;
456 output <<
"create = " <<
param.create << std::endl;
463 errorQuda(
"Casting a CloverField into ColorSpinorField not possible in half precision");
468 spinor_param.
nSpin = 4;
470 for (
int d=0; d<a.
Ndim(); d++) spinor_param.
x[d] = a.
X()[d];
472 spinor_param.
pad = a.
Pad();
void * Norm(bool inverse=false)
QudaCloverFieldOrder order
static CloverField * Create(const CloverFieldParam ¶m)
void setRho(double rho)
Bakes in the rho factor into the clover field, (for real diagonal additive Hasenbusch),...
void * V(bool inverse=false)
CloverField(const CloverFieldParam ¶m)
static ColorSpinorField * Create(const ColorSpinorParam ¶m)
QudaGammaBasis gammaBasis
QudaFieldLocation location
void setPrecision(QudaPrecision precision, QudaPrecision ghost_precision=QUDA_INVALID_PRECISION, bool force_native=false)
QudaFieldOrder fieldOrder
QudaPrecision Precision() const
QudaFieldLocation Location() const
void checkField(const LatticeField &a) const
virtual ~cpuCloverField()
virtual void copy_to_buffer(void *buffer) const
Copy all contents of the field to a host buffer.
cpuCloverField(const CloverFieldParam ¶m)
virtual void copy_from_buffer(void *buffer)
Copy all contents of the field from a host buffer to this field.
void copy(const CloverField &src, bool inverse=true)
Copy into this CloverField from the generic CloverField src.
void loadCPUField(const cpuCloverField &cpu)
cudaCloverField(const CloverFieldParam ¶m)
virtual void copy_from_buffer(void *buffer)
Copy all contents of the field from a host buffer to this field.
void saveCPUField(cpuCloverField &cpu) const
virtual void copy_to_buffer(void *buffer) const
Copy all contents of the field to a host buffer.
virtual ~cudaCloverField()
void prefetch(QudaFieldLocation mem_space, qudaStream_t stream=0) const
If managed memory and prefetch is enabled, prefetch the clover, the norm field (as appropriate),...
void * memset(void *s, int c, size_t n)
@ QUDA_QDPJIT_CLOVER_ORDER
@ QUDA_PACKED_CLOVER_ORDER
@ QUDA_CUDA_FIELD_LOCATION
@ QUDA_CPU_FIELD_LOCATION
enum QudaFieldLocation_s QudaFieldLocation
@ QUDA_EVEN_ODD_SITE_ORDER
@ QUDA_FLOAT2_FIELD_ORDER
@ QUDA_FLOAT4_FIELD_ORDER
@ QUDA_REFERENCE_FIELD_CREATE
enum QudaParity_s QudaParity
#define pool_pinned_malloc(size)
#define pool_device_malloc(size)
#define safe_malloc(size)
#define pool_pinned_free(ptr)
#define pool_device_free(ptr)
double norm1(const ColorSpinorField &b)
double norm2(const ColorSpinorField &a)
ColorSpinorParam colorSpinorParam(const CloverField &a, bool inverse)
double norm2(const CloverField &a, bool inverse=false)
__device__ __host__ Matrix< T, 3 > inverse(const Matrix< T, 3 > &u)
void copyGenericClover(CloverField &out, const CloverField &in, bool inverse, QudaFieldLocation location, void *Out=0, void *In=0, void *outNorm=0, void *inNorm=0)
This generic function is used for copying the clover field where in the input and output can be in an...
double norm1(const CloverField &u, bool inverse=false)
void computeClover(CloverField &clover, const GaugeField &fmunu, double coeff)
Driver for computing the clover field from the field strength tensor.
@ BOTH_CLOVER_PREFETCH_TYPE
@ INVERSE_CLOVER_PREFETCH_TYPE
@ CLOVER_CLOVER_PREFETCH_TYPE
QudaFieldLocation reorder_location()
Return whether data is reordered on the CPU or GPU. This can set at QUDA initialization using the env...
__host__ __device__ ValueType norm(const complex< ValueType > &z)
Returns the magnitude of z squared.
bool is_prefetch_enabled()
std::ostream & operator<<(std::ostream &output, const CloverFieldParam ¶m)
#define qudaMemPrefetchAsync(ptr, count, mem_space, stream)
#define qudaMemcpy(dst, src, count, kind)
cudaStream_t qudaStream_t
#define qudaDeviceSynchronize()
#define ALIGNMENT_ADJUST(n)
QudaFieldLocation location
QudaSiteSubset siteSubset