diff --git a/ethash.go b/ethash.go index 21c10c87..4831b60e 100644 --- a/ethash.go +++ b/ethash.go @@ -103,8 +103,8 @@ func freeCache(cache *cache) { cache.ptr = nil } -func (cache *cache) compute(dagSize uint64, hash common.Hash, nonce uint64) (ok bool, mixDigest, result common.Hash) { - ret := C.ethash_light_compute_internal(cache.ptr, C.uint64_t(dagSize), hashToH256(hash), C.uint64_t(nonce)) +func (cache *cache) compute(dagSize uint64, hash common.Hash, nonce uint64, block_number uint64) (ok bool, mixDigest, result common.Hash) { + ret := C.progpow_light_compute_internal(cache.ptr, C.uint64_t(dagSize), hashToH256(hash), C.uint64_t(nonce), C.uint64_t(block_number)) // Make sure cache is live until after the C call. // This is important because a GC might happen and execute // the finalizer before the call completes. @@ -130,7 +130,7 @@ func (l *Light) Verify(block Block) bool { // to prevent DOS attacks. blockNum := block.NumberU64() if blockNum >= epochLength*2048 { - log.Debug(fmt.Sprintf("block number %d too high, limit is %d", epochLength*2048)) + log.Debug(fmt.Sprintf("block number %d too high, limit is %d", blockNum, epochLength*2048)) return false } @@ -151,7 +151,7 @@ func (l *Light) Verify(block Block) bool { dagSize = dagSizeForTesting } // Recompute the hash using the cache. - ok, mixDigest, result := cache.compute(uint64(dagSize), block.HashNoNonce(), block.Nonce()) + ok, mixDigest, result := cache.compute(uint64(dagSize), block.HashNoNonce(), block.Nonce(), blockNum) if !ok { return false } diff --git a/ethashc.go b/ethashc.go index 1d2ba161..63b51f41 100644 --- a/ethashc.go +++ b/ethashc.go @@ -33,6 +33,7 @@ package ethash #cgo LDFLAGS: -lm #include "src/libethash/internal.c" +#include "src/libethash/progpow-internal.c" #include "src/libethash/sha3.c" #include "src/libethash/io.c" diff --git a/src/libethash/CMakeLists.txt b/src/libethash/CMakeLists.txt index a65621c3..5fd19009 100644 --- a/src/libethash/CMakeLists.txt +++ b/src/libethash/CMakeLists.txt @@ -13,6 +13,7 @@ endif() set(FILES util.h io.c internal.c + progpow-internal.c ethash.h endian.h compiler.h diff --git a/src/libethash/ethash.h b/src/libethash/ethash.h index 0c6a1f9e..04e87395 100644 --- a/src/libethash/ethash.h +++ b/src/libethash/ethash.h @@ -40,10 +40,17 @@ #define ETHASH_DAG_MAGIC_NUM_SIZE 8 #define ETHASH_DAG_MAGIC_NUM 0xFEE1DEADBADDCAFE +#define PROGPOW_MIX_BYTES 256 + #ifdef __cplusplus extern "C" { #endif +typedef struct +{ + uint32_t uint32s[32 / sizeof(uint32_t)]; +} hash32_t; + /// Type of a seedhash/blockhash e.t.c. typedef struct ethash_h256 { uint8_t b[32]; } ethash_h256_t; @@ -128,6 +135,39 @@ ethash_return_value_t ethash_full_compute( ethash_h256_t const header_hash, uint64_t nonce ); + +/** + * Calculate the light client data of the ProgPow + * + * @param light The light client handler + * @param header_hash The header hash to pack into the mix + * @param nonce The nonce to pack into the mix + * @param block_number The block_number + * @return an object of ethash_return_value_t holding the return values + */ +ethash_return_value_t progpow_light_compute( + ethash_light_t light, + ethash_h256_t const header_hash, + uint64_t nonce, + uint64_t block_number +); + +/** + * Calculate the full client data of the ProgPoW + * + * @param full The full client handler + * @param header_hash The header hash to pack into the mix + * @param nonce The nonce to pack into the mix + * @param block_number The current block_number + * @return An object of ethash_return_value to hold the return value + */ +ethash_return_value_t progpow_full_compute( + ethash_full_t full, + ethash_h256_t const header_hash, + uint64_t nonce, + uint64_t block_number +); + /** * Get a pointer to the full DAG data */ diff --git a/src/libethash/internal.c b/src/libethash/internal.c index 0a830fc8..0e69f009 100644 --- a/src/libethash/internal.c +++ b/src/libethash/internal.c @@ -56,7 +56,7 @@ uint64_t ethash_get_cachesize(uint64_t const block_number) // Follows Sergio's "STRICT MEMORY HARD HASHING FUNCTIONS" (2014) // https://bitslog.files.wordpress.com/2013/12/memohash-v0-3.pdf // SeqMemoHash(s, R, N) -bool static ethash_compute_cache_nodes( +static bool ethash_compute_cache_nodes( node* const nodes, uint64_t cache_size, ethash_h256_t const* seed @@ -108,6 +108,9 @@ void ethash_calculate_dag_item( __m128i xmm1 = ret->xmm[1]; __m128i xmm2 = ret->xmm[2]; __m128i xmm3 = ret->xmm[3]; +#elif defined(__MIC__) + __m512i const fnv_prime = _mm512_set1_epi32(FNV_PRIME); + __m512i zmm0 = ret->zmm[0]; #endif for (uint32_t i = 0; i != ETHASH_DATASET_PARENTS; ++i) { @@ -131,6 +134,14 @@ void ethash_calculate_dag_item( ret->xmm[2] = xmm2; ret->xmm[3] = xmm3; } + #elif defined(__MIC__) + { + zmm0 = _mm512_mullo_epi32(zmm0, fnv_prime); + + // have to write to ret as values are used to compute index + zmm0 = _mm512_xor_si512(zmm0, parent->zmm[0]); + ret->zmm[0] = zmm0; + } #else { for (unsigned w = 0; w != NODE_WORDS; ++w) { @@ -207,10 +218,10 @@ static bool ethash_hash( for (unsigned n = 0; n != MIX_NODES; ++n) { node const* dag_node; + node tmp_node; if (full_nodes) { dag_node = &full_nodes[MIX_NODES * index + n]; } else { - node tmp_node; ethash_calculate_dag_item(&tmp_node, index * MIX_NODES + n, light); dag_node = &tmp_node; } @@ -227,6 +238,14 @@ static bool ethash_hash( mix[n].xmm[2] = _mm_xor_si128(xmm2, dag_node->xmm[2]); mix[n].xmm[3] = _mm_xor_si128(xmm3, dag_node->xmm[3]); } + #elif defined(__MIC__) + { + // __m512i implementation via union + // Each vector register (zmm) can store sixteen 32-bit integer numbers + __m512i fnv_prime = _mm512_set1_epi32(FNV_PRIME); + __m512i zmm0 = _mm512_mullo_epi32(fnv_prime, mix[n].zmm[0]); + mix[n].zmm[0] = _mm512_xor_si512(zmm0, dag_node->zmm[0]); + } #else { for (unsigned w = 0; w != NODE_WORDS; ++w) { @@ -238,6 +257,22 @@ static bool ethash_hash( } +// Workaround for a GCC regression which causes a bogus -Warray-bounds warning. +// The regression was introduced in GCC 4.8.4, fixed in GCC 5.0.0 and backported to GCC 4.9.3 but +// never to the GCC 4.8.x line. +// +// See https://gcc.gnu.org/bugzilla/show_bug.cgi?id=56273 +// +// This regression is affecting Debian Jesse (8.5) builds of cpp-ethereum (GCC 4.9.2) and also +// manifests in the doublethinkco armel v5 cross-builds, which use crosstool-ng and resulting +// in the use of GCC 4.8.4. The Tizen runtime wants an even older GLIBC version - the one from +// GCC 4.6.0! + +#if defined(__GNUC__) && (__GNUC__ < 5) +#pragma GCC diagnostic push +#pragma GCC diagnostic ignored "-Warray-bounds" +#endif // define (__GNUC__) + // compress mix for (uint32_t w = 0; w != MIX_WORDS; w += 4) { uint32_t reduction = mix->words[w + 0]; @@ -247,6 +282,10 @@ static bool ethash_hash( mix->words[w / 4] = reduction; } +#if defined(__GNUC__) && (__GNUC__ < 5) +#pragma GCC diagnostic pop +#endif // define (__GNUC__) + fix_endian_arr32(mix->words, MIX_WORDS / 4); memcpy(&ret->mix_hash, mix->bytes, 32); // final Keccak hash @@ -300,7 +339,11 @@ ethash_light_t ethash_light_new_internal(uint64_t cache_size, ethash_h256_t cons if (!ret) { return NULL; } +#if defined(__MIC__) + ret->cache = _mm_malloc((size_t)cache_size, 64); +#else ret->cache = malloc((size_t)cache_size); +#endif if (!ret->cache) { goto fail_free_light; } @@ -312,7 +355,11 @@ ethash_light_t ethash_light_new_internal(uint64_t cache_size, ethash_h256_t cons return ret; fail_free_cache_mem: +#if defined(__MIC__) + _mm_free(ret->cache); +#else free(ret->cache); +#endif fail_free_light: free(ret); return NULL; @@ -399,31 +446,47 @@ ethash_full_t ethash_full_new_internal( return NULL; } ret->file_size = (size_t)full_size; - switch (ethash_io_prepare(dirname, seed_hash, &f, (size_t)full_size, false)) { - case ETHASH_IO_FAIL: - // ethash_io_prepare will do all ETHASH_CRITICAL() logging in fail case + + enum ethash_io_rc err = ethash_io_prepare(dirname, seed_hash, &f, (size_t)full_size, false); + if (err == ETHASH_IO_FAIL) goto fail_free_full; - case ETHASH_IO_MEMO_MATCH: - if (!ethash_mmap(ret, f)) { - ETHASH_CRITICAL("mmap failure()"); - goto fail_close_file; - } - return ret; - case ETHASH_IO_MEMO_SIZE_MISMATCH: + + if (err == ETHASH_IO_MEMO_SIZE_MISMATCH) { // if a DAG of same filename but unexpected size is found, silently force new file creation if (ethash_io_prepare(dirname, seed_hash, &f, (size_t)full_size, true) != ETHASH_IO_MEMO_MISMATCH) { ETHASH_CRITICAL("Could not recreate DAG file after finding existing DAG with unexpected size."); goto fail_free_full; } - // fallthrough to the mismatch case here, DO NOT go through match - case ETHASH_IO_MEMO_MISMATCH: + // we now need to go through the mismatch case, NOT the match case + err = ETHASH_IO_MEMO_MISMATCH; + } + + if (err == ETHASH_IO_MEMO_MISMATCH || err == ETHASH_IO_MEMO_MATCH) { if (!ethash_mmap(ret, f)) { ETHASH_CRITICAL("mmap failure()"); goto fail_close_file; } - break; + + if (err == ETHASH_IO_MEMO_MATCH) { +#if defined(__MIC__) + node* tmp_nodes = _mm_malloc((size_t)full_size, 64); + //copy all nodes from ret->data + //mmapped_nodes are not aligned properly + uint32_t const countnodes = (uint32_t) ((size_t)ret->file_size / sizeof(node)); + //fprintf(stderr,"ethash_full_new_internal:countnodes:%d",countnodes); + for (uint32_t i = 1; i != countnodes; ++i) { + tmp_nodes[i] = ret->data[i]; + } + ret->data = tmp_nodes; +#endif + return ret; + } } + +#if defined(__MIC__) + ret->data = _mm_malloc((size_t)full_size, 64); +#endif if (!ethash_compute_full_data(ret->data, full_size, light, callback)) { ETHASH_CRITICAL("Failure at computing DAG data."); goto fail_free_full_data; @@ -448,6 +511,9 @@ ethash_full_t ethash_full_new_internal( fail_free_full_data: // could check that munmap(..) == 0 but even if it did not can't really do anything here munmap(ret->data, (size_t)full_size); +#if defined(__MIC__) + _mm_free(ret->data); +#endif fail_close_file: fclose(ret->file); fail_free_full: diff --git a/src/libethash/internal.h b/src/libethash/internal.h index 26c395ad..6f045dc0 100644 --- a/src/libethash/internal.h +++ b/src/libethash/internal.h @@ -8,6 +8,8 @@ #if defined(_M_X64) && ENABLE_SSE #include +#elif defined(__MIC__) +#include #endif #ifdef __cplusplus @@ -27,6 +29,8 @@ typedef union node { #if defined(_M_X64) && ENABLE_SSE __m128i xmm[NODE_WORDS/4]; +#elif defined(__MIC__) + __m512i zmm[NODE_WORDS/16]; #endif } node; @@ -102,6 +106,7 @@ ethash_light_t ethash_light_new_internal(uint64_t cache_size, ethash_h256_t cons * @param full_size The size of the full data in bytes. * @param header_hash The header hash to pack into the mix * @param nonce The nonce to pack into the mix + * @param block_number The block_number * @return The resulting hash. */ ethash_return_value_t ethash_light_compute_internal( @@ -111,6 +116,29 @@ ethash_return_value_t ethash_light_compute_internal( uint64_t nonce ); +void keccak_f800_round(uint32_t st[25], const int r); +hash32_t keccak_f800_progpow(hash32_t header, uint64_t seed, hash32_t digest); +uint32_t progpowMath(uint32_t a, uint32_t b, uint32_t r); +void merge(uint32_t *a, uint32_t b, uint32_t r); + +/** + * Calculate the light client data of the ProgPow. Internal version. + * + * @param light The light client handler + * @param full_size The size of the full data in bytes. + * @param header_hash The header hash to pack into the mix + * @param nonce The nonce to pack into the mix + * @param block_number The block_number + * @return The resulting hash. + */ +ethash_return_value_t progpow_light_compute_internal( + ethash_light_t light, + uint64_t full_size, + ethash_h256_t const header_hash, + uint64_t nonce, + uint64_t block_number +); + struct ethash_full { FILE* file; uint64_t file_size; diff --git a/src/libethash/io.c b/src/libethash/io.c index f4db477c..ffb1a0e7 100644 --- a/src/libethash/io.c +++ b/src/libethash/io.c @@ -91,7 +91,7 @@ enum ethash_io_rc ethash_io_prepare( goto free_memo; } // make sure it's of the proper size - if (fseek(f, (long int)(file_size + ETHASH_DAG_MAGIC_NUM_SIZE - 1), SEEK_SET) != 0) { + if (ethash_fseek(f, file_size + ETHASH_DAG_MAGIC_NUM_SIZE - 1, SEEK_SET) != 0) { fclose(f); ETHASH_CRITICAL("Could not seek to the end of DAG file: \"%s\". Insufficient space?", tmpfile); goto free_memo; diff --git a/src/libethash/io.h b/src/libethash/io.h index 7a27089c..06bd8c3b 100644 --- a/src/libethash/io.h +++ b/src/libethash/io.h @@ -113,6 +113,16 @@ enum ethash_io_rc ethash_io_prepare( */ FILE* ethash_fopen(char const* file_name, char const* mode); +/** + * An fseek wrapper for crossplatform 64-bit seek. + * + * @param f The file stream whose fd to get + * @param offset Number of bytes from @a origin + * @param origin Initial position + * @return Current offset or -1 to indicate an error + */ +int ethash_fseek(FILE* f, size_t offset, int origin); + /** * An strncat wrapper for no-warnings crossplatform strncat. * diff --git a/src/libethash/io_posix.c b/src/libethash/io_posix.c index c9a17d84..b831acf1 100644 --- a/src/libethash/io_posix.c +++ b/src/libethash/io_posix.c @@ -34,6 +34,11 @@ FILE* ethash_fopen(char const* file_name, char const* mode) return fopen(file_name, mode); } +int ethash_fseek(FILE* f, size_t offset, int origin) +{ + return fseeko(f, offset, origin); +} + char* ethash_strncat(char* dest, size_t dest_size, char const* src, size_t count) { return strlen(dest) + count + 1 <= dest_size ? strncat(dest, src, count) : NULL; @@ -96,6 +101,8 @@ bool ethash_get_default_dirname(char* strbuf, size_t buffsize) struct passwd* pwd = getpwuid(getuid()); if (pwd) home_dir = pwd->pw_dir; + if (!home_dir) + return false; } size_t len = strlen(home_dir); diff --git a/src/libethash/io_win32.c b/src/libethash/io_win32.c index 34f1aaa7..2fb0e6c9 100644 --- a/src/libethash/io_win32.c +++ b/src/libethash/io_win32.c @@ -26,6 +26,7 @@ #include #include #include +#include FILE* ethash_fopen(char const* file_name, char const* mode) { @@ -33,6 +34,11 @@ FILE* ethash_fopen(char const* file_name, char const* mode) return fopen_s(&f, file_name, mode) == 0 ? f : NULL; } +int ethash_fseek(FILE* f, size_t offset, int origin) +{ + return _fseeki64(f, offset, origin); +} + char* ethash_strncat(char* dest, size_t dest_size, char const* src, size_t count) { return strncat_s(dest, dest_size, src, count) == 0 ? dest : NULL; diff --git a/src/libethash/progpow-internal.c b/src/libethash/progpow-internal.c new file mode 100644 index 00000000..e9981877 --- /dev/null +++ b/src/libethash/progpow-internal.c @@ -0,0 +1,543 @@ +/* + This file is part of ethash. + + ethash is free software: you can redistribute it and/or modify + it under the terms of the GNU General Public License as published by + the Free Software Foundation, either version 3 of the License, or + (at your option) any later version. + + ethash is distributed in the hope that it will be useful, + but WITHOUT ANY WARRANTY; without even the implied warranty of + MERCHANTABILITY or FITNESS FOR A PARTICULAR PURPOSE. See the + GNU General Public License for more details. + + You should have received a copy of the GNU General Public License + along with ethash. If not, see . +*/ +/** @file progpow-internal.c + * @license CC0 + * @author ifdefelse + * @date 2018 + */ + +#include +#include +#include +#include +#include +#include "mmap.h" +#include "ethash.h" +#include "fnv.h" +#include "endian.h" +#include "internal.h" +#include "io.h" + +#ifdef WITH_CRYPTOPP + +#include "sha3_cryptopp.h" + +#else +#include "sha3.h" +#endif // WITH_CRYPTOPP + +#define PROGPOW_LANES 16 +#define PROGPOW_REGS 32 +#define PROGPOW_DAG_LOADS 4 +#define PROGPOW_CACHE_BYTES (16*1024) +#define PROGPOW_CNT_DAG ETHASH_ACCESSES +#define PROGPOW_CNT_CACHE 12 +#define PROGPOW_CNT_MATH 20 +#define PROGPOW_CACHE_WORDS (PROGPOW_CACHE_BYTES / sizeof(uint32_t)) +#define PROGPOW_PERIOD 50 + +#define ROTL(x,n,w) (((x) << (n % w)) | ((x) >> ((w) - (n % w)))) +#define ROTL32(x,n) ROTL(x,n,32) /* 32 bits word */ + +#define ROTR(x,n,w) (((x) >> (n % w)) | ((x) << ((w) - (n % w)))) +#define ROTR32(x,n) ROTR(x,n,32) /* 32 bits word */ + +#define min(a,b) ((a>32); +} + +#ifdef __GNUC__ +#define clz(a) (a ? (uint32_t)__builtin_clz(a) : 32) +#define popcount(a) ((uint32_t)__builtin_popcount(a)) +#elif _MSC_VER +#include +static inline uint32_t clz(uint32_t value) +{ + unsigned long leading_zero = 0; + + if (_BitScanReverse(&leading_zero, value)) { + return 31 - leading_zero; + } else { + // Same remarks as above + return 32; + } +} + +static inline uint32_t popcount(uint32_t value) +{ + return (uint32_t)__popcnt(value); +} +#else +static inline uint32_t clz(uint32_t a) +{ + uint32_t result = 0; + for (int i = 31; i >= 0; i--) { + if (((a>>i)&1) == 0) + result ++; + else + break; + } + return result; +} + +static inline uint32_t popcount(uint32_t a) +{ + uint32_t result = 0; + for (int i = 31; i >= 0; i--) { + if(((a>>i)&1) == 1) + result ++; + } + return result; +} +#endif + +static inline void swap(uint32_t *a, uint32_t *b) +{ + uint32_t t = *a; + *a = *b; + *b = t; +} + +static inline uint32_t fnv1a(uint32_t *h, uint32_t d) +{ + return *h = (*h ^ d) * (uint32_t)0x1000193; +} + +// Implementation based on: +// https://github.com/mjosaarinen/tiny_sha3/blob/master/sha3.c +const uint32_t keccakf_rndc[24] = { + 0x00000001, 0x00008082, 0x0000808a, 0x80008000, 0x0000808b, 0x80000001, + 0x80008081, 0x00008009, 0x0000008a, 0x00000088, 0x80008009, 0x8000000a, + 0x8000808b, 0x0000008b, 0x00008089, 0x00008003, 0x00008002, 0x00000080, + 0x0000800a, 0x8000000a, 0x80008081, 0x00008080, 0x80000001, 0x80008008 +}; + +// Implementation of the permutation Keccakf with width 800. +void keccak_f800_round(uint32_t st[25], const int r) +{ + const uint32_t keccakf_rotc[24] = { + 1, 3, 6, 10, 15, 21, 28, 36, 45, 55, 2, 14, + 27, 41, 56, 8, 25, 43, 62, 18, 39, 61, 20, 44 + }; + const uint32_t keccakf_piln[24] = { + 10, 7, 11, 17, 18, 3, 5, 16, 8, 21, 24, 4, + 15, 23, 19, 13, 12, 2, 20, 14, 22, 9, 6, 1 + }; + + uint32_t t, bc[5]; + // Theta + for (int i = 0; i < 5; i++) + bc[i] = st[i] ^ st[i + 5] ^ st[i + 10] ^ st[i + 15] ^ st[i + 20]; + + for (int i = 0; i < 5; i++) { + t = bc[(i + 4) % 5] ^ ROTL32(bc[(i + 1) % 5], 1); + for (uint32_t j = 0; j < 25; j += 5) + st[j + i] ^= t; + } + + // Rho Pi + t = st[1]; + for (int i = 0; i < 24; i++) { + uint32_t j = keccakf_piln[i]; + bc[0] = st[j]; + st[j] = ROTL32(t, keccakf_rotc[i]); + t = bc[0]; + } + + // Chi + for (uint32_t j = 0; j < 25; j += 5) { + for (int i = 0; i < 5; i++) + bc[i] = st[j + i]; + for (int i = 0; i < 5; i++) + st[j + i] ^= (~bc[(i + 1) % 5]) & bc[(i + 2) % 5]; + } + + // Iota + st[0] ^= keccakf_rndc[r]; +} + +// Implementation of the Keccak sponge construction (with padding omitted) +// The width is 800, with a bitrate of 576, and a capacity of 224. +hash32_t keccak_f800_progpow(hash32_t header, uint64_t seed, hash32_t digest) +{ + uint32_t st[25]; + + for (int i = 0; i < 25; i++) + st[i] = 0; + for (int i = 0; i < 8; i++) + st[i] = header.uint32s[i]; + st[8] = seed; + st[9] = seed >> 32; + for (int i = 0; i < 8; i++) + st[10+i] = digest.uint32s[i]; + + for (int r = 0; r < 21; r++) { + keccak_f800_round(st, r); + } + // last round can be simplified due to partial output + keccak_f800_round(st, 21); + + hash32_t ret; + for (int i = 0; i < 8; i++) { + ret.uint32s[i] = st[i]; + } + + return ret; +} + +typedef struct { + uint32_t z, w, jsr, jcong; +} kiss99_t; + +// KISS99 is simple, fast, and passes the TestU01 suite +// https://en.wikipedia.org/wiki/KISS_(algorithm) +// http://www.cse.yorku.ca/~oz/marsaglia-rng.html +uint32_t kiss99(kiss99_t * st) +{ + st->z = 36969 * (st->z & 65535) + (st->z >> 16); + st->w = 18000 * (st->w & 65535) + (st->w >> 16); + uint32_t MWC = ((st->z << 16) + st->w); + st->jsr ^= (st->jsr << 17); + st->jsr ^= (st->jsr >> 13); + st->jsr ^= (st->jsr << 5); + st->jcong = 69069 * st->jcong + 1234567; + return ((MWC^st->jcong) + st->jsr); +} + +void fill_mix( + uint64_t seed, + uint32_t lane_id, + uint32_t mix[PROGPOW_REGS] +) +{ + // Use FNV to expand the per-warp seed to per-lane + // Use KISS to expand the per-lane seed to fill mix + uint32_t fnv_hash = 0x811c9dc5; + kiss99_t st; + st.z = fnv1a(&fnv_hash, (uint32_t)seed); + st.w = fnv1a(&fnv_hash, seed >> 32); + st.jsr = fnv1a(&fnv_hash, lane_id); + st.jcong = fnv1a(&fnv_hash, lane_id); + for (int i = 0; i < PROGPOW_REGS; i++) + mix[i] = kiss99(&st); +} + +kiss99_t progPowInit(uint64_t prog_seed, uint32_t mix_seq_dst[PROGPOW_REGS], uint32_t mix_seq_src[PROGPOW_REGS]) +{ + kiss99_t prog_rnd; + uint32_t fnv_hash = 0x811c9dc5; + prog_rnd.z = fnv1a(&fnv_hash, (uint32_t)prog_seed); + prog_rnd.w = fnv1a(&fnv_hash, prog_seed >> 32); + prog_rnd.jsr = fnv1a(&fnv_hash, (uint32_t)prog_seed); + prog_rnd.jcong = fnv1a(&fnv_hash, prog_seed >> 32); + // Create a random sequence of mix destinations for merge() and mix sources for cache reads + // guaranteeing every location is touched once + // guarantees no duplicate cache reads, which could be optimized away + // Uses Fisher–Yates shuffle + for (int i = 0; i < PROGPOW_REGS; i++) + { + mix_seq_dst[i] = i; + mix_seq_src[i] = i; + } + for (int i = PROGPOW_REGS - 1; i > 0; i--) + { + int j; + j = kiss99(&prog_rnd) % (i + 1); + swap(&(mix_seq_dst[i]), &(mix_seq_dst[j])); + j = kiss99(&prog_rnd) % (i + 1); + swap(&(mix_seq_src[i]), &(mix_seq_src[j])); + } + return prog_rnd; +} + +// Merge new data from b into the value in a +// Assuming A has high entropy only do ops that retain entropy +// even if B is low entropy +// (IE don't do A&B) +void merge(uint32_t *a, uint32_t b, uint32_t r) +{ + switch (r % 4) + { + case 0: *a = (*a * 33) + b; break; + case 1: *a = (*a ^ b) * 33; break; + // prevent rotate by 0 which is a NOP + case 2: *a = ROTL32(*a, ((r >> 16) % 31)+1) ^ b; break; + case 3: *a = ROTR32(*a, ((r >> 16) % 31)+1) ^ b; break; + } +} + +// Random math between two input values +uint32_t progpowMath(uint32_t a, uint32_t b, uint32_t r) +{ + switch (r % 11) + { + default: + case 0: return a + b; break; + case 1: return a * b; break; + case 2: return mul_hi32(a, b); break; + case 3: return min(a, b); break; + case 4: return ROTL32(a, b); break; + case 5: return ROTR32(a, b); break; + case 6: return a & b; break; + case 7: return a | b; break; + case 8: return a ^ b; break; + case 9: return clz(a) + clz(b); break; + case 10: return popcount(a) + popcount(b); break; + } + return 0; +} + +void progPowLoop( + const uint64_t prog_seed, + const uint32_t loop, + ethash_light_t const light, + uint32_t mix[PROGPOW_LANES][PROGPOW_REGS], + const uint32_t *g_dag, + const uint32_t *c_dag, + const uint32_t dag_words) +{ + // All lanes share a base address for the global load + // Global offset uses mix[0] to guarantee it depends on the load result + uint32_t offset_g = mix[loop%PROGPOW_LANES][0] % (64 * dag_words / (PROGPOW_LANES*PROGPOW_DAG_LOADS)); + + // global load to sequential locations + uint32_t data_g[PROGPOW_LANES][PROGPOW_DAG_LOADS]; + uint32_t dag_data[PROGPOW_LANES*PROGPOW_DAG_LOADS]; + if (g_dag) { + for (int i = 0; i < PROGPOW_DAG_LOADS; i++) { + memcpy((void *)&dag_data[PROGPOW_LANES*i], + (void *)&g_dag[offset_g*PROGPOW_LANES*PROGPOW_DAG_LOADS + i*PROGPOW_LANES], + PROGPOW_LANES*sizeof(uint32_t)); + } + } else { + node tmp_node; + for (int i = 0; i < PROGPOW_DAG_LOADS; i++) { + uint64_t k = offset_g*PROGPOW_LANES*PROGPOW_DAG_LOADS + i*PROGPOW_LANES; + ethash_calculate_dag_item(&tmp_node, k / 16, light); + memcpy((void *)&dag_data[PROGPOW_LANES*i], + (void *)&tmp_node.words[0], + PROGPOW_LANES*sizeof(uint32_t)); + } + } + + //int max_i = max(PROGPOW_CNT_CACHE, PROGPOW_CNT_MATH); + int max_i; + if (PROGPOW_CNT_CACHE > PROGPOW_CNT_MATH) + max_i = PROGPOW_CNT_CACHE; + else + max_i = PROGPOW_CNT_MATH; + + // Initialize the program seed and sequences + // When mining these are evaluated on the CPU and compiled away + uint32_t mix_seq_dst[PROGPOW_REGS]; + uint32_t mix_seq_src[PROGPOW_REGS]; + uint32_t mix_seq_dst_cnt = 0; + uint32_t mix_seq_src_cnt = 0; + kiss99_t prog_rnd = progPowInit(prog_seed, mix_seq_dst, mix_seq_src); + + for (int i = 0; i < max_i; i++) + { + if (i < PROGPOW_CNT_CACHE) + { + // Cached memory access + // lanes access random 32-bit locations within the first portion of the DAG + int src = mix_seq_src[(mix_seq_src_cnt++)%PROGPOW_REGS]; + int dst = mix_seq_dst[(mix_seq_dst_cnt++)%PROGPOW_REGS]; + int sel = kiss99(&prog_rnd); + for (int l = 0; l < PROGPOW_LANES; l++) + { + uint32_t offset = mix[l][src] % PROGPOW_CACHE_WORDS; + merge(&(mix[l][dst]), c_dag[offset], sel); + } + } + if (i < PROGPOW_CNT_MATH) + { + // Random Math + // Generate 2 unique sources + int src_rnd = kiss99(&prog_rnd) % (PROGPOW_REGS * (PROGPOW_REGS-1)); + int src1 = src_rnd % PROGPOW_REGS; // 0 <= src1 < PROGPOW_REGS + int src2 = src_rnd / PROGPOW_REGS; // 0 <= src2 < PROGPOW_REGS - 1 + if (src2 >= src1) ++src2; // src2 is now any reg other than src1 + int sel1 = kiss99(&prog_rnd); + int dst = mix_seq_dst[(mix_seq_dst_cnt++)%PROGPOW_REGS]; + int sel2 = kiss99(&prog_rnd); + for (int l = 0; l < PROGPOW_LANES; l++) + { + uint32_t data = progpowMath(mix[l][src1], mix[l][src2], sel1); + merge(&(mix[l][dst]), data, sel2); + } + } + } + for (int l = 0; l < PROGPOW_LANES; l++) + { + // global load to the 256 byte DAG entry + // every lane can access every part of the entry + uint32_t index = ((l ^ loop) % PROGPOW_LANES) * PROGPOW_DAG_LOADS; + for (int i = 0; i < PROGPOW_DAG_LOADS; i++) + data_g[l][i] = dag_data[index+i]; + } + + // Consume the global load data at the very end of the loop to allow full latency hiding + // Always merge into mix[0] to feed the offset calculation + for (int i = 0; i < PROGPOW_DAG_LOADS; i++) + { + int dst = (i==0) ? 0 : mix_seq_dst[(mix_seq_dst_cnt++)%PROGPOW_REGS]; + int sel = kiss99(&prog_rnd); + for (int l = 0; l < PROGPOW_LANES; l++) + merge(&(mix[l][dst]), data_g[l][i], sel); + } +} + +static bool progpow_hash( + ethash_return_value_t* ret, + node const* full_nodes, + ethash_light_t const light, + uint64_t full_size, + ethash_h256_t const header_hash, + uint64_t const nonce, + uint64_t const block_number +) +{ + uint32_t *g_dag = NULL; + + const hash32_t header; + memcpy((void *)&header, (void *)&header_hash, sizeof(header_hash)); + uint32_t c_dag[PROGPOW_CACHE_WORDS]; + if (full_nodes) { + g_dag = (uint32_t *) full_nodes; + for(int l = 0; l < PROGPOW_LANES; l++) + { + // Load random data into the cache + // TODO: should be a new blob of data, not existing DAG data + for (uint32_t word = l*4; word < PROGPOW_CACHE_WORDS; word += PROGPOW_LANES*4) + { + c_dag[word + 0] = g_dag[word + 0]; + c_dag[word + 1] = g_dag[word + 1]; + c_dag[word + 2] = g_dag[word + 2]; + c_dag[word + 3] = g_dag[word + 3]; + } + } + } else { + node tmp_node; + for(int l = 0; l < PROGPOW_LANES; l++) + { + for (uint32_t word = l*NODE_WORDS; word < PROGPOW_CACHE_WORDS; word += PROGPOW_LANES*NODE_WORDS) + { + ethash_calculate_dag_item(&tmp_node, word / NODE_WORDS, light); + memcpy((void *)&c_dag[word], (void *)&tmp_node.words[0], sizeof(uint32_t) * NODE_WORDS); + } + } + } + + uint32_t mix[PROGPOW_LANES][PROGPOW_REGS]; + hash32_t digest; + for (int i = 0; i < 8; i++) + digest.uint32s[i] = 0; + + // keccak(header..nonce) + hash32_t seed_256 = keccak_f800_progpow(header, nonce, digest); + // endian swap so byte 0 of the hash is the MSB of the value + uint64_t seed = (uint64_t)ethash_swap_u32(seed_256.uint32s[0]) << 32 | ethash_swap_u32(seed_256.uint32s[1]); + + // initialize mix for all lanes + for (int l = 0; l < PROGPOW_LANES; l++) + fill_mix(seed, l, mix[l]); + + uint64_t prog_seed = block_number / PROGPOW_PERIOD; + uint32_t dagWords = (unsigned)((uint32_t)full_size / PROGPOW_MIX_BYTES); + // execute the randomly generated inner loop + for (int i = 0; i < PROGPOW_CNT_DAG; i++) + { + progPowLoop(prog_seed, i, light, mix, g_dag, c_dag, dagWords); + } + + // Reduce mix data to a single per-lane result + uint32_t lane_hash[PROGPOW_LANES]; + for (int l = 0; l < PROGPOW_LANES; l++) + { + lane_hash[l] = 0x811c9dc5; + for (int i = 0; i < PROGPOW_REGS; i++) + fnv1a(&lane_hash[l], mix[l][i]); + } + // Reduce all lanes to a single 256-bit result + for (int i = 0; i < 8; i++) + digest.uint32s[i] = 0x811c9dc5; + + for (int l = 0; l < PROGPOW_LANES; l++) + fnv1a(&digest.uint32s[l%8], lane_hash[l]); + + memset((void *)&ret->mix_hash, 0, sizeof(ret->mix_hash)); + memcpy(&ret->mix_hash, (void *)&digest, sizeof(digest)); + memset((void *)&ret->result, 0, sizeof(ret->result)); + digest = keccak_f800_progpow(header, seed, digest); + memcpy((void *)&ret->result, (void *)&digest, sizeof(ret->result)); + + return true; +} + +ethash_return_value_t progpow_light_compute_internal( + ethash_light_t light, + uint64_t full_size, + ethash_h256_t const header_hash, + uint64_t nonce, + uint64_t block_number +) +{ + ethash_return_value_t ret; + ret.success = true; + if (!progpow_hash(&ret, NULL, light, full_size, header_hash, nonce, block_number)) { + ret.success = false; + } + return ret; +} + +ethash_return_value_t progpow_light_compute( + ethash_light_t light, + ethash_h256_t const header_hash, + uint64_t nonce, + uint64_t block_number +) +{ + uint64_t full_size = ethash_get_datasize(block_number); + return progpow_light_compute_internal(light, full_size, header_hash, nonce, block_number); +} + +ethash_return_value_t progpow_full_compute( + ethash_full_t full, + ethash_h256_t const header_hash, + uint64_t nonce, + uint64_t block_number +) +{ + ethash_return_value_t ret; + ret.success = true; + if (!progpow_hash( + &ret, + (node const*)full->data, + NULL, + full->file_size, + header_hash, + nonce, + block_number)) { + ret.success = false; + } + return ret; +} diff --git a/test/c/CMakeLists.txt b/test/c/CMakeLists.txt index f94531c3..e3025708 100644 --- a/test/c/CMakeLists.txt +++ b/test/c/CMakeLists.txt @@ -51,7 +51,7 @@ IF (Boost_FOUND) set(CMAKE_CXX_FLAGS "${CMAKE_CXX_FLAGS} -std=c++11 ") endif() - add_executable (Test "./test.cpp" ${HEADERS}) + add_executable (Test test.cpp test_progpow.cpp ${HEADERS}) target_link_libraries(Test ${ETHHASH_LIBS}) target_link_libraries(Test ${Boost_FILESYSTEM_LIBRARIES}) target_link_libraries(Test ${Boost_SYSTEM_LIBRARIES}) diff --git a/test/c/test_progpow.cpp b/test/c/test_progpow.cpp new file mode 100644 index 00000000..9742598b --- /dev/null +++ b/test/c/test_progpow.cpp @@ -0,0 +1,383 @@ +#include +#include +#include +#include +#include + +#ifdef WITH_CRYPTOPP + +#include + +#else +#include +#endif // WITH_CRYPTOPP + +#ifdef _WIN32 +#include +#include +#endif + +#include +#include +#include +#include +#include + +using namespace std; +using byte = uint8_t; +using bytes = std::vector; +namespace fs = boost::filesystem; + +// Just an alloca "wrapper" to silence uint64_t to size_t conversion warnings in windows +// consider replacing alloca calls with something better though! +#define our_alloca(param__) alloca((size_t)(param__)) + +// some functions taken from eth::dev for convenience. +static std::string bytesToHexString(const uint8_t *str, const uint64_t s) +{ + std::ostringstream ret; + + for (size_t i = 0; i < s; ++i) + ret << std::hex << std::setfill('0') << std::setw(2) << std::nouppercase << (int) str[i]; + + return ret.str(); +} + +static std::string blockhashToHexString(ethash_h256_t* _hash) +{ + return bytesToHexString((uint8_t*)_hash, 32); +} + +static int fromHex(char _i) +{ + if (_i >= '0' && _i <= '9') + return _i - '0'; + if (_i >= 'a' && _i <= 'f') + return _i - 'a' + 10; + if (_i >= 'A' && _i <= 'F') + return _i - 'A' + 10; + + BOOST_REQUIRE_MESSAGE(false, "should never get here"); + return -1; +} + +static bytes hexStringToBytes(std::string const& _s) +{ + unsigned s = (_s[0] == '0' && _s[1] == 'x') ? 2 : 0; + std::vector ret; + ret.reserve((_s.size() - s + 1) / 2); + + if (_s.size() % 2) + try + { + ret.push_back(fromHex(_s[s++])); + } + catch (...) + { + ret.push_back(0); + } + for (unsigned i = s; i < _s.size(); i += 2) + try + { + ret.push_back((byte)(fromHex(_s[i]) * 16 + fromHex(_s[i + 1]))); + } + catch (...){ + ret.push_back(0); + } + return ret; +} + +static ethash_h256_t stringToBlockhash(std::string const& _s) +{ + ethash_h256_t ret; + bytes b = hexStringToBytes(_s); + memcpy(&ret, b.data(), b.size()); + return ret; +} + +/* ProgPoW */ + +static void ethash_keccakf800(uint32_t state[25]) +{ + for (int i = 0; i < 22; ++i) + keccak_f800_round(state, i); +} + +BOOST_AUTO_TEST_CASE(test_progpow_math) +{ + typedef struct { + uint32_t a; + uint32_t b; + uint32_t exp; + } mytest; + + mytest tests[] = { + {20, 22, 42}, + {70000, 80000, 1305032704}, + {70000, 80000, 1}, + {1, 2, 1}, + {3, 10000, 196608}, + {3, 0, 3}, + {3, 6, 2}, + {3, 6, 7}, + {3, 6, 5}, + {0, 0xffffffff, 32}, + {3 << 13, 1 << 5, 3}, + {22, 20, 42}, + {80000, 70000, 1305032704}, + {80000, 70000, 1}, + {2, 1, 1}, + {10000, 3, 80000}, + {0, 3, 0}, + {6, 3, 2}, + {6, 3, 7}, + {6, 3, 5}, + {0, 0xffffffff, 32}, + {3 << 13, 1 << 5, 3}, + }; + + for (int i = 0; i < sizeof(tests) / sizeof(mytest); i++) { + uint32_t res = progpowMath(tests[i].a, tests[i].b, (uint32_t)i); + BOOST_REQUIRE_EQUAL(res, tests[i].exp); + } +} + +BOOST_AUTO_TEST_CASE(test_progpow_merge) +{ + typedef struct { + uint32_t a; + uint32_t b; + uint32_t exp; + } mytest; + mytest tests[] = { + {1000000, 101, 33000101}, + {2000000, 102, 66003366}, + {3000000, 103, 2999975}, + {4000000, 104, 4000104}, + {1000000, 0, 33000000}, + {2000000, 0, 66000000}, + {3000000, 0, 3000000}, + {4000000, 0, 4000000}, + }; + for (int i = 0; i < sizeof(tests) / sizeof(mytest); i++) { + uint32_t res = tests[i].a; + merge(&res, tests[i].b, (uint32_t)i); + BOOST_REQUIRE_EQUAL(res, tests[i].exp); + } +} + +BOOST_AUTO_TEST_CASE(test_progpow_keccak) +{ + // Test vectors from + // https://github.com/XKCP/XKCP/blob/master/tests/TestVectors/KeccakF-800-IntermediateValues.txt. + uint32_t state[25] = {}; + const uint32_t expected_state_0[] = {0xE531D45D, 0xF404C6FB, 0x23A0BF99, 0xF1F8452F, 0x51FFD042, + 0xE539F578, 0xF00B80A7, 0xAF973664, 0xBF5AF34C, 0x227A2424, 0x88172715, 0x9F685884, + 0xB15CD054, 0x1BF4FC0E, 0x6166FA91, 0x1A9E599A, 0xA3970A1F, 0xAB659687, 0xAFAB8D68, + 0xE74B1015, 0x34001A98, 0x4119EFF3, 0x930A0E76, 0x87B28070, 0x11EFE996}; + ethash_keccakf800(state); + for (size_t i = 0; i < 25; ++i) + BOOST_REQUIRE_EQUAL(state[i], expected_state_0[i]); + const uint32_t expected_state_1[] = {0x75BF2D0D, 0x9B610E89, 0xC826AF40, 0x64CD84AB, 0xF905BDD6, + 0xBC832835, 0x5F8001B9, 0x15662CCE, 0x8E38C95E, 0x701FE543, 0x1B544380, 0x89ACDEFF, + 0x51EDB5DE, 0x0E9702D9, 0x6C19AA16, 0xA2913EEE, 0x60754E9A, 0x9819063C, 0xF4709254, + 0xD09F9084, 0x772DA259, 0x1DB35DF7, 0x5AA60162, 0x358825D5, 0xB3783BAB}; + ethash_keccakf800(state); + for (size_t i = 0; i < 25; ++i) + BOOST_REQUIRE_EQUAL(state[i], expected_state_1[i]); +} + +BOOST_AUTO_TEST_CASE(test_progpow_block0_verification) { + // epoch 0 + ethash_light_t light = ethash_light_new(1045); + ethash_h256_t seedhash = stringToBlockhash("5fc898f16035bf5ac9c6d9077ae1e3d5fc1ecc3c9fd5bee8bb00e810fdacbaa0"); + BOOST_ASSERT(light); + ethash_return_value_t ret = progpow_light_compute( + light, + seedhash, + 0x50377003e5d830caU, + 1045 + ); + //ethash_h256_t difficulty = ethash_h256_static_init(0x25, 0xa6, 0x1e); + //BOOST_REQUIRE(ethash_check_difficulty(&ret.result, &difficulty)); + ethash_light_delete(light); +} + +BOOST_AUTO_TEST_CASE(test_progpow_keccak_f800) { + ethash_h256_t seedhash; + ethash_h256_t headerhash = stringToBlockhash("0000000000000000000000000000000000000000000000000000000000000000"); + + { + const std::string + seedexp = "5dd431e5fbc604f499bfa0232f45f8f142d0ff5178f539e5a7800bf0643697af"; + const std::string header_string = blockhashToHexString(&headerhash); + BOOST_REQUIRE_MESSAGE(true, + "\nheader: " << header_string.c_str() << "\n"); + hash32_t result; + for (int i = 0; i < 8; i++) + result.uint32s[i] = 0; + + hash32_t header; + memcpy((void *)&header, (void *)&headerhash, sizeof(headerhash)); + uint64_t nonce = 0x0; + // keccak(header..nonce) + hash32_t seed_256 = keccak_f800_progpow(header, nonce, result); + uint64_t seed = (uint64_t)ethash_swap_u32(seed_256.uint32s[0]) << 32 | ethash_swap_u32(seed_256.uint32s[1]); + uint64_t exp = 0x5dd431e5fbc604f4U; + + BOOST_REQUIRE_MESSAGE(seed == exp, + "\nseed: " << seed << "exp: " << exp << "\n"); + ethash_h256_t out; + memcpy((void *)&out, (void *)&seed_256, sizeof(result)); + const std::string out_string = blockhashToHexString(&out); + BOOST_REQUIRE_MESSAGE(out_string == seedexp, + "\nresult: " << out_string.c_str() << "\n"); + } +} + +BOOST_AUTO_TEST_CASE(test_progpow_full_client_checks) { + uint64_t full_size = ethash_get_datasize(0); + uint64_t cache_size = ethash_get_cachesize(0); + ethash_h256_t difficulty; + ethash_return_value_t light_out; + ethash_return_value_t full_out; + ethash_h256_t hash = stringToBlockhash("0000000000000000000000000000000000000000000000000000000000000000"); + ethash_h256_t seed = stringToBlockhash("0000000000000000000000000000000000000000000000000000000000000000"); + + // Set the difficulty + ethash_h256_set(&difficulty, 0, 197); + ethash_h256_set(&difficulty, 1, 90); + for (int i = 2; i < 32; i++) + ethash_h256_set(&difficulty, i, 255); + + ethash_light_t light = ethash_light_new_internal(cache_size, &seed); + ethash_full_t full = ethash_full_new_internal( + "./test_ethash_directory/", + seed, + full_size, + light, + NULL + ); + { + uint64_t nonce = 0x0; + full_out = progpow_full_compute(full, hash, nonce, 0); + BOOST_REQUIRE(full_out.success); + + const std::string + exphead = "7ea12cfc33f64616ab7dbbddf3362ee7dd3e1e20d60d860a85c51d6559c912c4", + expmix = "a09ffaa0f2b5d47a98c2d4fbc0e90936710dd2b2a220fce04e8d55a6c6a093d6"; + const std::string seed_string = blockhashToHexString(&seed); + const std::string hash_string = blockhashToHexString(&hash); + + const std::string full_mix_hash_string = blockhashToHexString(&full_out.mix_hash); + BOOST_REQUIRE_MESSAGE(full_mix_hash_string == expmix, + "\nfull mix hash: " << full_mix_hash_string.c_str() << "\n"); + const std::string full_result_string = blockhashToHexString(&full_out.result); + BOOST_REQUIRE_MESSAGE(full_result_string == exphead, + "\nfull result: " << full_result_string.c_str() << "\n"); + } + + ethash_light_delete(light); + ethash_full_delete(full); + //fs::remove_all("./test_ethash_directory/"); +} + +BOOST_AUTO_TEST_CASE(test_progpow_light_client_checks) { + uint64_t full_size = ethash_get_datasize(0); + uint64_t cache_size = ethash_get_cachesize(0); + ethash_return_value_t light_out; + ethash_h256_t hash = stringToBlockhash("0000000000000000000000000000000000000000000000000000000000000000"); + ethash_h256_t seed = stringToBlockhash("0000000000000000000000000000000000000000000000000000000000000000"); + ethash_light_t light = ethash_light_new_internal(cache_size, &seed); + { + uint64_t nonce = 0x0; + const std::string + exphead = "63155f732f2bf556967f906155b510c917e48e99685ead76ea83f4eca03ab12b", + expmix = "faeb1be51075b03a4ff44b335067951ead07a3b078539ace76fd56fc410557a3"; + const std::string hash_string = blockhashToHexString(&hash); + + light_out = progpow_light_compute_internal(light, full_size, hash, nonce, 0); + BOOST_REQUIRE(light_out.success); + + const std::string light_result_string = blockhashToHexString(&light_out.result); + BOOST_REQUIRE_MESSAGE(exphead == light_result_string, + "\nlight result: " << light_result_string.c_str() << "\n" + << "exp result: " << exphead.c_str() << "\n"); + const std::string light_mix_hash_string = blockhashToHexString(&light_out.mix_hash); + BOOST_REQUIRE_MESSAGE(expmix == light_mix_hash_string, + "\nlight mix hash: " << light_mix_hash_string.c_str() << "\n" + << "exp mix hash: " << expmix.c_str() << "\n"); + } + + ethash_light_delete(light); +} + +/// Defines a test case for ProgPoW hash() function. (from chfast/ethash/test/unittests/progpow_test_vectors.hpp) +struct progpow_hash_test_case +{ + int block_number; + const char* header_hash_hex; + const char* nonce_hex; + const char* mix_hash_hex; + const char* final_hash_hex; +}; + +progpow_hash_test_case progpow_hash_test_cases[] = { + {0, "0000000000000000000000000000000000000000000000000000000000000000", "0000000000000000", + "faeb1be51075b03a4ff44b335067951ead07a3b078539ace76fd56fc410557a3", + "63155f732f2bf556967f906155b510c917e48e99685ead76ea83f4eca03ab12b"}, + {49, "63155f732f2bf556967f906155b510c917e48e99685ead76ea83f4eca03ab12b", "0000000006ff2c47", + "c789c1180f890ec555ff42042913465481e8e6bc512cb981e1c1108dc3f2227d", + "9e7248f20914913a73d80a70174c331b1d34f260535ac3631d770e656b5dd922"}, + {50, "9e7248f20914913a73d80a70174c331b1d34f260535ac3631d770e656b5dd922", "00000000076e482e", + "c7340542c2a06b3a7dc7222635f7cd402abf8b528ae971ddac6bbe2b0c7cb518", + "de37e1824c86d35d154cf65a88de6d9286aec4f7f10c3fc9f0fa1bcc2687188d"}, + {99, "de37e1824c86d35d154cf65a88de6d9286aec4f7f10c3fc9f0fa1bcc2687188d", "000000003917afab", + "f5e60b2c5bfddd136167a30cbc3c8dbdbd15a512257dee7964e0bc6daa9f8ba7", + "ac7b55e801511b77e11d52e9599206101550144525b5679f2dab19386f23dcce"}, + {29950, "ac7b55e801511b77e11d52e9599206101550144525b5679f2dab19386f23dcce", "005d409dbc23a62a", + "07393d15805eb08ee6fc6cb3ad4ad1010533bd0ff92d6006850246829f18fd6e", + "e43d7e0bdc8a4a3f6e291a5ed790b9fa1a0948a2b9e33c844888690847de19f5"}, + {29999, "e43d7e0bdc8a4a3f6e291a5ed790b9fa1a0948a2b9e33c844888690847de19f5", "005db5fa4c2a3d03", + "7551bddf977491da2f6cfc1679299544b23483e8f8ee0931c4c16a796558a0b8", + "d34519f72c97cae8892c277776259db3320820cb5279a299d0ef1e155e5c6454"}, + {30000, "d34519f72c97cae8892c277776259db3320820cb5279a299d0ef1e155e5c6454", "005db8607994ff30", + "f1c2c7c32266af9635462e6ce1c98ebe4e7e3ecab7a38aaabfbf2e731e0fbff4", + "8b6ce5da0b06d18db7bd8492d9e5717f8b53e7e098d9fef7886d58a6e913ef64"}, + {30049, "8b6ce5da0b06d18db7bd8492d9e5717f8b53e7e098d9fef7886d58a6e913ef64", "005e2e215a8ca2e7", + "57fe6a9fbf920b4e91deeb66cb0efa971e08229d1a160330e08da54af0689add", + "c2c46173481b9ced61123d2e293b42ede5a1b323210eb2a684df0874ffe09047"}, + {30050, "c2c46173481b9ced61123d2e293b42ede5a1b323210eb2a684df0874ffe09047", "005e30899481055e", + "ba30c61cc5a2c74a5ecaf505965140a08f24a296d687e78720f0b48baf712f2d", + "ea42197eb2ba79c63cb5e655b8b1f612c5f08aae1a49ff236795a3516d87bc71"}, + {30099, "ea42197eb2ba79c63cb5e655b8b1f612c5f08aae1a49ff236795a3516d87bc71", "005ea6aef136f88b", + "cfd5e46048cd133d40f261fe8704e51d3f497fc14203ac6a9ef6a0841780b1cd", + "49e15ba4bf501ce8fe8876101c808e24c69a859be15de554bf85dbc095491bd6"}, + {59950, "49e15ba4bf501ce8fe8876101c808e24c69a859be15de554bf85dbc095491bd6", "02ebe0503bd7b1da", + "21511fbaa31fb9f5fc4998a754e97b3083a866f4de86fa7500a633346f56d773", + "f5c50ba5c0d6210ddb16250ec3efda178de857b2b1703d8d5403bd0f848e19cf"}, + {59999, "f5c50ba5c0d6210ddb16250ec3efda178de857b2b1703d8d5403bd0f848e19cf", "02edb6275bd221e3", + "653eda37d337e39d311d22be9bbd3458d3abee4e643bee4a7280a6d08106ef98", + "341562d10d4afb706ec2c8d5537cb0c810de02b4ebb0a0eea5ae335af6fb2e88"}, +}; + +BOOST_AUTO_TEST_CASE(test_progpow_test_cases) { + ethash_light_t light; + uint32_t epoch = -1; + for (int i = 0; i < sizeof(progpow_hash_test_cases) / sizeof(progpow_hash_test_case); i++) + { + progpow_hash_test_case *t; + t = &progpow_hash_test_cases[i]; + const auto epoch_number = t->block_number / ETHASH_EPOCH_LENGTH; + if (!light || epoch != epoch_number) + light = ethash_light_new(t->block_number); + epoch = epoch_number; + ethash_h256_t hash = stringToBlockhash(t->header_hash_hex); + uint64_t nonce = strtoul(t->nonce_hex, NULL, 16); + ethash_return_value_t light_out = progpow_light_compute(light, hash, nonce, t->block_number); + BOOST_REQUIRE_EQUAL(blockhashToHexString(&light_out.result), t->final_hash_hex); + BOOST_REQUIRE_EQUAL(blockhashToHexString(&light_out.mix_hash), t->mix_hash_hex); + printf("next...\n"); + } + ethash_light_delete(light); +}