diff --git a/src/common/crypto/Algorithm.cpp b/src/common/crypto/Algorithm.cpp
index 31035fb10..05c890dc7 100644
--- a/src/common/crypto/Algorithm.cpp
+++ b/src/common/crypto/Algorithm.cpp
@@ -60,6 +60,7 @@ static AlgoData const algorithms[] = {
{ "cryptonight/msr", "cn/msr", xmrig::CRYPTONIGHT, xmrig::VARIANT_MSR },
{ "cryptonight/xao", "cn/xao", xmrig::CRYPTONIGHT, xmrig::VARIANT_XAO },
{ "cryptonight/rto", "cn/rto", xmrig::CRYPTONIGHT, xmrig::VARIANT_RTO },
+ { "cryptonight/2", "cn/2", xmrig::CRYPTONIGHT, xmrig::VARIANT_2 },
# ifndef XMRIG_NO_AEON
{ "cryptonight-lite", "cn-lite", xmrig::CRYPTONIGHT_LITE, xmrig::VARIANT_AUTO },
@@ -81,6 +82,8 @@ static AlgoData const algorithms[] = {
static AlgoData const xmrStakAlgorithms[] = {
{ "cryptonight-monerov7", nullptr, xmrig::CRYPTONIGHT, xmrig::VARIANT_1 },
{ "cryptonight_v7", nullptr, xmrig::CRYPTONIGHT, xmrig::VARIANT_1 },
+ { "cryptonight-monerov8", nullptr, xmrig::CRYPTONIGHT, xmrig::VARIANT_2 },
+ { "cryptonight_v8", nullptr, xmrig::CRYPTONIGHT, xmrig::VARIANT_2 },
{ "cryptonight_v7_stellite", nullptr, xmrig::CRYPTONIGHT, xmrig::VARIANT_XTL },
{ "cryptonight_lite", nullptr, xmrig::CRYPTONIGHT_LITE, xmrig::VARIANT_0 },
{ "cryptonight-aeonv7", nullptr, xmrig::CRYPTONIGHT_LITE, xmrig::VARIANT_1 },
@@ -103,7 +106,8 @@ static const char *variants[] = {
"msr",
"xhv",
"xao",
- "rto"
+ "rto",
+ "2",
};
diff --git a/src/common/net/Job.cpp b/src/common/net/Job.cpp
index 80b521ea1..e9f81e02e 100644
--- a/src/common/net/Job.cpp
+++ b/src/common/net/Job.cpp
@@ -178,7 +178,12 @@ xmrig::Variant Job::variant() const
}
if (m_algorithm.variant() == xmrig::VARIANT_AUTO) {
- return m_algorithm.algo() == xmrig::CRYPTONIGHT_HEAVY ? xmrig::VARIANT_0 : xmrig::VARIANT_1;
+ if (m_algorithm.algo() == xmrig::CRYPTONIGHT_HEAVY) {
+ return xmrig::VARIANT_0;
+ } else if (m_algorithm.algo() == xmrig::CRYPTONIGHT_LITE) {
+ return xmrig::VARIANT_1;
+ }
+ return (m_blob[0] >= 8) ? xmrig::VARIANT_2 : xmrig::VARIANT_1;
}
return m_algorithm.variant();
diff --git a/src/common/net/Pool.cpp b/src/common/net/Pool.cpp
index 053f25070..357cb330b 100644
--- a/src/common/net/Pool.cpp
+++ b/src/common/net/Pool.cpp
@@ -211,6 +211,7 @@ rapidjson::Value Pool::toJSON(rapidjson::Document &doc) const
case xmrig::VARIANT_AUTO:
case xmrig::VARIANT_0:
case xmrig::VARIANT_1:
+ case xmrig::VARIANT_2:
obj.AddMember("variant", m_algorithm.variant(), allocator);
break;
@@ -377,6 +378,7 @@ void Pool::rebuild()
m_algorithms.push_back(m_algorithm);
# ifndef XMRIG_PROXY_PROJECT
+ addVariant(xmrig::VARIANT_2);
addVariant(xmrig::VARIANT_1);
addVariant(xmrig::VARIANT_0);
addVariant(xmrig::VARIANT_XTL);
diff --git a/src/common/xmrig.h b/src/common/xmrig.h
index 58a3540c7..3e1b65dfd 100644
--- a/src/common/xmrig.h
+++ b/src/common/xmrig.h
@@ -67,6 +67,7 @@ enum Variant {
VARIANT_XHV = 5, // Modified CryptoNight-Heavy (Haven Protocol only)
VARIANT_XAO = 6, // Modified CryptoNight variant 0 (Alloy only)
VARIANT_RTO = 7, // Modified CryptoNight variant 1 (Arto only)
+ VARIANT_2 = 8, // CryptoNight variant 2
VARIANT_MAX
};
diff --git a/src/crypto/CryptoNight_arm.h b/src/crypto/CryptoNight_arm.h
index efb5759e1..7f273765d 100644
--- a/src/crypto/CryptoNight_arm.h
+++ b/src/crypto/CryptoNight_arm.h
@@ -404,19 +404,27 @@ static inline __m128i aes_round_tweak_div(const __m128i &in, const __m128i &key)
}
-template<int SHIFT>
-static inline void cryptonight_monero_tweak(uint64_t* mem_out, __m128i tmp)
+template<int VARIANT>
+static inline void cryptonight_monero_tweak(const uint8_t* l, uint64_t idx, __m128i ax0, __m128i bx0, __m128i bx1, __m128i cx)
{
- mem_out[0] = EXTRACT64(tmp);
+ uint64_t* mem_out = (uint64_t*)&l[idx];
- uint64_t vh = vgetq_lane_u64(tmp, 1);
+ if (VARIANT == xmrig::VARIANT_2) {
+ VARIANT2_SHUFFLE(l, idx, ax0, bx0, bx1);
+ _mm_store_si128((__m128i *)mem_out, _mm_xor_si128(bx0, cx));
+ } else {
+ __m128i tmp = _mm_xor_si128(bx0, cx);
+ mem_out[0] = EXTRACT64(tmp);
- uint8_t x = vh >> 24;
- static const uint16_t table = 0x7531;
- const uint8_t index = (((x >> SHIFT) & 6) | (x & 1)) << 1;
- vh ^= ((table >> index) & 0x3) << 28;
+ uint64_t vh = vgetq_lane_u64(tmp, 1);
- mem_out[1] = vh;
+ uint8_t x = vh >> 24;
+ static const uint16_t table = 0x7531;
+ const uint8_t index = (((x >> (VARIANT == xmrig::VARIANT_XTL ? 4 : 3)) & 6) | (x & 1)) << 1;
+ vh ^= ((table >> index) & 0x3) << 28;
+
+ mem_out[1] = vh;
+ }
}
@@ -426,27 +434,29 @@ inline void cryptonight_single_hash(const uint8_t *__restrict__ input, size_t si
constexpr size_t MASK = xmrig::cn_select_mask<ALGO>();
constexpr size_t ITERATIONS = xmrig::cn_select_iter<ALGO, VARIANT>();
constexpr size_t MEM = xmrig::cn_select_memory<ALGO>();
- constexpr bool IS_MONERO = xmrig::cn_is_monero<VARIANT>();
+ constexpr bool IS_V1 = xmrig::cn_uses_variant1<VARIANT>();
- if (IS_MONERO && size < 43) {
+ if (IS_V1 && size < 43) {
memset(output, 0, 32);
return;
}
xmrig::keccak(input, size, ctx[0]->state);
- VARIANT1_INIT(0);
-
cn_explode_scratchpad<ALGO, MEM, SOFT_AES>((__m128i*) ctx[0]->state, (__m128i*) ctx[0]->memory);
const uint8_t* l0 = ctx[0]->memory;
uint64_t* h0 = reinterpret_cast<uint64_t*>(ctx[0]->state);
+ VARIANT1_INIT(0);
+ VARIANT2_INIT(0);
+
uint64_t al0 = h0[0] ^ h0[4];
uint64_t ah0 = h0[1] ^ h0[5];
__m128i bx0 = _mm_set_epi64x(h0[3] ^ h0[7], h0[2] ^ h0[6]);
+ __m128i bx1 = _mm_set_epi64x(h0[9] ^ h0[11], h0[8] ^ h0[10]);
- uint64_t idx0 = h0[0] ^ h0[4];
+ uint64_t idx0 = al0;
for (size_t i = 0; i < ITERATIONS; i++) {
__m128i cx;
@@ -454,44 +464,47 @@ inline void cryptonight_single_hash(const uint8_t *__restrict__ input, size_t si
cx = _mm_load_si128((__m128i *) &l0[idx0 & MASK]);
}
+ const __m128i ax0 = _mm_set_epi64x(ah0, al0);
if (VARIANT == xmrig::VARIANT_TUBE) {
- cx = aes_round_tweak_div(cx, _mm_set_epi64x(ah0, al0));
+ cx = aes_round_tweak_div(cx, ax0);
}
else if (SOFT_AES) {
- cx = soft_aesenc((uint32_t*)&l0[idx0 & MASK], _mm_set_epi64x(ah0, al0));
+ cx = soft_aesenc((uint32_t*)&l0[idx0 & MASK], ax0);
}
else {
- cx = _mm_aesenc_si128(cx, _mm_set_epi64x(ah0, al0));
+ cx = _mm_aesenc_si128(cx, ax0);
}
- if (IS_MONERO) {
- cryptonight_monero_tweak<VARIANT == xmrig::VARIANT_XTL ? 4 : 3>((uint64_t*)&l0[idx0 & MASK], _mm_xor_si128(bx0, cx));
+ if (IS_V1 || VARIANT == xmrig::VARIANT_2) {
+ cryptonight_monero_tweak<VARIANT>(l0, idx0 & MASK, ax0, bx0, bx1, cx);
} else {
_mm_store_si128((__m128i *)&l0[idx0 & MASK], _mm_xor_si128(bx0, cx));
}
idx0 = EXTRACT64(cx);
- bx0 = cx;
uint64_t hi, lo, cl, ch;
cl = ((uint64_t*) &l0[idx0 & MASK])[0];
ch = ((uint64_t*) &l0[idx0 & MASK])[1];
- lo = __umul128(idx0, cl, &hi);
+ if (VARIANT == xmrig::VARIANT_2) {
+ VARIANT2_INTEGER_MATH(0, cl, cx);
+ lo = __umul128(idx0, cl, &hi);
+ VARIANT2_SHUFFLE(l0, idx0 & MASK, ax0, bx0, bx1);
+ }
+ else {
+ lo = __umul128(idx0, cl, &hi);
+ }
al0 += hi;
ah0 += lo;
((uint64_t*)&l0[idx0 & MASK])[0] = al0;
- if (IS_MONERO) {
- if (VARIANT == xmrig::VARIANT_TUBE || VARIANT == xmrig::VARIANT_RTO) {
- ((uint64_t*)&l0[idx0 & MASK])[1] = ah0 ^ tweak1_2_0 ^ al0;
- }
- else {
- ((uint64_t*)&l0[idx0 & MASK])[1] = ah0 ^ tweak1_2_0;
- }
- }
- else {
+ if (IS_V1 && (VARIANT == xmrig::VARIANT_TUBE || VARIANT == xmrig::VARIANT_RTO)) {
+ ((uint64_t*)&l0[idx0 & MASK])[1] = ah0 ^ tweak1_2_0 ^ al0;
+ } else if (IS_V1) {
+ ((uint64_t*)&l0[idx0 & MASK])[1] = ah0 ^ tweak1_2_0;
+ } else {
((uint64_t*)&l0[idx0 & MASK])[1] = ah0;
}
@@ -514,6 +527,8 @@ inline void cryptonight_single_hash(const uint8_t *__restrict__ input, size_t si
idx0 = d ^ q;
}
}
+ bx1 = bx0;
+ bx0 = cx;
}
cn_implode_scratchpad<ALGO, MEM, SOFT_AES>((__m128i*) ctx[0]->memory, (__m128i*) ctx[0]->state);
@@ -529,9 +544,9 @@ inline void cryptonight_double_hash(const uint8_t *__restrict__ input, size_t si
constexpr size_t MASK = xmrig::cn_select_mask<ALGO>();
constexpr size_t ITERATIONS = xmrig::cn_select_iter<ALGO, VARIANT>();
constexpr size_t MEM = xmrig::cn_select_memory<ALGO>();
- constexpr bool IS_MONERO = xmrig::cn_is_monero<VARIANT>();
+ constexpr bool IS_V1 = xmrig::cn_uses_variant1<VARIANT>();
- if (IS_MONERO && size < 43) {
+ if (IS_V1 && size < 43) {
memset(output, 0, 64);
return;
}
@@ -539,14 +554,16 @@ inline void cryptonight_double_hash(const uint8_t *__restrict__ input, size_t si
xmrig::keccak(input, size, ctx[0]->state);
xmrig::keccak(input + size, size, ctx[1]->state);
- VARIANT1_INIT(0);
- VARIANT1_INIT(1);
-
const uint8_t* l0 = ctx[0]->memory;
const uint8_t* l1 = ctx[1]->memory;
uint64_t* h0 = reinterpret_cast<uint64_t*>(ctx[0]->state);
uint64_t* h1 = reinterpret_cast<uint64_t*>(ctx[1]->state);
+ VARIANT1_INIT(0);
+ VARIANT1_INIT(1);
+ VARIANT2_INIT(0);
+ VARIANT2_INIT(1);
+
cn_explode_scratchpad<ALGO, MEM, SOFT_AES>((__m128i*) h0, (__m128i*) l0);
cn_explode_scratchpad<ALGO, MEM, SOFT_AES>((__m128i*) h1, (__m128i*) l1);
@@ -555,11 +572,13 @@ inline void cryptonight_double_hash(const uint8_t *__restrict__ input, size_t si
uint64_t ah0 = h0[1] ^ h0[5];
uint64_t ah1 = h1[1] ^ h1[5];
- __m128i bx0 = _mm_set_epi64x(h0[3] ^ h0[7], h0[2] ^ h0[6]);
- __m128i bx1 = _mm_set_epi64x(h1[3] ^ h1[7], h1[2] ^ h1[6]);
+ __m128i bx00 = _mm_set_epi64x(h0[3] ^ h0[7], h0[2] ^ h0[6]);
+ __m128i bx01 = _mm_set_epi64x(h0[9] ^ h0[11], h0[8] ^ h0[10]);
+ __m128i bx10 = _mm_set_epi64x(h1[3] ^ h1[7], h1[2] ^ h1[6]);
+ __m128i bx11 = _mm_set_epi64x(h1[9] ^ h1[11], h1[8] ^ h1[10]);
- uint64_t idx0 = h0[0] ^ h0[4];
- uint64_t idx1 = h1[0] ^ h1[4];
+ uint64_t idx0 = al0;
+ uint64_t idx1 = al1;
for (size_t i = 0; i < ITERATIONS; i++) {
__m128i cx0, cx1;
@@ -568,52 +587,53 @@ inline void cryptonight_double_hash(const uint8_t *__restrict__ input, size_t si
cx1 = _mm_load_si128((__m128i *) &l1[idx1 & MASK]);
}
+ const __m128i ax0 = _mm_set_epi64x(ah0, al0);
+ const __m128i ax1 = _mm_set_epi64x(ah1, al1);
if (VARIANT == xmrig::VARIANT_TUBE) {
- cx0 = aes_round_tweak_div(cx0, _mm_set_epi64x(ah0, al0));
- cx1 = aes_round_tweak_div(cx1, _mm_set_epi64x(ah1, al1));
+ cx0 = aes_round_tweak_div(cx0, ax0);
+ cx1 = aes_round_tweak_div(cx1, ax1);
}
else if (SOFT_AES) {
- cx0 = soft_aesenc((uint32_t*)&l0[idx0 & MASK], _mm_set_epi64x(ah0, al0));
- cx1 = soft_aesenc((uint32_t*)&l1[idx1 & MASK], _mm_set_epi64x(ah1, al1));
+ cx0 = soft_aesenc((uint32_t*)&l0[idx0 & MASK], ax0);
+ cx1 = soft_aesenc((uint32_t*)&l1[idx1 & MASK], ax1);
}
else {
- cx0 = _mm_aesenc_si128(cx0, _mm_set_epi64x(ah0, al0));
- cx1 = _mm_aesenc_si128(cx1, _mm_set_epi64x(ah1, al1));
+ cx0 = _mm_aesenc_si128(cx0, ax0);
+ cx1 = _mm_aesenc_si128(cx1, ax1);
}
- if (IS_MONERO) {
- cryptonight_monero_tweak<VARIANT == xmrig::VARIANT_XTL ? 4 : 3>((uint64_t*)&l0[idx0 & MASK], _mm_xor_si128(bx0, cx0));
- cryptonight_monero_tweak<VARIANT == xmrig::VARIANT_XTL ? 4 : 3>((uint64_t*)&l1[idx1 & MASK], _mm_xor_si128(bx1, cx1));
+ if (IS_V1 || (VARIANT == xmrig::VARIANT_2)) {
+ cryptonight_monero_tweak<VARIANT>(l0, idx0 & MASK, ax0, bx00, bx01, cx0);
+ cryptonight_monero_tweak<VARIANT>(l1, idx1 & MASK, ax1, bx10, bx11, cx1);
} else {
- _mm_store_si128((__m128i *) &l0[idx0 & MASK], _mm_xor_si128(bx0, cx0));
- _mm_store_si128((__m128i *) &l1[idx1 & MASK], _mm_xor_si128(bx1, cx1));
- };
+ _mm_store_si128((__m128i *) &l0[idx0 & MASK], _mm_xor_si128(bx00, cx0));
+ _mm_store_si128((__m128i *) &l1[idx1 & MASK], _mm_xor_si128(bx10, cx1));
+ }
idx0 = EXTRACT64(cx0);
idx1 = EXTRACT64(cx1);
- bx0 = cx0;
- bx1 = cx1;
-
uint64_t hi, lo, cl, ch;
cl = ((uint64_t*) &l0[idx0 & MASK])[0];
ch = ((uint64_t*) &l0[idx0 & MASK])[1];
- lo = __umul128(idx0, cl, &hi);
+ if (VARIANT == xmrig::VARIANT_2) {
+ VARIANT2_INTEGER_MATH(0, cl, cx0);
+ lo = __umul128(idx0, cl, &hi);
+ VARIANT2_SHUFFLE(l0, idx0 & MASK, ax0, bx00, bx01);
+ } else {
+ lo = __umul128(idx0, cl, &hi);
+ }
al0 += hi;
ah0 += lo;
((uint64_t*)&l0[idx0 & MASK])[0] = al0;
- if (IS_MONERO) {
- if (VARIANT == xmrig::VARIANT_TUBE || VARIANT == xmrig::VARIANT_RTO) {
- ((uint64_t*)&l0[idx0 & MASK])[1] = ah0 ^ tweak1_2_0 ^ al0;
- }
- else {
- ((uint64_t*)&l0[idx0 & MASK])[1] = ah0 ^ tweak1_2_0;
- }
- }
- else {
+ if (IS_V1 && (VARIANT == xmrig::VARIANT_TUBE || VARIANT == xmrig::VARIANT_RTO)) {
+ ((uint64_t*)&l0[idx0 & MASK])[1] = ah0 ^ tweak1_2_0 ^ al0;
+ } else if (IS_V1) {
+ ((uint64_t*)&l0[idx0 & MASK])[1] = ah0 ^ tweak1_2_0;
+ } else {
((uint64_t*)&l0[idx0 & MASK])[1] = ah0;
}
@@ -639,22 +659,24 @@ inline void cryptonight_double_hash(const uint8_t *__restrict__ input, size_t si
cl = ((uint64_t*) &l1[idx1 & MASK])[0];
ch = ((uint64_t*) &l1[idx1 & MASK])[1];
- lo = __umul128(idx1, cl, &hi);
+ if (VARIANT == xmrig::VARIANT_2) {
+ VARIANT2_INTEGER_MATH(1, cl, cx1);
+ lo = __umul128(idx1, cl, &hi);
+ VARIANT2_SHUFFLE(l1, idx1 & MASK, ax1, bx10, bx11);
+ } else {
+ lo = __umul128(idx1, cl, &hi);
+ }
al1 += hi;
ah1 += lo;
((uint64_t*)&l1[idx1 & MASK])[0] = al1;
- if (IS_MONERO) {
- if (VARIANT == xmrig::VARIANT_TUBE || VARIANT == xmrig::VARIANT_RTO) {
- ((uint64_t*)&l1[idx1 & MASK])[1] = ah1 ^ tweak1_2_1 ^ al1;
- }
- else {
- ((uint64_t*)&l1[idx1 & MASK])[1] = ah1 ^ tweak1_2_1;
- }
- }
- else {
+ if (IS_V1 && (VARIANT == xmrig::VARIANT_TUBE || VARIANT == xmrig::VARIANT_RTO)) {
+ ((uint64_t*)&l1[idx1 & MASK])[1] = ah1 ^ tweak1_2_1 ^ al1;
+ } else if (IS_V1) {
+ ((uint64_t*)&l1[idx1 & MASK])[1] = ah1 ^ tweak1_2_1;
+ } else {
((uint64_t*)&l1[idx1 & MASK])[1] = ah1;
}
@@ -677,6 +699,10 @@ inline void cryptonight_double_hash(const uint8_t *__restrict__ input, size_t si
idx1 = d ^ q;
}
}
+ bx01 = bx00;
+ bx00 = cx0;
+ bx11 = bx10;
+ bx10 = cx1;
}
cn_implode_scratchpad<ALGO, MEM, SOFT_AES>((__m128i*) l0, (__m128i*) h0);
diff --git a/src/crypto/CryptoNight_constants.h b/src/crypto/CryptoNight_constants.h
index 08a755d40..97a77bbdb 100644
--- a/src/crypto/CryptoNight_constants.h
+++ b/src/crypto/CryptoNight_constants.h
@@ -107,6 +107,7 @@ inline uint32_t cn_select_mask(Algo algorithm)
template<Algo ALGO, Variant variant> inline constexpr uint32_t cn_select_iter() { return 0; }
template<> inline constexpr uint32_t cn_select_iter<CRYPTONIGHT, VARIANT_0>() { return CRYPTONIGHT_ITER; }
template<> inline constexpr uint32_t cn_select_iter<CRYPTONIGHT, VARIANT_1>() { return CRYPTONIGHT_ITER; }
+template<> inline constexpr uint32_t cn_select_iter<CRYPTONIGHT, VARIANT_2>() { return CRYPTONIGHT_ITER; }
template<> inline constexpr uint32_t cn_select_iter<CRYPTONIGHT, VARIANT_XTL>() { return CRYPTONIGHT_ITER; }
template<> inline constexpr uint32_t cn_select_iter<CRYPTONIGHT, VARIANT_MSR>() { return CRYPTONIGHT_MSR_ITER; }
template<> inline constexpr uint32_t cn_select_iter<CRYPTONIGHT, VARIANT_XAO>() { return CRYPTONIGHT_XAO_ITER; }
@@ -150,29 +151,16 @@ inline uint32_t cn_select_iter(Algo algorithm, Variant variant)
}
-template<Variant variant> inline constexpr bool cn_is_monero() { return false; }
-template<> inline constexpr bool cn_is_monero<VARIANT_0>() { return false; }
-template<> inline constexpr bool cn_is_monero<VARIANT_1>() { return true; }
-template<> inline constexpr bool cn_is_monero<VARIANT_TUBE>() { return true; }
-template<> inline constexpr bool cn_is_monero<VARIANT_XTL>() { return true; }
-template<> inline constexpr bool cn_is_monero<VARIANT_MSR>() { return true; }
-template<> inline constexpr bool cn_is_monero<VARIANT_XHV>() { return false; }
-template<> inline constexpr bool cn_is_monero<VARIANT_XAO>() { return false; }
-template<> inline constexpr bool cn_is_monero<VARIANT_RTO>() { return true; }
-
-
-inline bool cn_is_monero(Variant variant)
-{
- switch (variant) {
- case VARIANT_0:
- case VARIANT_XHV:
- case VARIANT_RTO:
- return false;
-
- default:
- return true;
- }
-}
+template<Variant variant> inline constexpr bool cn_uses_variant1() { return false; }
+template<> inline constexpr bool cn_uses_variant1<VARIANT_0>() { return false; }
+template<> inline constexpr bool cn_uses_variant1<VARIANT_1>() { return true; }
+template<> inline constexpr bool cn_uses_variant1<VARIANT_TUBE>() { return true; }
+template<> inline constexpr bool cn_uses_variant1<VARIANT_XTL>() { return true; }
+template<> inline constexpr bool cn_uses_variant1<VARIANT_MSR>() { return true; }
+template<> inline constexpr bool cn_uses_variant1<VARIANT_XHV>() { return false; }
+template<> inline constexpr bool cn_uses_variant1<VARIANT_XAO>() { return false; }
+template<> inline constexpr bool cn_uses_variant1<VARIANT_RTO>() { return true; }
+template<> inline constexpr bool cn_uses_variant1<VARIANT_2>() { return false; }
} /* namespace xmrig */
diff --git a/src/crypto/CryptoNight_monero.h b/src/crypto/CryptoNight_monero.h
index a758fdbc5..3fb18d00f 100644
--- a/src/crypto/CryptoNight_monero.h
+++ b/src/crypto/CryptoNight_monero.h
@@ -25,26 +25,27 @@
#ifndef __CRYPTONIGHT_MONERO_H__
#define __CRYPTONIGHT_MONERO_H__
+#include <math.h>
// VARIANT ALTERATIONS
#ifndef XMRIG_ARM
# define VARIANT1_INIT(part) \
uint64_t tweak1_2_##part = 0; \
- if (IS_MONERO) { \
+ if (IS_V1) { \
tweak1_2_##part = (*reinterpret_cast<const uint64_t*>(input + 35 + part * size) ^ \
*(reinterpret_cast<const uint64_t*>(ctx[part]->state) + 24)); \
}
#else
# define VARIANT1_INIT(part) \
uint64_t tweak1_2_##part = 0; \
- if (IS_MONERO) { \
+ if (IS_V1) { \
memcpy(&tweak1_2_##part, input + 35 + part * size, sizeof tweak1_2_##part); \
tweak1_2_##part ^= *(reinterpret_cast<const uint64_t*>(ctx[part]->state) + 24); \
}
#endif
#define VARIANT1_1(p) \
- if (IS_MONERO) { \
+ if (IS_V1) { \
const uint8_t tmp = reinterpret_cast<const uint8_t*>(p)[11]; \
static const uint32_t table = 0x75310; \
const uint8_t index = (((tmp >> 3) & 6) | (tmp & 1)) << 1; \
@@ -52,9 +53,72 @@
}
#define VARIANT1_2(p, part) \
- if (IS_MONERO) { \
+ if (IS_V1) { \
(p) ^= tweak1_2_##part; \
}
+#ifndef XMRIG_ARM
+# define VARIANT2_INIT(part) \
+ __m128i division_result_xmm_##part = _mm_cvtsi64_si128(h##part[12]); \
+ __m128i sqrt_result_xmm_##part = _mm_cvtsi64_si128(h##part[13]);
+
+#ifdef _MSC_VER
+# define VARIANT2_SET_ROUNDING_MODE() if (VARIANT == xmrig::VARIANT_2) { _control87(RC_DOWN, MCW_RC); }
+#else
+# define VARIANT2_SET_ROUNDING_MODE() if (VARIANT == xmrig::VARIANT_2) { std::fesetround(FE_DOWNWARD); }
+#endif
+
+# define VARIANT2_INTEGER_MATH(part, cl, cx) \
+ do { \
+ const uint64_t sqrt_result = static_cast<uint64_t>(_mm_cvtsi128_si64(sqrt_result_xmm_##part)); \
+ const uint64_t cx_0 = _mm_cvtsi128_si64(cx); \
+ cl ^= static_cast<uint64_t>(_mm_cvtsi128_si64(division_result_xmm_##part)) ^ (sqrt_result << 32); \
+ const uint32_t d = static_cast<uint32_t>(cx_0 + (sqrt_result << 1)) | 0x80000001UL; \
+ const uint64_t cx_1 = _mm_cvtsi128_si64(_mm_srli_si128(cx, 8)); \
+ const uint64_t division_result = static_cast<uint32_t>(cx_1 / d) + ((cx_1 % d) << 32); \
+ division_result_xmm_##part = _mm_cvtsi64_si128(static_cast<int64_t>(division_result)); \
+ sqrt_result_xmm_##part = int_sqrt_v2(cx_0 + division_result); \
+ } while (0)
+
+# define VARIANT2_SHUFFLE(base_ptr, offset, _a, _b, _b1) \
+ do { \
+ const __m128i chunk1 = _mm_load_si128((__m128i *)((base_ptr) + ((offset) ^ 0x10))); \
+ const __m128i chunk2 = _mm_load_si128((__m128i *)((base_ptr) + ((offset) ^ 0x20))); \
+ const __m128i chunk3 = _mm_load_si128((__m128i *)((base_ptr) + ((offset) ^ 0x30))); \
+ _mm_store_si128((__m128i *)((base_ptr) + ((offset) ^ 0x10)), _mm_add_epi64(chunk3, _b1)); \
+ _mm_store_si128((__m128i *)((base_ptr) + ((offset) ^ 0x20)), _mm_add_epi64(chunk1, _b)); \
+ _mm_store_si128((__m128i *)((base_ptr) + ((offset) ^ 0x30)), _mm_add_epi64(chunk2, _a)); \
+ } while (0)
+
+#else
+# define VARIANT2_INIT(part) \
+ uint64_t division_result_##part = h##part[12]; \
+ uint64_t sqrt_result_##part = h##part[13];
+
+# define VARIANT2_INTEGER_MATH(part, cl, cx) \
+ do { \
+ const uint64_t cx_0 = _mm_cvtsi128_si64(cx); \
+ cl ^= division_result_##part ^ (sqrt_result_##part << 32); \
+ const uint32_t d = static_cast<uint32_t>(cx_0 + (sqrt_result_##part << 1)) | 0x80000001UL; \
+ const uint64_t cx_1 = _mm_cvtsi128_si64(_mm_srli_si128(cx, 8)); \
+ division_result_##part = static_cast<uint32_t>(cx_1 / d) + ((cx_1 % d) << 32); \
+ const uint64_t sqrt_input = cx_0 + division_result_##part; \
+ sqrt_result_##part = sqrt(sqrt_input + 18446744073709551616.0) * 2.0 - 8589934592.0; \
+ const uint64_t s = sqrt_result_##part >> 1; \
+ const uint64_t b = sqrt_result_##part & 1; \
+ const uint64_t r2 = (uint64_t)(s) * (s + b) + (sqrt_result_##part << 32); \
+ sqrt_result_##part += ((r2 + b > sqrt_input) ? -1 : 0) + ((r2 + (1ULL << 32) < sqrt_input - s) ? 1 : 0); \
+ } while (0)
+
+# define VARIANT2_SHUFFLE(base_ptr, offset, _a, _b, _b1) \
+ do { \
+ const uint64x2_t chunk1 = vld1q_u64((uint64_t*)((base_ptr) + ((offset) ^ 0x10))); \
+ const uint64x2_t chunk2 = vld1q_u64((uint64_t*)((base_ptr) + ((offset) ^ 0x20))); \
+ const uint64x2_t chunk3 = vld1q_u64((uint64_t*)((base_ptr) + ((offset) ^ 0x30))); \
+ vst1q_u64((uint64_t*)((base_ptr) + ((offset) ^ 0x10)), vaddq_u64(chunk3, vreinterpretq_u64_u8(_b1))); \
+ vst1q_u64((uint64_t*)((base_ptr) + ((offset) ^ 0x20)), vaddq_u64(chunk1, vreinterpretq_u64_u8(_b))); \
+ vst1q_u64((uint64_t*)((base_ptr) + ((offset) ^ 0x30)), vaddq_u64(chunk2, vreinterpretq_u64_u8(_a))); \
+ } while (0)
+#endif
#endif /* __CRYPTONIGHT_MONERO_H__ */
diff --git a/src/crypto/CryptoNight_test.h b/src/crypto/CryptoNight_test.h
index 16296efbf..953f88d0e 100644
--- a/src/crypto/CryptoNight_test.h
+++ b/src/crypto/CryptoNight_test.h
@@ -69,7 +69,7 @@ const static uint8_t test_output_v0[160] = {
};
-// Monero v7
+// Cryptonight variant 1 (Monero v7)
const static uint8_t test_output_v1[160] = {
0xF2, 0x2D, 0x3D, 0x62, 0x03, 0xD2, 0xA0, 0x8B, 0x41, 0xD9, 0x02, 0x72, 0x78, 0xD8, 0xBC, 0xC9,
0x83, 0xAC, 0xAD, 0xA9, 0xB6, 0x8E, 0x52, 0xE3, 0xC6, 0x89, 0x69, 0x2A, 0x50, 0xE9, 0x21, 0xD9,
@@ -84,6 +84,21 @@ const static uint8_t test_output_v1[160] = {
};
+// Cryptonight variant 2 (Monero v8)
+const static uint8_t test_output_v2[160] = {
+ 0x6E, 0xEE, 0x53, 0xA3, 0xDA, 0xD1, 0x8C, 0x05, 0xB8, 0xCB, 0x32, 0x17, 0xAA, 0xEA, 0xEA, 0xB4,
+ 0x16, 0x11, 0x01, 0xA9, 0x08, 0x76, 0x37, 0x36, 0x6F, 0xDC, 0xCA, 0xC6, 0x92, 0x0D, 0xEA, 0x09,
+ 0x91, 0x03, 0x2F, 0x5B, 0x27, 0x4D, 0x94, 0x1D, 0x60, 0x50, 0xDC, 0x1F, 0x35, 0x57, 0xEC, 0x20,
+ 0xA6, 0xAC, 0x10, 0xDB, 0xCF, 0x36, 0x23, 0x8F, 0x96, 0xC7, 0x72, 0x8B, 0xF9, 0xE7, 0x30, 0xEB,
+ 0x50, 0x58, 0x4B, 0xFE, 0xAD, 0xC5, 0x13, 0x79, 0x50, 0x98, 0x1C, 0x67, 0xB2, 0xEB, 0xDA, 0x64,
+ 0xD4, 0xAA, 0xC4, 0xE8, 0xE5, 0xC9, 0xE7, 0x6B, 0x84, 0xC2, 0xD2, 0xE9, 0x1F, 0xA1, 0x0F, 0xDF,
+ 0x45, 0x06, 0x80, 0x25, 0x32, 0x6B, 0xC4, 0x66, 0x2A, 0x69, 0x9F, 0x1E, 0x1F, 0x4C, 0xBE, 0x89,
+ 0xFE, 0x61, 0xBB, 0x04, 0x79, 0xB5, 0x3B, 0x45, 0x58, 0xD9, 0x9C, 0x18, 0x7C, 0x48, 0x1B, 0x44,
+ 0x92, 0xC4, 0x4C, 0xD0, 0x8F, 0x16, 0x44, 0x79, 0x71, 0x48, 0x63, 0x0B, 0x51, 0xB6, 0x33, 0x8B,
+ 0x6B, 0x3F, 0xCC, 0x0A, 0x3A, 0x14, 0x3B, 0x49, 0x68, 0x46, 0xB9, 0x46, 0xC6, 0xA3, 0x03, 0x41
+};
+
+
// Stellite (XTL)
const static uint8_t test_output_xtl[160] = {
0x8F, 0xE5, 0xF0, 0x5F, 0x02, 0x2A, 0x61, 0x7D, 0xE5, 0x3F, 0x79, 0x36, 0x4B, 0x25, 0xCB, 0xC3,
diff --git a/src/crypto/CryptoNight_x86.h b/src/crypto/CryptoNight_x86.h
index 199c190bb..e234d41d0 100644
--- a/src/crypto/CryptoNight_x86.h
+++ b/src/crypto/CryptoNight_x86.h
@@ -408,20 +408,46 @@ static inline __m128i aes_round_tweak_div(const __m128i &in, const __m128i &key)
}
-template<int SHIFT>
-static inline void cryptonight_monero_tweak(uint64_t* mem_out, __m128i tmp)
+static inline __m128i int_sqrt_v2(const uint64_t n0)
{
- mem_out[0] = EXTRACT64(tmp);
+ __m128d x = _mm_castsi128_pd(_mm_add_epi64(_mm_cvtsi64_si128(n0 >> 12), _mm_set_epi64x(0, 1023ULL << 52)));
+ x = _mm_sqrt_sd(_mm_setzero_pd(), x);
+ uint64_t r = static_cast<uint64_t>(_mm_cvtsi128_si64(_mm_castpd_si128(x)));
- tmp = _mm_castps_si128(_mm_movehl_ps(_mm_castsi128_ps(tmp), _mm_castsi128_ps(tmp)));
- uint64_t vh = EXTRACT64(tmp);
+ const uint64_t s = r >> 20;
+ r >>= 19;
- uint8_t x = vh >> 24;
- static const uint16_t table = 0x7531;
- const uint8_t index = (((x >> SHIFT) & 6) | (x & 1)) << 1;
- vh ^= ((table >> index) & 0x3) << 28;
+ uint64_t x2 = (s - (1022ULL << 32)) * (r - s - (1022ULL << 32) + 1);
+#if defined _MSC_VER || (__GNUC__ >= 7)
+ _addcarry_u64(_subborrow_u64(0, x2, n0, (unsigned long long int*)&x2), r, 0, (unsigned long long int*)&r);
+#else
+ if (x2 < n0) ++r;
+#endif
- mem_out[1] = vh;
+ return _mm_cvtsi64_si128(r);
+}
+
+
+template<int VARIANT>
+static inline void cryptonight_monero_tweak(uint64_t* mem_out, const uint8_t* l, uint64_t idx, __m128i ax0, __m128i bx0, __m128i bx1, __m128i cx)
+{
+ if (VARIANT == xmrig::VARIANT_2) {
+ VARIANT2_SHUFFLE(l, idx, ax0, bx0, bx1);
+ _mm_store_si128((__m128i *)mem_out, _mm_xor_si128(bx0, cx));
+ } else {
+ __m128i tmp = _mm_xor_si128(bx0, cx);
+ mem_out[0] = EXTRACT64(tmp);
+
+ tmp = _mm_castps_si128(_mm_movehl_ps(_mm_castsi128_ps(tmp), _mm_castsi128_ps(tmp)));
+ uint64_t vh = EXTRACT64(tmp);
+
+ uint8_t x = static_cast<uint8_t>(vh >> 24);
+ static const uint16_t table = 0x7531;
+ const uint8_t index = (((x >> (VARIANT == xmrig::VARIANT_XTL ? 4 : 3)) & 6) | (x & 1)) << 1;
+ vh ^= ((table >> index) & 0x3) << 28;
+
+ mem_out[1] = vh;
+ }
}
@@ -431,92 +457,104 @@ inline void cryptonight_single_hash(const uint8_t *__restrict__ input, size_t si
constexpr size_t MASK = xmrig::cn_select_mask<ALGO>();
constexpr size_t ITERATIONS = xmrig::cn_select_iter<ALGO, VARIANT>();
constexpr size_t MEM = xmrig::cn_select_memory<ALGO>();
- constexpr bool IS_MONERO = xmrig::cn_is_monero<VARIANT>();
+ constexpr bool IS_V1 = xmrig::cn_uses_variant1<VARIANT>();
- if (IS_MONERO && size < 43) {
+ if (IS_V1 && size < 43) {
memset(output, 0, 32);
return;
}
xmrig::keccak(input, size, ctx[0]->state);
- VARIANT1_INIT(0)
-
cn_explode_scratchpad<ALGO, MEM, SOFT_AES>((__m128i*) ctx[0]->state, (__m128i*) ctx[0]->memory);
const uint8_t* l0 = ctx[0]->memory;
uint64_t* h0 = reinterpret_cast<uint64_t*>(ctx[0]->state);
+ VARIANT1_INIT(0);
+ VARIANT2_INIT(0);
+ VARIANT2_SET_ROUNDING_MODE();
+
uint64_t al0 = h0[0] ^ h0[4];
uint64_t ah0 = h0[1] ^ h0[5];
__m128i bx0 = _mm_set_epi64x(h0[3] ^ h0[7], h0[2] ^ h0[6]);
+ __m128i bx1 = _mm_set_epi64x(h0[9] ^ h0[11], h0[8] ^ h0[10]);
uint64_t idx0 = al0;
+ uint64_t* ptr0 = (uint64_t*) &l0[idx0 & MASK];
for (size_t i = 0; i < ITERATIONS; i++) {
__m128i cx;
if (VARIANT == xmrig::VARIANT_TUBE || !SOFT_AES) {
- cx = _mm_load_si128((__m128i *) &l0[idx0 & MASK]);
+ cx = _mm_load_si128((__m128i *) ptr0);
}
+ const __m128i ax0 = _mm_set_epi64x(ah0, al0);
if (VARIANT == xmrig::VARIANT_TUBE) {
- cx = aes_round_tweak_div(cx, _mm_set_epi64x(ah0, al0));
+ cx = aes_round_tweak_div(cx, ax0);
}
else if (SOFT_AES) {
- cx = soft_aesenc((uint32_t*)&l0[idx0 & MASK], _mm_set_epi64x(ah0, al0));
+ cx = soft_aesenc((uint32_t*) ptr0, ax0);
}
else {
- cx = _mm_aesenc_si128(cx, _mm_set_epi64x(ah0, al0));
+ cx = _mm_aesenc_si128(cx, ax0);
}
- if (IS_MONERO) {
- cryptonight_monero_tweak<VARIANT == xmrig::VARIANT_XTL ? 4 : 3>((uint64_t*)&l0[idx0 & MASK], _mm_xor_si128(bx0, cx));
+ if (IS_V1 || VARIANT == xmrig::VARIANT_2) {
+ cryptonight_monero_tweak<VARIANT>(ptr0, l0, idx0 & MASK, ax0, bx0, bx1, cx);
} else {
- _mm_store_si128((__m128i *)&l0[idx0 & MASK], _mm_xor_si128(bx0, cx));
+ _mm_store_si128((__m128i *) ptr0, _mm_xor_si128(bx0, cx));
}
idx0 = EXTRACT64(cx);
- bx0 = cx;
+ ptr0 = (uint64_t*) &l0[idx0 & MASK];
uint64_t hi, lo, cl, ch;
- cl = ((uint64_t*) &l0[idx0 & MASK])[0];
- ch = ((uint64_t*) &l0[idx0 & MASK])[1];
- lo = __umul128(idx0, cl, &hi);
+ cl = ptr0[0];
+ ch = ptr0[1];
+ if (VARIANT == xmrig::VARIANT_2) {
+ VARIANT2_INTEGER_MATH(0, cl, cx);
+ lo = __umul128(idx0, cl, &hi);
+ VARIANT2_SHUFFLE(l0, idx0 & MASK, ax0, bx0, bx1);
+ }
+ else {
+ lo = __umul128(idx0, cl, &hi);
+ }
al0 += hi;
ah0 += lo;
- ((uint64_t*)&l0[idx0 & MASK])[0] = al0;
+ ptr0[0] = al0;
- if (IS_MONERO) {
- if (VARIANT == xmrig::VARIANT_TUBE || VARIANT == xmrig::VARIANT_RTO) {
- ((uint64_t*)&l0[idx0 & MASK])[1] = ah0 ^ tweak1_2_0 ^ al0;
- }
- else {
- ((uint64_t*)&l0[idx0 & MASK])[1] = ah0 ^ tweak1_2_0;
- }
- }
- else {
- ((uint64_t*)&l0[idx0 & MASK])[1] = ah0;
+ if (IS_V1 && (VARIANT == xmrig::VARIANT_TUBE || VARIANT == xmrig::VARIANT_RTO)) {
+ ptr0[1] = ah0 ^ tweak1_2_0 ^ al0;
+ } else if (IS_V1) {
+ ptr0[1] = ah0 ^ tweak1_2_0;
+ } else {
+ ptr0[1] = ah0;
}
al0 ^= cl;
ah0 ^= ch;
idx0 = al0;
+ ptr0 = (uint64_t*) &l0[idx0 & MASK];
if (ALGO == xmrig::CRYPTONIGHT_HEAVY) {
- int64_t n = ((int64_t*)&l0[idx0 & MASK])[0];
- int32_t d = ((int32_t*)&l0[idx0 & MASK])[2];
+ int64_t n = ((int64_t*)ptr0)[0];
+ int32_t d = ((int32_t*)ptr0)[2];
int64_t q = n / (d | 0x5);
- ((int64_t*)&l0[idx0 & MASK])[0] = n ^ q;
+ ((int64_t*) ptr0)[0] = n ^ q;
if (VARIANT == xmrig::VARIANT_XHV) {
d = ~d;
}
idx0 = d ^ q;
+ ptr0 = (uint64_t*)&l0[idx0 & MASK];
}
+ bx1 = bx0;
+ bx0 = cx;
}
cn_implode_scratchpad<ALGO, MEM, SOFT_AES>((__m128i*) ctx[0]->memory, (__m128i*) ctx[0]->state);
@@ -532,9 +570,9 @@ inline void cryptonight_double_hash(const uint8_t *__restrict__ input, size_t si
constexpr size_t MASK = xmrig::cn_select_mask<ALGO>();
constexpr size_t ITERATIONS = xmrig::cn_select_iter<ALGO, VARIANT>();
constexpr size_t MEM = xmrig::cn_select_memory<ALGO>();
- constexpr bool IS_MONERO = xmrig::cn_is_monero<VARIANT>();
+ constexpr bool IS_V1 = xmrig::cn_uses_variant1<VARIANT>();
- if (IS_MONERO && size < 43) {
+ if (IS_V1 && size < 43) {
memset(output, 0, 64);
return;
}
@@ -542,14 +580,17 @@ inline void cryptonight_double_hash(const uint8_t *__restrict__ input, size_t si
xmrig::keccak(input, size, ctx[0]->state);
xmrig::keccak(input + size, size, ctx[1]->state);
- VARIANT1_INIT(0);
- VARIANT1_INIT(1);
-
const uint8_t* l0 = ctx[0]->memory;
const uint8_t* l1 = ctx[1]->memory;
uint64_t* h0 = reinterpret_cast<uint64_t*>(ctx[0]->state);
uint64_t* h1 = reinterpret_cast<uint64_t*>(ctx[1]->state);
+ VARIANT1_INIT(0);
+ VARIANT1_INIT(1);
+ VARIANT2_INIT(0);
+ VARIANT2_INIT(1);
+ VARIANT2_SET_ROUNDING_MODE();
+
cn_explode_scratchpad<ALGO, MEM, SOFT_AES>((__m128i*) h0, (__m128i*) l0);
cn_explode_scratchpad<ALGO, MEM, SOFT_AES>((__m128i*) h1, (__m128i*) l1);
@@ -558,124 +599,142 @@ inline void cryptonight_double_hash(const uint8_t *__restrict__ input, size_t si
uint64_t ah0 = h0[1] ^ h0[5];
uint64_t ah1 = h1[1] ^ h1[5];
- __m128i bx0 = _mm_set_epi64x(h0[3] ^ h0[7], h0[2] ^ h0[6]);
- __m128i bx1 = _mm_set_epi64x(h1[3] ^ h1[7], h1[2] ^ h1[6]);
+ __m128i bx00 = _mm_set_epi64x(h0[3] ^ h0[7], h0[2] ^ h0[6]);
+ __m128i bx01 = _mm_set_epi64x(h0[9] ^ h0[11], h0[8] ^ h0[10]);
+ __m128i bx10 = _mm_set_epi64x(h1[3] ^ h1[7], h1[2] ^ h1[6]);
+ __m128i bx11 = _mm_set_epi64x(h1[9] ^ h1[11], h1[8] ^ h1[10]);
uint64_t idx0 = al0;
uint64_t idx1 = al1;
+ uint64_t* ptr0 = (uint64_t*)&l0[idx0 & MASK];
+ uint64_t* ptr1 = (uint64_t*)&l1[idx1 & MASK];
for (size_t i = 0; i < ITERATIONS; i++) {
__m128i cx0, cx1;
if (VARIANT == xmrig::VARIANT_TUBE || !SOFT_AES) {
- cx0 = _mm_load_si128((__m128i *) &l0[idx0 & MASK]);
- cx1 = _mm_load_si128((__m128i *) &l1[idx1 & MASK]);
+ cx0 = _mm_load_si128((__m128i *) ptr0);
+ cx1 = _mm_load_si128((__m128i *) ptr1);
}
+ const __m128i ax0 = _mm_set_epi64x(ah0, al0);
+ const __m128i ax1 = _mm_set_epi64x(ah1, al1);
if (VARIANT == xmrig::VARIANT_TUBE) {
- cx0 = aes_round_tweak_div(cx0, _mm_set_epi64x(ah0, al0));
- cx1 = aes_round_tweak_div(cx1, _mm_set_epi64x(ah1, al1));
+ cx0 = aes_round_tweak_div(cx0, ax0);
+ cx1 = aes_round_tweak_div(cx1, ax1);
}
else if (SOFT_AES) {
- cx0 = soft_aesenc((uint32_t*)&l0[idx0 & MASK], _mm_set_epi64x(ah0, al0));
- cx1 = soft_aesenc((uint32_t*)&l1[idx1 & MASK], _mm_set_epi64x(ah1, al1));
+ cx0 = soft_aesenc((uint32_t*)ptr0, ax0);
+ cx1 = soft_aesenc((uint32_t*)ptr1, ax1);
}
else {
- cx0 = _mm_aesenc_si128(cx0, _mm_set_epi64x(ah0, al0));
- cx1 = _mm_aesenc_si128(cx1, _mm_set_epi64x(ah1, al1));
+ cx0 = _mm_aesenc_si128(cx0, ax0);
+ cx1 = _mm_aesenc_si128(cx1, ax1);
}
- if (IS_MONERO) {
- cryptonight_monero_tweak<VARIANT == xmrig::VARIANT_XTL ? 4 : 3>((uint64_t*)&l0[idx0 & MASK], _mm_xor_si128(bx0, cx0));
- cryptonight_monero_tweak<VARIANT == xmrig::VARIANT_XTL ? 4 : 3>((uint64_t*)&l1[idx1 & MASK], _mm_xor_si128(bx1, cx1));
+ if (IS_V1 || (VARIANT == xmrig::VARIANT_2)) {
+ cryptonight_monero_tweak<VARIANT>(ptr0, l0, idx0 & MASK, ax0, bx00, bx01, cx0);
+ cryptonight_monero_tweak<VARIANT>(ptr1, l1, idx1 & MASK, ax1, bx10, bx11, cx1);
} else {
- _mm_store_si128((__m128i *) &l0[idx0 & MASK], _mm_xor_si128(bx0, cx0));
- _mm_store_si128((__m128i *) &l1[idx1 & MASK], _mm_xor_si128(bx1, cx1));
+ _mm_store_si128((__m128i *) ptr0, _mm_xor_si128(bx00, cx0));
+ _mm_store_si128((__m128i *) ptr1, _mm_xor_si128(bx10, cx1));
}
idx0 = EXTRACT64(cx0);
idx1 = EXTRACT64(cx1);
-
- bx0 = cx0;
- bx1 = cx1;
+ ptr0 = (uint64_t*)&l0[idx0 & MASK];
+ ptr1 = (uint64_t*)&l1[idx1 & MASK];
uint64_t hi, lo, cl, ch;
- cl = ((uint64_t*) &l0[idx0 & MASK])[0];
- ch = ((uint64_t*) &l0[idx0 & MASK])[1];
- lo = __umul128(idx0, cl, &hi);
+ cl = ((uint64_t*)ptr0)[0];
+ ch = ((uint64_t*)ptr0)[1];
+ if (VARIANT == xmrig::VARIANT_2) {
+ VARIANT2_INTEGER_MATH(0, cl, cx0);
+ lo = __umul128(idx0, cl, &hi);
+ VARIANT2_SHUFFLE(l0, idx0 & MASK, ax0, bx00, bx01);
+ } else {
+ lo = __umul128(idx0, cl, &hi);
+ }
al0 += hi;
ah0 += lo;
- ((uint64_t*)&l0[idx0 & MASK])[0] = al0;
+ ((uint64_t*)ptr0)[0] = al0;
- if (IS_MONERO) {
- if (VARIANT == xmrig::VARIANT_TUBE || VARIANT == xmrig::VARIANT_RTO) {
- ((uint64_t*)&l0[idx0 & MASK])[1] = ah0 ^ tweak1_2_0 ^ al0;
- }
- else {
- ((uint64_t*)&l0[idx0 & MASK])[1] = ah0 ^ tweak1_2_0;
- }
- }
- else {
- ((uint64_t*)&l0[idx0 & MASK])[1] = ah0;
+ if (IS_V1 && (VARIANT == xmrig::VARIANT_TUBE || VARIANT == xmrig::VARIANT_RTO)) {
+ ((uint64_t*)ptr0)[1] = ah0 ^ tweak1_2_0 ^ al0;
+ } else if (IS_V1) {
+ ((uint64_t*)ptr0)[1] = ah0 ^ tweak1_2_0;
+ } else {
+ ((uint64_t*)ptr0)[1] = ah0;
}
al0 ^= cl;
ah0 ^= ch;
idx0 = al0;
+ ptr0 = (uint64_t*)&l0[idx0 & MASK];
if (ALGO == xmrig::CRYPTONIGHT_HEAVY) {
- int64_t n = ((int64_t*)&l0[idx0 & MASK])[0];
- int32_t d = ((int32_t*)&l0[idx0 & MASK])[2];
+ int64_t n = ((int64_t*)ptr0)[0];
+ int32_t d = ((int32_t*)ptr0)[2];
int64_t q = n / (d | 0x5);
- ((int64_t*)&l0[idx0 & MASK])[0] = n ^ q;
+ ((int64_t*)ptr0)[0] = n ^ q;
if (VARIANT == xmrig::VARIANT_XHV) {
d = ~d;
}
idx0 = d ^ q;
+ ptr0 = (uint64_t*)&l0[idx0 & MASK];
}
- cl = ((uint64_t*) &l1[idx1 & MASK])[0];
- ch = ((uint64_t*) &l1[idx1 & MASK])[1];
- lo = __umul128(idx1, cl, &hi);
+ cl = ptr1[0];
+ ch = ptr1[1];
+ if (VARIANT == xmrig::VARIANT_2) {
+ VARIANT2_INTEGER_MATH(1, cl, cx1);
+ lo = __umul128(idx1, cl, &hi);
+ VARIANT2_SHUFFLE(l1, idx1 & MASK, ax1, bx10, bx11);
+ } else {
+ lo = __umul128(idx1, cl, &hi);
+ }
al1 += hi;
ah1 += lo;
- ((uint64_t*)&l1[idx1 & MASK])[0] = al1;
+ ptr1[0] = al1;
- if (IS_MONERO) {
- if (VARIANT == xmrig::VARIANT_TUBE || VARIANT == xmrig::VARIANT_RTO) {
- ((uint64_t*)&l1[idx1 & MASK])[1] = ah1 ^ tweak1_2_1 ^ al1;
- }
- else {
- ((uint64_t*)&l1[idx1 & MASK])[1] = ah1 ^ tweak1_2_1;
- }
- }
- else {
- ((uint64_t*)&l1[idx1 & MASK])[1] = ah1;
+ if (IS_V1 && (VARIANT == xmrig::VARIANT_TUBE || VARIANT == xmrig::VARIANT_RTO)) {
+ ptr1[1] = ah1 ^ tweak1_2_1 ^ al1;
+ } else if (IS_V1) {
+ ptr1[1] = ah1 ^ tweak1_2_1;
+ } else {
+ ptr1[1] = ah1;
}
al1 ^= cl;
ah1 ^= ch;
idx1 = al1;
+ ptr1 = (uint64_t*)&l1[idx1 & MASK];
if (ALGO == xmrig::CRYPTONIGHT_HEAVY) {
- int64_t n = ((int64_t*)&l1[idx1 & MASK])[0];
- int32_t d = ((int32_t*)&l1[idx1 & MASK])[2];
+ int64_t n = ((int64_t*)ptr1)[0];
+ int32_t d = ((int32_t*)ptr1)[2];
int64_t q = n / (d | 0x5);
- ((int64_t*)&l1[idx1 & MASK])[0] = n ^ q;
+ ((int64_t*)ptr1)[0] = n ^ q;
if (VARIANT == xmrig::VARIANT_XHV) {
d = ~d;
}
idx1 = d ^ q;
+ ptr1 = (uint64_t*)&l1[idx1 & MASK];
}
+
+ bx01 = bx00;
+ bx00 = cx0;
+ bx11 = bx10;
+ bx10 = cx1;
}
cn_implode_scratchpad<ALGO, MEM, SOFT_AES>((__m128i*) l0, (__m128i*) h0);
@@ -689,12 +748,12 @@ inline void cryptonight_double_hash(const uint8_t *__restrict__ input, size_t si
}
-#define CN_STEP1(a, b, c, l, ptr, idx) \
+#define CN_STEP1(a, b0, b1, c, l, ptr, idx) \
ptr = reinterpret_cast<__m128i*>(&l[idx & MASK]); \
c = _mm_load_si128(ptr);
-#define CN_STEP2(a, b, c, l, ptr, idx) \
+#define CN_STEP2(a, b0, b1, c, l, ptr, idx) \
if (VARIANT == xmrig::VARIANT_TUBE) { \
c = aes_round_tweak_div(c, a); \
} \
@@ -704,26 +763,31 @@ inline void cryptonight_double_hash(const uint8_t *__restrict__ input, size_t si
c = _mm_aesenc_si128(c, a); \
} \
\
- b = _mm_xor_si128(b, c); \
- \
- if (IS_MONERO) { \
- cryptonight_monero_tweak<VARIANT == xmrig::VARIANT_XTL ? 4 : 3>(reinterpret_cast<uint64_t*>(ptr), b); \
+ if (IS_V1 || (VARIANT == xmrig::VARIANT_2)) { \
+ cryptonight_monero_tweak<VARIANT>((uint64_t*)ptr, l, idx & MASK, a, b0, b1, c); \
} else { \
- _mm_store_si128(ptr, b); \
+ _mm_store_si128(ptr, _mm_xor_si128(b0, c)); \
}
-#define CN_STEP3(a, b, c, l, ptr, idx) \
+#define CN_STEP3(part, a, b0, b1, c, l, ptr, idx) \
idx = EXTRACT64(c); \
ptr = reinterpret_cast<__m128i*>(&l[idx & MASK]); \
- b = _mm_load_si128(ptr);
+ uint64_t cl##part = ((uint64_t*)ptr)[0]; \
+ uint64_t ch##part = ((uint64_t*)ptr)[1];
-#define CN_STEP4(a, b, c, l, mc, ptr, idx) \
- lo = __umul128(idx, EXTRACT64(b), &hi); \
+#define CN_STEP4(part, a, b0, b1, c, l, mc, ptr, idx) \
+ if (VARIANT == xmrig::VARIANT_2) { \
+ VARIANT2_INTEGER_MATH(part, cl##part, c); \
+ lo = __umul128(idx, cl##part, &hi); \
+ VARIANT2_SHUFFLE(l, idx & MASK, a, b0, b1); \
+ } else { \
+ lo = __umul128(idx, cl##part, &hi); \
+ } \
a = _mm_add_epi64(a, _mm_set_epi64x(lo, hi)); \
\
- if (IS_MONERO) { \
+ if (IS_V1) { \
_mm_store_si128(ptr, _mm_xor_si128(a, mc)); \
\
if (VARIANT == xmrig::VARIANT_TUBE || \
@@ -734,7 +798,7 @@ inline void cryptonight_double_hash(const uint8_t *__restrict__ input, size_t si
_mm_store_si128(ptr, a); \
} \
\
- a = _mm_xor_si128(a, b); \
+ a = _mm_xor_si128(a, _mm_set_epi64x(ch##part, cl##part)); \
idx = EXTRACT64(a); \
\
if (ALGO == xmrig::CRYPTONIGHT_HEAVY) { \
@@ -747,15 +811,27 @@ inline void cryptonight_double_hash(const uint8_t *__restrict__ input, size_t si
} \
\
idx = d ^ q; \
- }
+ } \
+ b1 = b0; \
+ b0 = c;
#define CONST_INIT(ctx, n) \
__m128i mc##n; \
- if (IS_MONERO) { \
+ __m128i division_result_xmm_##n; \
+ __m128i sqrt_result_xmm_##n; \
+ if (IS_V1) { \
mc##n = _mm_set_epi64x(*reinterpret_cast<const uint64_t*>(input + n * size + 35) ^ \
*(reinterpret_cast<const uint64_t*>((ctx)->state) + 24), 0); \
- }
+ } \
+ if (VARIANT == xmrig::VARIANT_2) { \
+ division_result_xmm_##n = _mm_cvtsi64_si128(h##n[12]); \
+ sqrt_result_xmm_##n = _mm_cvtsi64_si128(h##n[13]); \
+ } \
+ __m128i ax##n = _mm_set_epi64x(h##n[1] ^ h##n[5], h##n[0] ^ h##n[4]); \
+ __m128i bx##n##0 = _mm_set_epi64x(h##n[3] ^ h##n[7], h##n[2] ^ h##n[6]); \
+ __m128i bx##n##1 = _mm_set_epi64x(h##n[9] ^ h##n[11], h##n[8] ^ h##n[10]); \
+ __m128i cx##n = _mm_setzero_si128();
template<xmrig::Algo ALGO, bool SOFT_AES, xmrig::Variant VARIANT>
@@ -764,9 +840,9 @@ inline void cryptonight_triple_hash(const uint8_t *__restrict__ input, size_t si
constexpr size_t MASK = xmrig::cn_select_mask<ALGO>();
constexpr size_t ITERATIONS = xmrig::cn_select_iter<ALGO, VARIANT>();
constexpr size_t MEM = xmrig::cn_select_memory<ALGO>();
- constexpr bool IS_MONERO = xmrig::cn_is_monero<VARIANT>();
+ constexpr bool IS_V1 = xmrig::cn_uses_variant1<VARIANT>();
- if (IS_MONERO && size < 43) {
+ if (IS_V1 && size < 43) {
memset(output, 0, 32 * 3);
return;
}
@@ -776,10 +852,6 @@ inline void cryptonight_triple_hash(const uint8_t *__restrict__ input, size_t si
cn_explode_scratchpad<ALGO, MEM, SOFT_AES>(reinterpret_cast<__m128i*>(ctx[i]->state), reinterpret_cast<__m128i*>(ctx[i]->memory));
}
- CONST_INIT(ctx[0], 0);
- CONST_INIT(ctx[1], 1);
- CONST_INIT(ctx[2], 2);
-
uint8_t* l0 = ctx[0]->memory;
uint8_t* l1 = ctx[1]->memory;
uint8_t* l2 = ctx[2]->memory;
@@ -787,58 +859,35 @@ inline void cryptonight_triple_hash(const uint8_t *__restrict__ input, size_t si
uint64_t* h1 = reinterpret_cast<uint64_t*>(ctx[1]->state);
uint64_t* h2 = reinterpret_cast<uint64_t*>(ctx[2]->state);
- __m128i ax0 = _mm_set_epi64x(h0[1] ^ h0[5], h0[0] ^ h0[4]);
- __m128i bx0 = _mm_set_epi64x(h0[3] ^ h0[7], h0[2] ^ h0[6]);
- __m128i ax1 = _mm_set_epi64x(h1[1] ^ h1[5], h1[0] ^ h1[4]);
- __m128i bx1 = _mm_set_epi64x(h1[3] ^ h1[7], h1[2] ^ h1[6]);
- __m128i ax2 = _mm_set_epi64x(h2[1] ^ h2[5], h2[0] ^ h2[4]);
- __m128i bx2 = _mm_set_epi64x(h2[3] ^ h2[7], h2[2] ^ h2[6]);
- __m128i cx0 = _mm_set_epi64x(0, 0);
- __m128i cx1 = _mm_set_epi64x(0, 0);
- __m128i cx2 = _mm_set_epi64x(0, 0);
+ CONST_INIT(ctx[0], 0);
+ CONST_INIT(ctx[1], 1);
+ CONST_INIT(ctx[2], 2);
+ VARIANT2_SET_ROUNDING_MODE();
uint64_t idx0, idx1, idx2;
idx0 = EXTRACT64(ax0);
idx1 = EXTRACT64(ax1);
idx2 = EXTRACT64(ax2);
- for (size_t i = 0; i < ITERATIONS / 2; i++) {
+ for (size_t i = 0; i < ITERATIONS; i++) {
uint64_t hi, lo;
__m128i *ptr0, *ptr1, *ptr2;
- // EVEN ROUND
- CN_STEP1(ax0, bx0, cx0, l0, ptr0, idx0);
- CN_STEP1(ax1, bx1, cx1, l1, ptr1, idx1);
- CN_STEP1(ax2, bx2, cx2, l2, ptr2, idx2);
+ CN_STEP1(ax0, bx00, bx01, cx0, l0, ptr0, idx0);
+ CN_STEP1(ax1, bx10, bx11, cx1, l1, ptr1, idx1);
+ CN_STEP1(ax2, bx20, bx21, cx2, l2, ptr2, idx2);
- CN_STEP2(ax0, bx0, cx0, l0, ptr0, idx0);
- CN_STEP2(ax1, bx1, cx1, l1, ptr1, idx1);
- CN_STEP2(ax2, bx2, cx2, l2, ptr2, idx2);
+ CN_STEP2(ax0, bx00, bx01, cx0, l0, ptr0, idx0);
+ CN_STEP2(ax1, bx10, bx11, cx1, l1, ptr1, idx1);
+ CN_STEP2(ax2, bx20, bx21, cx2, l2, ptr2, idx2);
- CN_STEP3(ax0, bx0, cx0, l0, ptr0, idx0);
- CN_STEP3(ax1, bx1, cx1, l1, ptr1, idx1);
- CN_STEP3(ax2, bx2, cx2, l2, ptr2, idx2);
+ CN_STEP3(0, ax0, bx00, bx01, cx0, l0, ptr0, idx0);
+ CN_STEP3(1, ax1, bx10, bx11, cx1, l1, ptr1, idx1);
+ CN_STEP3(2, ax2, bx20, bx21, cx2, l2, ptr2, idx2);
- CN_STEP4(ax0, bx0, cx0, l0, mc0, ptr0, idx0);
- CN_STEP4(ax1, bx1, cx1, l1, mc1, ptr1, idx1);
- CN_STEP4(ax2, bx2, cx2, l2, mc2, ptr2, idx2);
-
- // ODD ROUND
- CN_STEP1(ax0, cx0, bx0, l0, ptr0, idx0);
- CN_STEP1(ax1, cx1, bx1, l1, ptr1, idx1);
- CN_STEP1(ax2, cx2, bx2, l2, ptr2, idx2);
-
- CN_STEP2(ax0, cx0, bx0, l0, ptr0, idx0);
- CN_STEP2(ax1, cx1, bx1, l1, ptr1, idx1);
- CN_STEP2(ax2, cx2, bx2, l2, ptr2, idx2);
-
- CN_STEP3(ax0, cx0, bx0, l0, ptr0, idx0);
- CN_STEP3(ax1, cx1, bx1, l1, ptr1, idx1);
- CN_STEP3(ax2, cx2, bx2, l2, ptr2, idx2);
-
- CN_STEP4(ax0, cx0, bx0, l0, mc0, ptr0, idx0);
- CN_STEP4(ax1, cx1, bx1, l1, mc1, ptr1, idx1);
- CN_STEP4(ax2, cx2, bx2, l2, mc2, ptr2, idx2);
+ CN_STEP4(0, ax0, bx00, bx01, cx0, l0, mc0, ptr0, idx0);
+ CN_STEP4(1, ax1, bx10, bx11, cx1, l1, mc1, ptr1, idx1);
+ CN_STEP4(2, ax2, bx20, bx21, cx2, l2, mc2, ptr2, idx2);
}
for (size_t i = 0; i < 3; i++) {
@@ -855,9 +904,9 @@ inline void cryptonight_quad_hash(const uint8_t *__restrict__ input, size_t size
constexpr size_t MASK = xmrig::cn_select_mask<ALGO>();
constexpr size_t ITERATIONS = xmrig::cn_select_iter<ALGO, VARIANT>();
constexpr size_t MEM = xmrig::cn_select_memory<ALGO>();
- constexpr bool IS_MONERO = xmrig::cn_is_monero<VARIANT>();
+ constexpr bool IS_V1 = xmrig::cn_uses_variant1<VARIANT>();
- if (IS_MONERO && size < 43) {
+ if (IS_V1 && size < 43) {
memset(output, 0, 32 * 4);
return;
}
@@ -867,11 +916,6 @@ inline void cryptonight_quad_hash(const uint8_t *__restrict__ input, size_t size
cn_explode_scratchpad<ALGO, MEM, SOFT_AES>(reinterpret_cast<__m128i*>(ctx[i]->state), reinterpret_cast<__m128i*>(ctx[i]->memory));
}
- CONST_INIT(ctx[0], 0);
- CONST_INIT(ctx[1], 1);
- CONST_INIT(ctx[2], 2);
- CONST_INIT(ctx[3], 3);
-
uint8_t* l0 = ctx[0]->memory;
uint8_t* l1 = ctx[1]->memory;
uint8_t* l2 = ctx[2]->memory;
@@ -881,18 +925,11 @@ inline void cryptonight_quad_hash(const uint8_t *__restrict__ input, size_t size
uint64_t* h2 = reinterpret_cast<uint64_t*>(ctx[2]->state);
uint64_t* h3 = reinterpret_cast<uint64_t*>(ctx[3]->state);
- __m128i ax0 = _mm_set_epi64x(h0[1] ^ h0[5], h0[0] ^ h0[4]);
- __m128i bx0 = _mm_set_epi64x(h0[3] ^ h0[7], h0[2] ^ h0[6]);
- __m128i ax1 = _mm_set_epi64x(h1[1] ^ h1[5], h1[0] ^ h1[4]);
- __m128i bx1 = _mm_set_epi64x(h1[3] ^ h1[7], h1[2] ^ h1[6]);
- __m128i ax2 = _mm_set_epi64x(h2[1] ^ h2[5], h2[0] ^ h2[4]);
- __m128i bx2 = _mm_set_epi64x(h2[3] ^ h2[7], h2[2] ^ h2[6]);
- __m128i ax3 = _mm_set_epi64x(h3[1] ^ h3[5], h3[0] ^ h3[4]);
- __m128i bx3 = _mm_set_epi64x(h3[3] ^ h3[7], h3[2] ^ h3[6]);
- __m128i cx0 = _mm_set_epi64x(0, 0);
- __m128i cx1 = _mm_set_epi64x(0, 0);
- __m128i cx2 = _mm_set_epi64x(0, 0);
- __m128i cx3 = _mm_set_epi64x(0, 0);
+ CONST_INIT(ctx[0], 0);
+ CONST_INIT(ctx[1], 1);
+ CONST_INIT(ctx[2], 2);
+ CONST_INIT(ctx[3], 3);
+ VARIANT2_SET_ROUNDING_MODE();
uint64_t idx0, idx1, idx2, idx3;
idx0 = EXTRACT64(ax0);
@@ -900,52 +937,30 @@ inline void cryptonight_quad_hash(const uint8_t *__restrict__ input, size_t size
idx2 = EXTRACT64(ax2);
idx3 = EXTRACT64(ax3);
- for (size_t i = 0; i < ITERATIONS / 2; i++)
+ for (size_t i = 0; i < ITERATIONS; i++)
{
uint64_t hi, lo;
__m128i *ptr0, *ptr1, *ptr2, *ptr3;
- // EVEN ROUND
- CN_STEP1(ax0, bx0, cx0, l0, ptr0, idx0);
- CN_STEP1(ax1, bx1, cx1, l1, ptr1, idx1);
- CN_STEP1(ax2, bx2, cx2, l2, ptr2, idx2);
- CN_STEP1(ax3, bx3, cx3, l3, ptr3, idx3);
+ CN_STEP1(ax0, bx00, bx01, cx0, l0, ptr0, idx0);
+ CN_STEP1(ax1, bx10, bx11, cx1, l1, ptr1, idx1);
+ CN_STEP1(ax2, bx20, bx21, cx2, l2, ptr2, idx2);
+ CN_STEP1(ax3, bx30, bx31, cx3, l3, ptr3, idx3);
- CN_STEP2(ax0, bx0, cx0, l0, ptr0, idx0);
- CN_STEP2(ax1, bx1, cx1, l1, ptr1, idx1);
- CN_STEP2(ax2, bx2, cx2, l2, ptr2, idx2);
- CN_STEP2(ax3, bx3, cx3, l3, ptr3, idx3);
+ CN_STEP2(ax0, bx00, bx01, cx0, l0, ptr0, idx0);
+ CN_STEP2(ax1, bx10, bx11, cx1, l1, ptr1, idx1);
+ CN_STEP2(ax2, bx20, bx21, cx2, l2, ptr2, idx2);
+ CN_STEP2(ax3, bx30, bx31, cx3, l3, ptr3, idx3);
- CN_STEP3(ax0, bx0, cx0, l0, ptr0, idx0);
- CN_STEP3(ax1, bx1, cx1, l1, ptr1, idx1);
- CN_STEP3(ax2, bx2, cx2, l2, ptr2, idx2);
- CN_STEP3(ax3, bx3, cx3, l3, ptr3, idx3);
+ CN_STEP3(0, ax0, bx00, bx01, cx0, l0, ptr0, idx0);
+ CN_STEP3(1, ax1, bx10, bx11, cx1, l1, ptr1, idx1);
+ CN_STEP3(2, ax2, bx20, bx21, cx2, l2, ptr2, idx2);
+ CN_STEP3(3, ax3, bx30, bx31, cx3, l3, ptr3, idx3);
- CN_STEP4(ax0, bx0, cx0, l0, mc0, ptr0, idx0);
- CN_STEP4(ax1, bx1, cx1, l1, mc1, ptr1, idx1);
- CN_STEP4(ax2, bx2, cx2, l2, mc2, ptr2, idx2);
- CN_STEP4(ax3, bx3, cx3, l3, mc3, ptr3, idx3);
-
- // ODD ROUND
- CN_STEP1(ax0, cx0, bx0, l0, ptr0, idx0);
- CN_STEP1(ax1, cx1, bx1, l1, ptr1, idx1);
- CN_STEP1(ax2, cx2, bx2, l2, ptr2, idx2);
- CN_STEP1(ax3, cx3, bx3, l3, ptr3, idx3);
-
- CN_STEP2(ax0, cx0, bx0, l0, ptr0, idx0);
- CN_STEP2(ax1, cx1, bx1, l1, ptr1, idx1);
- CN_STEP2(ax2, cx2, bx2, l2, ptr2, idx2);
- CN_STEP2(ax3, cx3, bx3, l3, ptr3, idx3);
-
- CN_STEP3(ax0, cx0, bx0, l0, ptr0, idx0);
- CN_STEP3(ax1, cx1, bx1, l1, ptr1, idx1);
- CN_STEP3(ax2, cx2, bx2, l2, ptr2, idx2);
- CN_STEP3(ax3, cx3, bx3, l3, ptr3, idx3);
-
- CN_STEP4(ax0, cx0, bx0, l0, mc0, ptr0, idx0);
- CN_STEP4(ax1, cx1, bx1, l1, mc1, ptr1, idx1);
- CN_STEP4(ax2, cx2, bx2, l2, mc2, ptr2, idx2);
- CN_STEP4(ax3, cx3, bx3, l3, mc3, ptr3, idx3);
+ CN_STEP4(0, ax0, bx00, bx01, cx0, l0, mc0, ptr0, idx0);
+ CN_STEP4(1, ax1, bx10, bx11, cx1, l1, mc1, ptr1, idx1);
+ CN_STEP4(2, ax2, bx20, bx21, cx2, l2, mc2, ptr2, idx2);
+ CN_STEP4(3, ax3, bx30, bx31, cx3, l3, mc3, ptr3, idx3);
}
for (size_t i = 0; i < 4; i++) {
@@ -962,9 +977,9 @@ inline void cryptonight_penta_hash(const uint8_t *__restrict__ input, size_t siz
constexpr size_t MASK = xmrig::cn_select_mask<ALGO>();
constexpr size_t ITERATIONS = xmrig::cn_select_iter<ALGO, VARIANT>();
constexpr size_t MEM = xmrig::cn_select_memory<ALGO>();
- constexpr bool IS_MONERO = xmrig::cn_is_monero<VARIANT>();
+ constexpr bool IS_V1 = xmrig::cn_uses_variant1<VARIANT>();
- if (IS_MONERO && size < 43) {
+ if (IS_V1 && size < 43) {
memset(output, 0, 32 * 5);
return;
}
@@ -974,12 +989,6 @@ inline void cryptonight_penta_hash(const uint8_t *__restrict__ input, size_t siz
cn_explode_scratchpad<ALGO, MEM, SOFT_AES>(reinterpret_cast<__m128i*>(ctx[i]->state), reinterpret_cast<__m128i*>(ctx[i]->memory));
}
- CONST_INIT(ctx[0], 0);
- CONST_INIT(ctx[1], 1);
- CONST_INIT(ctx[2], 2);
- CONST_INIT(ctx[3], 3);
- CONST_INIT(ctx[4], 4);
-
uint8_t* l0 = ctx[0]->memory;
uint8_t* l1 = ctx[1]->memory;
uint8_t* l2 = ctx[2]->memory;
@@ -991,21 +1000,12 @@ inline void cryptonight_penta_hash(const uint8_t *__restrict__ input, size_t siz
uint64_t* h3 = reinterpret_cast<uint64_t*>(ctx[3]->state);
uint64_t* h4 = reinterpret_cast<uint64_t*>(ctx[4]->state);
- __m128i ax0 = _mm_set_epi64x(h0[1] ^ h0[5], h0[0] ^ h0[4]);
- __m128i bx0 = _mm_set_epi64x(h0[3] ^ h0[7], h0[2] ^ h0[6]);
- __m128i ax1 = _mm_set_epi64x(h1[1] ^ h1[5], h1[0] ^ h1[4]);
- __m128i bx1 = _mm_set_epi64x(h1[3] ^ h1[7], h1[2] ^ h1[6]);
- __m128i ax2 = _mm_set_epi64x(h2[1] ^ h2[5], h2[0] ^ h2[4]);
- __m128i bx2 = _mm_set_epi64x(h2[3] ^ h2[7], h2[2] ^ h2[6]);
- __m128i ax3 = _mm_set_epi64x(h3[1] ^ h3[5], h3[0] ^ h3[4]);
- __m128i bx3 = _mm_set_epi64x(h3[3] ^ h3[7], h3[2] ^ h3[6]);
- __m128i ax4 = _mm_set_epi64x(h4[1] ^ h4[5], h4[0] ^ h4[4]);
- __m128i bx4 = _mm_set_epi64x(h4[3] ^ h4[7], h4[2] ^ h4[6]);
- __m128i cx0 = _mm_set_epi64x(0, 0);
- __m128i cx1 = _mm_set_epi64x(0, 0);
- __m128i cx2 = _mm_set_epi64x(0, 0);
- __m128i cx3 = _mm_set_epi64x(0, 0);
- __m128i cx4 = _mm_set_epi64x(0, 0);
+ CONST_INIT(ctx[0], 0);
+ CONST_INIT(ctx[1], 1);
+ CONST_INIT(ctx[2], 2);
+ CONST_INIT(ctx[3], 3);
+ CONST_INIT(ctx[4], 4);
+ VARIANT2_SET_ROUNDING_MODE();
uint64_t idx0, idx1, idx2, idx3, idx4;
idx0 = EXTRACT64(ax0);
@@ -1014,60 +1014,34 @@ inline void cryptonight_penta_hash(const uint8_t *__restrict__ input, size_t siz
idx3 = EXTRACT64(ax3);
idx4 = EXTRACT64(ax4);
- for (size_t i = 0; i < ITERATIONS / 2; i++)
+ for (size_t i = 0; i < ITERATIONS; i++)
{
uint64_t hi, lo;
__m128i *ptr0, *ptr1, *ptr2, *ptr3, *ptr4;
- // EVEN ROUND
- CN_STEP1(ax0, bx0, cx0, l0, ptr0, idx0);
- CN_STEP1(ax1, bx1, cx1, l1, ptr1, idx1);
- CN_STEP1(ax2, bx2, cx2, l2, ptr2, idx2);
- CN_STEP1(ax3, bx3, cx3, l3, ptr3, idx3);
- CN_STEP1(ax4, bx4, cx4, l4, ptr4, idx4);
+ CN_STEP1(ax0, bx00, bx01, cx0, l0, ptr0, idx0);
+ CN_STEP1(ax1, bx10, bx11, cx1, l1, ptr1, idx1);
+ CN_STEP1(ax2, bx20, bx21, cx2, l2, ptr2, idx2);
+ CN_STEP1(ax3, bx30, bx31, cx3, l3, ptr3, idx3);
+ CN_STEP1(ax4, bx40, bx41, cx4, l4, ptr4, idx4);
- CN_STEP2(ax0, bx0, cx0, l0, ptr0, idx0);
- CN_STEP2(ax1, bx1, cx1, l1, ptr1, idx1);
- CN_STEP2(ax2, bx2, cx2, l2, ptr2, idx2);
- CN_STEP2(ax3, bx3, cx3, l3, ptr3, idx3);
- CN_STEP2(ax4, bx4, cx4, l4, ptr4, idx4);
+ CN_STEP2(ax0, bx00, bx01, cx0, l0, ptr0, idx0);
+ CN_STEP2(ax1, bx10, bx11, cx1, l1, ptr1, idx1);
+ CN_STEP2(ax2, bx20, bx21, cx2, l2, ptr2, idx2);
+ CN_STEP2(ax3, bx30, bx31, cx3, l3, ptr3, idx3);
+ CN_STEP2(ax4, bx40, bx41, cx4, l4, ptr4, idx4);
- CN_STEP3(ax0, bx0, cx0, l0, ptr0, idx0);
- CN_STEP3(ax1, bx1, cx1, l1, ptr1, idx1);
- CN_STEP3(ax2, bx2, cx2, l2, ptr2, idx2);
- CN_STEP3(ax3, bx3, cx3, l3, ptr3, idx3);
- CN_STEP3(ax4, bx4, cx4, l4, ptr4, idx4);
+ CN_STEP3(0, ax0, bx00, bx01, cx0, l0, ptr0, idx0);
+ CN_STEP3(1, ax1, bx10, bx11, cx1, l1, ptr1, idx1);
+ CN_STEP3(2, ax2, bx20, bx21, cx2, l2, ptr2, idx2);
+ CN_STEP3(3, ax3, bx30, bx31, cx3, l3, ptr3, idx3);
+ CN_STEP3(4, ax4, bx40, bx41, cx4, l4, ptr4, idx4);
- CN_STEP4(ax0, bx0, cx0, l0, mc0, ptr0, idx0);
- CN_STEP4(ax1, bx1, cx1, l1, mc1, ptr1, idx1);
- CN_STEP4(ax2, bx2, cx2, l2, mc2, ptr2, idx2);
- CN_STEP4(ax3, bx3, cx3, l3, mc3, ptr3, idx3);
- CN_STEP4(ax4, bx4, cx4, l4, mc4, ptr4, idx4);
-
- // ODD ROUND
- CN_STEP1(ax0, cx0, bx0, l0, ptr0, idx0);
- CN_STEP1(ax1, cx1, bx1, l1, ptr1, idx1);
- CN_STEP1(ax2, cx2, bx2, l2, ptr2, idx2);
- CN_STEP1(ax3, cx3, bx3, l3, ptr3, idx3);
- CN_STEP1(ax4, cx4, bx4, l4, ptr4, idx4);
-
- CN_STEP2(ax0, cx0, bx0, l0, ptr0, idx0);
- CN_STEP2(ax1, cx1, bx1, l1, ptr1, idx1);
- CN_STEP2(ax2, cx2, bx2, l2, ptr2, idx2);
- CN_STEP2(ax3, cx3, bx3, l3, ptr3, idx3);
- CN_STEP2(ax4, cx4, bx4, l4, ptr4, idx4);
-
- CN_STEP3(ax0, cx0, bx0, l0, ptr0, idx0);
- CN_STEP3(ax1, cx1, bx1, l1, ptr1, idx1);
- CN_STEP3(ax2, cx2, bx2, l2, ptr2, idx2);
- CN_STEP3(ax3, cx3, bx3, l3, ptr3, idx3);
- CN_STEP3(ax4, cx4, bx4, l4, ptr4, idx4);
-
- CN_STEP4(ax0, cx0, bx0, l0, mc0, ptr0, idx0);
- CN_STEP4(ax1, cx1, bx1, l1, mc1, ptr1, idx1);
- CN_STEP4(ax2, cx2, bx2, l2, mc2, ptr2, idx2);
- CN_STEP4(ax3, cx3, bx3, l3, mc3, ptr3, idx3);
- CN_STEP4(ax4, cx4, bx4, l4, mc4, ptr4, idx4);
+ CN_STEP4(0, ax0, bx00, bx01, cx0, l0, mc0, ptr0, idx0);
+ CN_STEP4(1, ax1, bx10, bx11, cx1, l1, mc1, ptr1, idx1);
+ CN_STEP4(2, ax2, bx20, bx21, cx2, l2, mc2, ptr2, idx2);
+ CN_STEP4(3, ax3, bx30, bx31, cx3, l3, mc3, ptr3, idx3);
+ CN_STEP4(4, ax4, bx40, bx41, cx4, l4, mc4, ptr4, idx4);
}
for (size_t i = 0; i < 5; i++) {
diff --git a/src/workers/CpuThread.cpp b/src/workers/CpuThread.cpp
index 7cef4f3ad..bdf09af4b 100644
--- a/src/workers/CpuThread.cpp
+++ b/src/workers/CpuThread.cpp
@@ -135,6 +135,17 @@ xmrig::CpuThread::cn_hash_fun xmrig::CpuThread::fn(Algo algorithm, AlgoVariant a
cryptonight_quad_hash<CRYPTONIGHT, true, VARIANT_RTO>,
cryptonight_penta_hash<CRYPTONIGHT, true, VARIANT_RTO>,
+ cryptonight_single_hash<CRYPTONIGHT, false, VARIANT_2>,
+ cryptonight_double_hash<CRYPTONIGHT, false, VARIANT_2>,
+ cryptonight_single_hash<CRYPTONIGHT, true, VARIANT_2>,
+ cryptonight_double_hash<CRYPTONIGHT, true, VARIANT_2>,
+ cryptonight_triple_hash<CRYPTONIGHT, false, VARIANT_2>,
+ cryptonight_quad_hash<CRYPTONIGHT, false, VARIANT_2>,
+ cryptonight_penta_hash<CRYPTONIGHT, false, VARIANT_2>,
+ cryptonight_triple_hash<CRYPTONIGHT, true, VARIANT_2>,
+ cryptonight_quad_hash<CRYPTONIGHT, true, VARIANT_2>,
+ cryptonight_penta_hash<CRYPTONIGHT, true, VARIANT_2>,
+
# ifndef XMRIG_NO_AEON
cryptonight_single_hash<CRYPTONIGHT_LITE, false, VARIANT_0>,
cryptonight_double_hash<CRYPTONIGHT_LITE, false, VARIANT_0>,
@@ -164,6 +175,7 @@ xmrig::CpuThread::cn_hash_fun xmrig::CpuThread::fn(Algo algorithm, AlgoVariant a
nullptr, nullptr, nullptr, nullptr, nullptr, nullptr, nullptr, nullptr, nullptr, nullptr, // VARIANT_XHV
nullptr, nullptr, nullptr, nullptr, nullptr, nullptr, nullptr, nullptr, nullptr, nullptr, // VARIANT_XAO
nullptr, nullptr, nullptr, nullptr, nullptr, nullptr, nullptr, nullptr, nullptr, nullptr, // VARIANT_RTO
+ nullptr, nullptr, nullptr, nullptr, nullptr, nullptr, nullptr, nullptr, nullptr, nullptr, // VARIANT_2
# else
nullptr, nullptr, nullptr, nullptr, nullptr, nullptr, nullptr, nullptr, nullptr, nullptr,
nullptr, nullptr, nullptr, nullptr, nullptr, nullptr, nullptr, nullptr, nullptr, nullptr,
@@ -173,6 +185,7 @@ xmrig::CpuThread::cn_hash_fun xmrig::CpuThread::fn(Algo algorithm, AlgoVariant a
nullptr, nullptr, nullptr, nullptr, nullptr, nullptr, nullptr, nullptr, nullptr, nullptr,
nullptr, nullptr, nullptr, nullptr, nullptr, nullptr, nullptr, nullptr, nullptr, nullptr,
nullptr, nullptr, nullptr, nullptr, nullptr, nullptr, nullptr, nullptr, nullptr, nullptr,
+ nullptr, nullptr, nullptr, nullptr, nullptr, nullptr, nullptr, nullptr, nullptr, nullptr,
# endif
# ifndef XMRIG_NO_SUMO
@@ -216,6 +229,7 @@ xmrig::CpuThread::cn_hash_fun xmrig::CpuThread::fn(Algo algorithm, AlgoVariant a
nullptr, nullptr, nullptr, nullptr, nullptr, nullptr, nullptr, nullptr, nullptr, nullptr, // VARIANT_XAO
nullptr, nullptr, nullptr, nullptr, nullptr, nullptr, nullptr, nullptr, nullptr, nullptr, // VARIANT_RTO
+ nullptr, nullptr, nullptr, nullptr, nullptr, nullptr, nullptr, nullptr, nullptr, nullptr, // VARIANT_2
# else
nullptr, nullptr, nullptr, nullptr, nullptr, nullptr, nullptr, nullptr, nullptr, nullptr,
nullptr, nullptr, nullptr, nullptr, nullptr, nullptr, nullptr, nullptr, nullptr, nullptr,
@@ -225,6 +239,7 @@ xmrig::CpuThread::cn_hash_fun xmrig::CpuThread::fn(Algo algorithm, AlgoVariant a
nullptr, nullptr, nullptr, nullptr, nullptr, nullptr, nullptr, nullptr, nullptr, nullptr,
nullptr, nullptr, nullptr, nullptr, nullptr, nullptr, nullptr, nullptr, nullptr, nullptr,
nullptr, nullptr, nullptr, nullptr, nullptr, nullptr, nullptr, nullptr, nullptr, nullptr,
+ nullptr, nullptr, nullptr, nullptr, nullptr, nullptr, nullptr, nullptr, nullptr, nullptr,
# endif
};
diff --git a/src/workers/MultiWorker.cpp b/src/workers/MultiWorker.cpp
index 5d43875c4..475f99bee 100644
--- a/src/workers/MultiWorker.cpp
+++ b/src/workers/MultiWorker.cpp
@@ -55,6 +55,7 @@ bool MultiWorker<N>::selfTest()
if (m_thread->algorithm() == CRYPTONIGHT) {
return verify(VARIANT_0, test_output_v0) &&
verify(VARIANT_1, test_output_v1) &&
+ verify(VARIANT_2, test_output_v2) &&
verify(VARIANT_XTL, test_output_xtl) &&
verify(VARIANT_MSR, test_output_msr) &&
verify(VARIANT_XAO, test_output_xao) &&