quda::blas::HeavyQuarkResidualNorm_< ReduceType, Float2, FloatN > Struct Template Reference

Inheritance diagram for quda::blas::HeavyQuarkResidualNorm_< ReduceType, Float2, FloatN >:

Collaboration diagram for quda::blas::HeavyQuarkResidualNorm_< ReduceType, Float2, FloatN >:

Public Types
typedef scalar< ReduceType >::type	real

Public Member Functions
	HeavyQuarkResidualNorm_ (const Float2 &a, const Float2 &b)
__device__ __host__ void	pre ()
	pre-computation routine called before the "M-loop" More...
__device__ __host__ void	operator() (ReduceType &sum, FloatN &x, FloatN &y, FloatN &z, FloatN &w, FloatN &v)
	where the reduction is usually computed and any auxiliary operations More...
__device__ __host__ void	post (ReduceType &sum)
	sum the solution and residual norms, and compute the heavy-quark norm More...

Static Public Member Functions
static int	streams ()
static int	flops ()
	total number of input and output streams More...

Public Attributes
Float2	a
Float2	b
ReduceType	aux

Detailed Description

template<typename ReduceType, typename Float2, typename FloatN>
struct quda::blas::HeavyQuarkResidualNorm_< ReduceType, Float2, FloatN >

This kernel returns (x, x) and (r,r) and also returns the so-called heavy quark norm as used by MILC: 1 / N * (r, r)_i / (x, x)_i, where i is site index and N is the number of sites. When this kernel is launched, we must enforce that the parameter M in the launcher corresponds to the number of FloatN fields used to represent the spinor, e.g., M=6 for Wilson and M=3 for staggered. This is only the case for half-precision kernels by default. To enable this, the siteUnroll template parameter must be set true when reduceCuda is instantiated.

Definition at line 471 of file reduce_core.cuh.

Member Typedef Documentation

real

template<typename ReduceType , typename Float2 , typename FloatN >

typedef scalar<ReduceType>::type quda::blas::HeavyQuarkResidualNorm_< ReduceType, Float2, FloatN >::real

Definition at line 472 of file reduce_core.cuh.

Constructor & Destructor Documentation

HeavyQuarkResidualNorm_()

template<typename ReduceType , typename Float2 , typename FloatN >

quda::blas::HeavyQuarkResidualNorm_< ReduceType, Float2, FloatN >::HeavyQuarkResidualNorm_ ( const Float2 &  a, const Float2 &  b  )

inline

Definition at line 476 of file reduce_core.cuh.

Member Function Documentation

flops()

template<typename ReduceType , typename Float2 , typename FloatN >

static int quda::blas::HeavyQuarkResidualNorm_< ReduceType, Float2, FloatN >::flops ( )

inlinestatic

total number of input and output streams

Definition at line 499 of file reduce_core.cuh.

operator()()

template<typename ReduceType , typename Float2 , typename FloatN >

__device__ __host__ void quda::blas::HeavyQuarkResidualNorm_< ReduceType, Float2, FloatN >::operator() ( ReduceType &  sum, FloatN &  x, FloatN &  y, FloatN &  z, FloatN &  w, FloatN &  v  )

inlinevirtual

where the reduction is usually computed and any auxiliary operations

Implements quda::blas::ReduceFunctor< ReduceType, Float2, FloatN >.

Definition at line 484 of file reduce_core.cuh.

post()

template<typename ReduceType , typename Float2 , typename FloatN >

__device__ __host__ void quda::blas::HeavyQuarkResidualNorm_< ReduceType, Float2, FloatN >::post ( ReduceType &  sum )

inlinevirtual

sum the solution and residual norms, and compute the heavy-quark norm

Reimplemented from quda::blas::ReduceFunctor< ReduceType, Float2, FloatN >.

Definition at line 491 of file reduce_core.cuh.

pre()

template<typename ReduceType , typename Float2 , typename FloatN >

__device__ __host__ void quda::blas::HeavyQuarkResidualNorm_< ReduceType, Float2, FloatN >::pre ( )

inlinevirtual

pre-computation routine called before the "M-loop"

Reimplemented from quda::blas::ReduceFunctor< ReduceType, Float2, FloatN >.

Definition at line 478 of file reduce_core.cuh.

streams()

template<typename ReduceType , typename Float2 , typename FloatN >

static int quda::blas::HeavyQuarkResidualNorm_< ReduceType, Float2, FloatN >::streams ( )

inlinestatic

Definition at line 498 of file reduce_core.cuh.

Member Data Documentation

a

template<typename ReduceType , typename Float2 , typename FloatN >

Float2 quda::blas::HeavyQuarkResidualNorm_< ReduceType, Float2, FloatN >::a

Definition at line 473 of file reduce_core.cuh.

aux

template<typename ReduceType , typename Float2 , typename FloatN >

ReduceType quda::blas::HeavyQuarkResidualNorm_< ReduceType, Float2, FloatN >::aux

Definition at line 475 of file reduce_core.cuh.

b

template<typename ReduceType , typename Float2 , typename FloatN >

Float2 quda::blas::HeavyQuarkResidualNorm_< ReduceType, Float2, FloatN >::b

Definition at line 474 of file reduce_core.cuh.

---

The documentation for this struct was generated from the following file:

• include/kernels/reduce_core.cuh