Skip to content

Commit

Permalink
CryptoNight v8 ReverseWaltz
Browse files Browse the repository at this point in the history
  • Loading branch information
@SChernykh
SChernykh committed Mar 4, 2019
1 parent 4ee84fe commit 7f80713
Show file tree
Hide file tree
Showing 38 changed files with 3,309 additions and 3,860 deletions.
4 changes: 2 additions & 2 deletions cmake/asm.cmake
View file
Original file line number Diff line number Diff line change
Expand Up @@ -23,7 +23,7 @@ if (WITH_ASM AND NOT XMRIG_ARM AND CMAKE_SIZEOF_VOID_P EQUAL 8)
if (WIN32 AND CMAKE_C_COMPILER_ID MATCHES GNU)
set(XMRIG_ASM_FILES
"src/crypto/asm/win64/cn_main_loop.S"
"src/crypto/asm/win64/CryptonightR_template.S"
"src/crypto/asm/CryptonightR_template.S"
)
else()
set(XMRIG_ASM_FILES
Expand All @@ -36,7 +36,7 @@ if (WITH_ASM AND NOT XMRIG_ARM AND CMAKE_SIZEOF_VOID_P EQUAL 8)
endif()

add_library(${XMRIG_ASM_LIBRARY} STATIC ${XMRIG_ASM_FILES})
set(XMRIG_ASM_SOURCES "")
set(XMRIG_ASM_SOURCES src/crypto/CryptonightR_gen.cpp)
set_property(TARGET ${XMRIG_ASM_LIBRARY} PROPERTY LINKER_LANGUAGE C)
else()
set(XMRIG_ASM_SOURCES "")
Expand Down
2 changes: 2 additions & 0 deletions src/base/net/Pool.cpp
View file
Original file line number Diff line number Diff line change
Expand Up @@ -132,6 +132,7 @@ xmrig::Pool::Pool(const rapidjson::Value &object) :


xmrig::Pool::Pool(const char *host, uint16_t port, const char *user, const char *password, int keepAlive, bool nicehash, bool tls) :
m_enabled(true),
m_nicehash(nicehash),
m_tls(tls),
m_keepAlive(keepAlive),
Expand Down Expand Up @@ -483,6 +484,7 @@ void xmrig::Pool::rebuild()
m_algorithms.push_back(m_algorithm);

# ifndef XMRIG_PROXY_PROJECT
addVariant(VARIANT_RWZ);
addVariant(VARIANT_4);
addVariant(VARIANT_WOW);
addVariant(VARIANT_2);
Expand Down
2 changes: 2 additions & 0 deletions src/common/crypto/Algorithm.cpp
View file
Original file line number Diff line number Diff line change
Expand Up @@ -66,6 +66,7 @@ static AlgoData const algorithms[] = {
{ "cryptonight/xtlv9", "cn/xtlv9", xmrig::CRYPTONIGHT, xmrig::VARIANT_HALF },
{ "cryptonight/wow", "cn/wow", xmrig::CRYPTONIGHT, xmrig::VARIANT_WOW },
{ "cryptonight/r", "cn/r", xmrig::CRYPTONIGHT, xmrig::VARIANT_4 },
{ "cryptonight/rwz", "cn/rwz", xmrig::CRYPTONIGHT, xmrig::VARIANT_RWZ },

# ifndef XMRIG_NO_AEON
{ "cryptonight-lite", "cn-lite", xmrig::CRYPTONIGHT_LITE, xmrig::VARIANT_AUTO },
Expand Down Expand Up @@ -133,6 +134,7 @@ static const char *variants[] = {
"gpu",
"wow",
"r",
"rwz"
};


Expand Down
1 change: 1 addition & 0 deletions src/common/xmrig.h
View file
Original file line number Diff line number Diff line change
Expand Up @@ -76,6 +76,7 @@ enum Variant {
VARIANT_GPU = 11, // CryptoNight-GPU (Ryo)
VARIANT_WOW = 12, // CryptoNightR (Wownero)
VARIANT_4 = 13, // CryptoNightR (Monero's variant 4)
VARIANT_RWZ = 14, // CryptoNight variant 2 with 3/4 iterations and reversed shuffle operation (Graft)
VARIANT_MAX
};

Expand Down
95 changes: 37 additions & 58 deletions src/crypto/CryptoNight.cpp
View file
Original file line number Diff line number Diff line change
Expand Up @@ -205,6 +205,13 @@ CryptoNight::cn_hash_fun CryptoNight::fn(xmrig::Algo algorithm, xmrig::AlgoVerif
# endif
cryptonight_single_hash<CRYPTONIGHT, true, VARIANT_4>,

# ifdef XMRIG_NO_ASM
cryptonight_single_hash<CRYPTONIGHT, false, VARIANT_RWZ>,
# else
cryptonight_single_hash_asm<CRYPTONIGHT, VARIANT_RWZ, ASM_AUTO>,
# endif
cryptonight_single_hash<CRYPTONIGHT, true, VARIANT_RWZ>,

# ifndef XMRIG_NO_AEON
cryptonight_single_hash<CRYPTONIGHT_LITE, false, VARIANT_0>,
cryptonight_single_hash<CRYPTONIGHT_LITE, true, VARIANT_0>,
Expand All @@ -224,6 +231,7 @@ CryptoNight::cn_hash_fun CryptoNight::fn(xmrig::Algo algorithm, xmrig::AlgoVerif
nullptr, nullptr, // VARIANT_GPU
nullptr, nullptr, // VARIANT_WOW
nullptr, nullptr, // VARIANT_4
nullptr, nullptr, // VARIANT_RWZ
# else
nullptr, nullptr, nullptr, nullptr,
nullptr, nullptr, nullptr, nullptr,
Expand All @@ -232,6 +240,7 @@ CryptoNight::cn_hash_fun CryptoNight::fn(xmrig::Algo algorithm, xmrig::AlgoVerif
nullptr, nullptr, nullptr, nullptr,
nullptr, nullptr, nullptr, nullptr,
nullptr, nullptr, nullptr, nullptr,
nullptr, nullptr,
# endif

# ifndef XMRIG_NO_SUMO
Expand All @@ -257,6 +266,7 @@ CryptoNight::cn_hash_fun CryptoNight::fn(xmrig::Algo algorithm, xmrig::AlgoVerif
nullptr, nullptr, // VARIANT_GPU
nullptr, nullptr, // VARIANT_WOW
nullptr, nullptr, // VARIANT_4
nullptr, nullptr, // VARIANT_RWZ
# else
nullptr, nullptr, nullptr, nullptr,
nullptr, nullptr, nullptr, nullptr,
Expand All @@ -265,6 +275,7 @@ CryptoNight::cn_hash_fun CryptoNight::fn(xmrig::Algo algorithm, xmrig::AlgoVerif
nullptr, nullptr, nullptr, nullptr,
nullptr, nullptr, nullptr, nullptr,
nullptr, nullptr, nullptr, nullptr,
nullptr, nullptr,
# endif
# ifndef XMRIG_NO_CN_PICO
nullptr, nullptr, // VARIANT_0
Expand All @@ -288,6 +299,7 @@ CryptoNight::cn_hash_fun CryptoNight::fn(xmrig::Algo algorithm, xmrig::AlgoVerif
nullptr, nullptr, // VARIANT_GPU
nullptr, nullptr, // VARIANT_WOW
nullptr, nullptr, // VARIANT_4
nullptr, nullptr, // VARIANT_RWZ
#else
nullptr, nullptr, nullptr, nullptr,
nullptr, nullptr, nullptr, nullptr,
Expand All @@ -296,6 +308,7 @@ CryptoNight::cn_hash_fun CryptoNight::fn(xmrig::Algo algorithm, xmrig::AlgoVerif
nullptr, nullptr, nullptr, nullptr,
nullptr, nullptr, nullptr, nullptr,
nullptr, nullptr, nullptr, nullptr,
nullptr, nullptr,
# endif
};

Expand All @@ -322,27 +335,27 @@ bool CryptoNight::selfTest() {
Mem::create(&m_ctx, m_algorithm, 1);

if (m_algorithm == xmrig::CRYPTONIGHT) {
if (!verify2(VARIANT_WOW, test_input_WOW)) {
LOG_WARN("CryptonightR (Wownero) self-test failed");
return false;
}
const bool rc = 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) &&
verify(VARIANT_RTO, test_output_rto) &&
verify(VARIANT_HALF, test_output_half) &&
verify2(VARIANT_WOW, test_output_wow) &&
verify2(VARIANT_4, test_output_r) &&
verify(VARIANT_RWZ, test_output_rwz);

if (!verify2(VARIANT_4, test_input_R)) {
LOG_WARN("CryptonightR self-test failed");
return false;
# ifndef XMRIG_NO_CN_GPU
if (!rc) {
return rc;
}

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) &&
verify(VARIANT_RTO, test_output_rto) &&
# ifndef XMRIG_NO_CN_GPU
verify(VARIANT_GPU, test_output_gpu) &&
# endif
verify(VARIANT_HALF, test_output_half);
return verify(VARIANT_GPU, test_output_gpu);
# else
return rc;
# endif
}

# ifndef XMRIG_NO_AEON
Expand Down Expand Up @@ -388,55 +401,21 @@ bool CryptoNight::verify(xmrig::Variant variant, const uint8_t *referenceValue)
return memcmp(output, referenceValue, 32) == 0;
}

bool CryptoNight::verify2(xmrig::Variant variant, const char *test_data)
bool CryptoNight::verify2(xmrig::Variant variant, const uint8_t *referenceValue)
{
cn_hash_fun func = fn(variant);
if (!func) {
return false;
}

std::stringstream s(test_data);
std::string expected_hex;
std::string input_hex;
uint64_t height;
while (!s.eof())
{
uint8_t referenceValue[32];
uint8_t input[256];

s >> expected_hex;
s >> input_hex;
s >> height;

if ((expected_hex.length() != 64) || (input_hex.length() > 512))
{
return false;
}

bool err = false;

for (int i = 0; i < 32; ++i)
{
referenceValue[i] = (hf_hex2bin(expected_hex[i * 2], err) << 4) + hf_hex2bin(expected_hex[i * 2 + 1], err);
}

const size_t input_len = input_hex.length() / 2;
for (size_t i = 0; i < input_len; ++i)
{
input[i] = (hf_hex2bin(input_hex[i * 2], err) << 4) + hf_hex2bin(input_hex[i * 2 + 1], err);
}

if (err)
{
return false;
}

for (size_t i = 0; i < (sizeof(cn_r_test_input) / sizeof(cn_r_test_input[0])); ++i) {
uint8_t hash[32];
func(input, input_len, hash, &m_ctx, height);
if (memcmp(hash, referenceValue, sizeof(hash)) != 0)
{
func(cn_r_test_input[i].data, cn_r_test_input[i].size, hash, &m_ctx, cn_r_test_input[i].height);

if (memcmp(hash, referenceValue + i * 32, sizeof hash) != 0) {
return false;
}
}

return true;
}
6 changes: 5 additions & 1 deletion src/crypto/CryptoNight.h
View file
Original file line number Diff line number Diff line change
Expand Up @@ -65,6 +65,10 @@ struct cryptonight_r_data {
struct cryptonight_ctx {
alignas(16) uint8_t state[224];
alignas(16) uint8_t *memory;

uint8_t unused[40];
const uint32_t* saes_table;

cn_mainloop_fun_ms_abi generated_code;
cn_mainloop_double_fun_ms_abi generated_code_double;
cryptonight_r_data generated_code_data;
Expand All @@ -86,7 +90,7 @@ class CryptoNight
private:
static bool selfTest();
static bool verify(xmrig::Variant variant, const uint8_t *referenceValue);
static bool verify2(xmrig::Variant variant, const char *test_data);
static bool verify2(xmrig::Variant variant, const uint8_t *test_data);

alignas(16) static cryptonight_ctx *m_ctx;
static xmrig::Algo m_algorithm;
Expand Down
10 changes: 8 additions & 2 deletions src/crypto/CryptoNight_constants.h
View file
Original file line number Diff line number Diff line change
Expand Up @@ -42,6 +42,7 @@ constexpr const uint32_t CRYPTONIGHT_MASK = 0x1FFFF0;
constexpr const uint32_t CRYPTONIGHT_ITER = 0x80000;
constexpr const uint32_t CRYPTONIGHT_HALF_ITER = 0x40000;
constexpr const uint32_t CRYPTONIGHT_XAO_ITER = 0x100000;
constexpr const uint32_t CRYPTONIGHT_WALTZ_ITER = 0x60000;

constexpr const uint32_t CRYPTONIGHT_GPU_ITER = 0xC000;
constexpr const uint32_t CRYPTONIGHT_GPU_MASK = 0x1FFFC0;
Expand Down Expand Up @@ -134,6 +135,7 @@ template<> inline constexpr uint32_t cn_select_iter<CRYPTONIGHT, VARIANT_MSR>()
template<> inline constexpr uint32_t cn_select_iter<CRYPTONIGHT, VARIANT_XAO>() { return CRYPTONIGHT_XAO_ITER; }
template<> inline constexpr uint32_t cn_select_iter<CRYPTONIGHT, VARIANT_RTO>() { return CRYPTONIGHT_ITER; }
template<> inline constexpr uint32_t cn_select_iter<CRYPTONIGHT, VARIANT_GPU>() { return CRYPTONIGHT_GPU_ITER; }
template<> inline constexpr uint32_t cn_select_iter<CRYPTONIGHT, VARIANT_RWZ>() { return CRYPTONIGHT_WALTZ_ITER; }
template<> inline constexpr uint32_t cn_select_iter<CRYPTONIGHT_LITE, VARIANT_0>() { return CRYPTONIGHT_LITE_ITER; }
template<> inline constexpr uint32_t cn_select_iter<CRYPTONIGHT_LITE, VARIANT_1>() { return CRYPTONIGHT_LITE_ITER; }
template<> inline constexpr uint32_t cn_select_iter<CRYPTONIGHT_HEAVY, VARIANT_0>() { return CRYPTONIGHT_HEAVY_ITER; }
Expand All @@ -158,6 +160,9 @@ inline uint32_t cn_select_iter(Algo algorithm, Variant variant)
case VARIANT_TRTL:
return CRYPTONIGHT_TRTL_ITER;

case VARIANT_RWZ:
return CRYPTONIGHT_WALTZ_ITER;

default:
break;
}
Expand Down Expand Up @@ -199,11 +204,12 @@ template<> inline constexpr Variant cn_base_variant<VARIANT_TRTL>() { return VA
template<> inline constexpr Variant cn_base_variant<VARIANT_GPU>() { return VARIANT_GPU; }
template<> inline constexpr Variant cn_base_variant<VARIANT_WOW>() { return VARIANT_2; }
template<> inline constexpr Variant cn_base_variant<VARIANT_4>() { return VARIANT_2; }
template<> inline constexpr Variant cn_base_variant<VARIANT_RWZ>() { return VARIANT_2; }


template<Variant variant> inline constexpr bool cn_is_cryptonight_r() { return false; }
template<> inline constexpr bool cn_is_cryptonight_r<VARIANT_WOW>() { return true; }
template<> inline constexpr bool cn_is_cryptonight_r<VARIANT_4>() { return true; }
template<> inline constexpr bool cn_is_cryptonight_r<VARIANT_WOW>() { return true; }
template<> inline constexpr bool cn_is_cryptonight_r<VARIANT_4>() { return true; }

} /* namespace xmrig */

Expand Down
44 changes: 32 additions & 12 deletions src/crypto/CryptoNight_monero.h
View file
Original file line number Diff line number Diff line change
Expand Up @@ -83,11 +83,11 @@
sqrt_result_xmm_##part = int_sqrt_v2(cx_0 + division_result); \
} while (0)

# define VARIANT2_SHUFFLE(base_ptr, offset, _a, _b, _b1, _c) \
# define VARIANT2_SHUFFLE(base_ptr, offset, _a, _b, _b1, _c, reverse) \
do { \
const __m128i chunk1 = _mm_load_si128((__m128i *)((base_ptr) + ((offset) ^ 0x10))); \
const __m128i chunk1 = _mm_load_si128((__m128i *)((base_ptr) + ((offset) ^ (reverse ? 0x30 : 0x10)))); \
const __m128i chunk2 = _mm_load_si128((__m128i *)((base_ptr) + ((offset) ^ 0x20))); \
const __m128i chunk3 = _mm_load_si128((__m128i *)((base_ptr) + ((offset) ^ 0x30))); \
const __m128i chunk3 = _mm_load_si128((__m128i *)((base_ptr) + ((offset) ^ (reverse ? 0x10 : 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)); \
Expand All @@ -96,15 +96,20 @@
} \
} while (0)

# define VARIANT2_SHUFFLE2(base_ptr, offset, _a, _b, _b1, hi, lo) \
# define VARIANT2_SHUFFLE2(base_ptr, offset, _a, _b, _b1, hi, lo, reverse) \
do { \
const __m128i chunk1 = _mm_xor_si128(_mm_load_si128((__m128i *)((base_ptr) + ((offset) ^ 0x10))), _mm_set_epi64x(lo, hi)); \
const __m128i chunk2 = _mm_load_si128((__m128i *)((base_ptr) + ((offset) ^ 0x20))); \
hi ^= ((uint64_t*)((base_ptr) + ((offset) ^ 0x20)))[0]; \
lo ^= ((uint64_t*)((base_ptr) + ((offset) ^ 0x20)))[1]; \
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)); \
if (reverse) { \
_mm_store_si128((__m128i *)((base_ptr) + ((offset) ^ 0x10)), _mm_add_epi64(chunk1, _b1)); \
_mm_store_si128((__m128i *)((base_ptr) + ((offset) ^ 0x20)), _mm_add_epi64(chunk3, _b)); \
} else { \
_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)

Expand All @@ -128,11 +133,11 @@
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, _c) \
# define VARIANT2_SHUFFLE(base_ptr, offset, _a, _b, _b1, _c, reverse) \
do { \
const uint64x2_t chunk1 = vld1q_u64((uint64_t*)((base_ptr) + ((offset) ^ 0x10))); \
const uint64x2_t chunk1 = vld1q_u64((uint64_t*)((base_ptr) + ((offset) ^ (reverse ? 0x30 : 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))); \
const uint64x2_t chunk3 = vld1q_u64((uint64_t*)((base_ptr) + ((offset) ^ (reverse ? 0x10 : 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))); \
Expand All @@ -141,15 +146,20 @@
} \
} while (0)

# define VARIANT2_SHUFFLE2(base_ptr, offset, _a, _b, _b1, hi, lo) \
# define VARIANT2_SHUFFLE2(base_ptr, offset, _a, _b, _b1, hi, lo, reverse) \
do { \
const uint64x2_t chunk1 = veorq_u64(vld1q_u64((uint64_t*)((base_ptr) + ((offset) ^ 0x10))), vcombine_u64(vcreate_u64(hi), vcreate_u64(lo))); \
const uint64x2_t chunk2 = vld1q_u64((uint64_t*)((base_ptr) + ((offset) ^ 0x20))); \
hi ^= ((uint64_t*)((base_ptr) + ((offset) ^ 0x20)))[0]; \
lo ^= ((uint64_t*)((base_ptr) + ((offset) ^ 0x20)))[1]; \
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))); \
if (reverse) { \
vst1q_u64((uint64_t*)((base_ptr) + ((offset) ^ 0x10)), vaddq_u64(chunk1, vreinterpretq_u64_u8(_b1))); \
vst1q_u64((uint64_t*)((base_ptr) + ((offset) ^ 0x20)), vaddq_u64(chunk3, vreinterpretq_u64_u8(_b))); \
} else { \
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
Expand All @@ -158,6 +168,16 @@
#define SWAP64LE(x) x
#define hash_extra_blake(data, length, hash) blake256_hash((uint8_t*)(hash), (uint8_t*)(data), (length))

#ifndef NOINLINE
#ifdef __GNUC__
#define NOINLINE __attribute__ ((noinline))
#elif _MSC_VER
#define NOINLINE __declspec(noinline)
#else
#define NOINLINE
#endif
#endif

#include "common/xmrig.h"
#include "variant4_random_math.h"

Expand Down
Loading

0 comments on commit 7f80713

Please sign in to comment.