Skip to content
New issue

Have a question about this project? Sign up for a free GitHub account to open an issue and contact its maintainers and the community.

Sign up for GitHub

By clicking “Sign up for GitHub”, you agree to our terms of service and privacy statement. We’ll occasionally send you account related emails.

Already on GitHub? Sign in to your account

CryptoNight v8 ReverseWaltz #242

Merged
merged 1 commit into from
Mar 5, 2019
Merged

CryptoNight v8 ReverseWaltz #242

Show file tree
Hide file tree
Changes from all commits
Commits
Show all changes
1 commit
Select commit
File filter

Filter by extension

Filter by extension
Conversations
Failed to load comments.
Loading
Jump to
Jump to file
Failed to load files.
Loading
Diff view
Diff view
4 changes: 2 additions & 2 deletions cmake/asm.cmake
View file
Open in desktop
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
Open in desktop
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
Open in desktop
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
Open in desktop
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
Open in desktop
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
Open in desktop
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
Open in desktop
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
Open in desktop
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