1 #include <gauge_field_order.h>
5 template <typename T, QudaGaugeFieldOrder order, int Nc>
7 static constexpr int nColor = Nc;
8 typedef typename mapper<T>::type real;
9 typedef typename gauge_order_mapper<T,order,Nc>::type G;
12 ChecksumArg(const GaugeField &U, bool mini) : U(U), volumeCB(mini ? 1 : U.VolumeCB()) { }
15 template <typename Arg>
16 __device__ __host__ inline uint64_t siteChecksum(const Arg &arg, int d, int parity, int x_cb) {
17 const Matrix<complex<typename Arg::real>,Arg::nColor> u = arg.U(d, x_cb, parity);
21 template <typename Arg>
22 uint64_t ChecksumCPU(const Arg &arg)
24 uint64_t checksum_ = 0;
25 for (int parity=0; parity<2; parity++)
26 for (int x_cb=0; x_cb<arg.volumeCB; x_cb++)
27 for (int d=0; d<arg.U.geometry; d++)
28 checksum_ ^= siteChecksum(arg, d, parity, x_cb);
32 template <typename T, int Nc>
33 uint64_t Checksum(const GaugeField &u, bool mini)
35 uint64_t checksum = 0;
36 if (u.Order() == QUDA_QDP_GAUGE_ORDER) {
37 ChecksumArg<T,QUDA_QDP_GAUGE_ORDER,Nc> arg(u,mini);
38 checksum = ChecksumCPU(arg);
39 } else if (u.Order() == QUDA_QDPJIT_GAUGE_ORDER) {
40 ChecksumArg<T,QUDA_QDPJIT_GAUGE_ORDER,Nc> arg(u,mini);
41 checksum = ChecksumCPU(arg);
42 } else if (u.Order() == QUDA_MILC_GAUGE_ORDER) {
43 ChecksumArg<T,QUDA_MILC_GAUGE_ORDER,Nc> arg(u,mini);
44 checksum = ChecksumCPU(arg);
45 } else if (u.Order() == QUDA_BQCD_GAUGE_ORDER) {
46 ChecksumArg<T,QUDA_BQCD_GAUGE_ORDER,Nc> arg(u,mini);
47 checksum = ChecksumCPU(arg);
48 } else if (u.Order() == QUDA_TIFR_GAUGE_ORDER) {
49 ChecksumArg<T,QUDA_TIFR_GAUGE_ORDER,Nc> arg(u,mini);
50 checksum = ChecksumCPU(arg);
51 } else if (u.Order() == QUDA_TIFR_PADDED_GAUGE_ORDER) {
52 ChecksumArg<T,QUDA_TIFR_PADDED_GAUGE_ORDER,Nc> arg(u,mini);
53 checksum = ChecksumCPU(arg);
55 errorQuda("Checksum not implemented");
62 uint64_t Checksum(const GaugeField &u, bool mini)
64 uint64_t checksum = 0;
66 case 3: checksum = Checksum<T,3>(u,mini); break;
67 default: errorQuda("Unsupported nColor = %d", u.Ncolor());
72 uint64_t Checksum(const GaugeField &u, bool mini)
74 uint64_t checksum = 0;
75 switch (u.Precision()) {
76 case QUDA_DOUBLE_PRECISION: checksum = Checksum<double>(u,mini); break;
77 case QUDA_SINGLE_PRECISION: checksum = Checksum<float>(u,mini); break;
78 default: errorQuda("Unsupported precision = %d", u.Precision());
81 comm_allreduce_xor(&checksum);