popov/cpuminer-opt-gpu
1
0
Fork 0
mirror of https://github.com/JayDDee/cpuminer-opt.git synced 2025-09-17 23:44:27 +00:00
Code Issues Packages Projects Releases Wiki Activity

v3.9.5

Browse Source
This commit is contained in:
Jay D Dee
2019-06-26 14:16:01 -04:00
parent d6e8d7a46e
commit 0d48d573ce
174 changed files with 1352 additions and 1556 deletions
Download Patch File Download Diff File Expand all files Collapse all files

6
algo/x17/sonoa-4way.c
View File

@@ -803,7 +803,7 @@ void sonoa_4way_hash( void *state, const void *input )
haval256_5_4way_close( &ctx.haval, state );
}
int scanhash_sonoa_4way( int thr_id, struct work *work, uint32_t max_nonce,
int scanhash_sonoa_4way( struct work *work, uint32_t max_nonce,
uint64_t *hashes_done, struct thr_info *mythr )
{
uint32_t hash[4*8] __attribute__ ((aligned (64)));
@@ -816,7 +816,7 @@ int scanhash_sonoa_4way( int thr_id, struct work *work, uint32_t max_nonce,
const uint32_t first_nonce = pdata[19];
__m256i *noncev = (__m256i*)vdata + 9; // aligned
const uint32_t Htarg = ptarget[7];
/* int */ thr_id = mythr->id; // thr_id arg is deprecated
int thr_id = mythr->id; // thr_id arg is deprecated
uint64_t htmax[] = { 0, 0xF, 0xFF,
0xFFF, 0xFFFF, 0x10000000 };
uint32_t masks[] = { 0xFFFFFFFF, 0xFFFFFFF0, 0xFFFFFF00,
@@ -841,7 +841,7 @@ int scanhash_sonoa_4way( int thr_id, struct work *work, uint32_t max_nonce,
if ( fulltest( lane_hash, ptarget ) && !opt_benchmark )
{
pdata[19] = n + lane;
submit_solution( work, lane_hash, mythr, lane );
submit_lane_solution( work, lane_hash, mythr, lane );
}
}
n += 4;

4
algo/x17/sonoa-gate.h
View File

@@ -14,7 +14,7 @@ bool register_sonoa_algo( algo_gate_t* gate );
void sonoa_4way_hash( void *state, const void *input );
int scanhash_sonoa_4way( int thr_id, struct work *work, uint32_t max_nonce,
int scanhash_sonoa_4way( struct work *work, uint32_t max_nonce,
uint64_t *hashes_done, struct thr_info *mythr );
//void init_sonoa_4way_ctx();
@@ -23,7 +23,7 @@ int scanhash_sonoa_4way( int thr_id, struct work *work, uint32_t max_nonce,
void sonoa_hash( void *state, const void *input );
int scanhash_sonoa( int thr_id, struct work *work, uint32_t max_nonce,
int scanhash_sonoa( struct work *work, uint32_t max_nonce,
uint64_t *hashes_done, struct thr_info *mythr );
void init_sonoa_ctx();

65
algo/x17/sonoa.c
View File

@@ -564,17 +564,17 @@ void sonoa_hash( void *state, const void *input )
memcpy(state, hash, 32);
}
int scanhash_sonoa( int thr_id, struct work *work, uint32_t max_nonce,
int scanhash_sonoa( struct work *work, uint32_t max_nonce,
uint64_t *hashes_done, struct thr_info *mythr )
{
uint32_t _ALIGN(128) hash32[8];
uint32_t _ALIGN(128) endiandata[20];
uint32_t *pdata = work->data;
uint32_t *ptarget = work->target;
const uint32_t first_nonce = pdata[19];
const uint32_t Htarg = ptarget[7];
uint32_t n = pdata[19] - 1;
/* int */ thr_id = mythr->id; // thr_id arg is deprecated
uint32_t _ALIGN(128) hash32[8];
uint32_t _ALIGN(128) endiandata[20];
uint32_t *pdata = work->data;
uint32_t *ptarget = work->target;
const uint32_t first_nonce = pdata[19];
const uint32_t Htarg = ptarget[7];
uint32_t n = pdata[19] - 1;
int thr_id = mythr->id; // thr_id arg is deprecated
uint64_t htmax[] =
{
@@ -603,45 +603,20 @@ uint32_t n = pdata[19] - 1;
casti_m128i( endiandata, 3 ) = mm128_bswap_32( casti_m128i( pdata, 3 ) );
casti_m128i( endiandata, 4 ) = mm128_bswap_32( casti_m128i( pdata, 4 ) );
#ifdef DEBUG_ALGO
printf("[%d] Htarg=%X\n", thr_id, Htarg);
#endif
for ( int m = 0; m < 6; m++ )
for ( int m = 0; m < 6; m++ ) if ( Htarg <= htmax[m] )
{
if ( Htarg <= htmax[m] )
{
uint32_t mask = masks[m];
do
{
pdata[19] = ++n;
be32enc(&endiandata[19], n);
sonoa_hash(hash32, endiandata);
#ifndef DEBUG_ALGO
if ( ( !( hash32[7] & mask ) ) && fulltest( hash32, ptarget ) )
{
work_set_target_ratio( work, hash32 );
*hashes_done = n - first_nonce + 1;
return 1;
}
#else
if (!(n % 0x1000) && !thr_id) printf(".");
if ( !(hash32[7] & mask) )
{
printf("[%d]",thr_id);
if ( fulltest( hash32, ptarget ) )
{
work_set_target_ratio( work, hash32 );
*hashes_done = n - first_nonce + 1;
return 1;
}
}
#endif
uint32_t mask = masks[m];
do
{
pdata[19] = ++n;
be32enc(&endiandata[19], n);
sonoa_hash(hash32, endiandata);
if ( !( hash32[7] & mask ) )
if ( fulltest( hash32, ptarget ) && !opt_benchmark )
submit_solution( work, hash32, mythr );
} while (n < max_nonce && !work_restart[thr_id].restart);
// see blake.c if else to understand the loop on htmax => mask
break;
break;
}
}
*hashes_done = n - first_nonce + 1;
pdata[19] = n;
return 0;

6
algo/x17/x17-4way.c
View File

@@ -202,7 +202,7 @@ void x17_4way_hash( void *state, const void *input )
haval256_5_4way_close( &ctx.haval, state );
}
int scanhash_x17_4way( int thr_id, struct work *work, uint32_t max_nonce,
int scanhash_x17_4way( struct work *work, uint32_t max_nonce,
uint64_t *hashes_done, struct thr_info *mythr )
{
uint32_t hash[4*8] __attribute__ ((aligned (64)));
@@ -214,7 +214,7 @@ int scanhash_x17_4way( int thr_id, struct work *work, uint32_t max_nonce,
uint32_t n = pdata[19];
const uint32_t first_nonce = pdata[19];
__m256i *noncev = (__m256i*)vdata + 9; // aligned
/* int */ thr_id = mythr->id; // thr_id arg is deprecated
int thr_id = mythr->id; // thr_id arg is deprecated
const uint32_t Htarg = ptarget[7];
uint64_t htmax[] = { 0, 0xF, 0xFF,
0xFFF, 0xFFFF, 0x10000000 };
@@ -239,7 +239,7 @@ int scanhash_x17_4way( int thr_id, struct work *work, uint32_t max_nonce,
if ( fulltest( lane_hash, ptarget ) && !opt_benchmark )
{
pdata[19] = n + lane;
submit_solution( work, lane_hash, mythr, lane );
submit_lane_solution( work, lane_hash, mythr, lane );
}
}
n += 4;

4
algo/x17/x17-gate.h
View File

@@ -13,13 +13,13 @@ bool register_x17_algo( algo_gate_t* gate );
#if defined(X17_4WAY)
void x17_4way_hash( void *state, const void *input );
int scanhash_x17_4way( int thr_id, struct work *work, uint32_t max_nonce,
int scanhash_x17_4way( struct work *work, uint32_t max_nonce,
uint64_t *hashes_done, struct thr_info *mythr );
#endif
void x17_hash( void *state, const void *input );
int scanhash_x17( int thr_id, struct work *work, uint32_t max_nonce,
int scanhash_x17( struct work *work, uint32_t max_nonce,
uint64_t *hashes_done, struct thr_info *mythr );
#endif

36
algo/x17/x17.c
View File

@@ -181,7 +181,7 @@ void x17_hash(void *output, const void *input)
sph_haval256_5_close( &ctx.haval, output );
}
int scanhash_x17( int thr_id, struct work *work, uint32_t max_nonce,
int scanhash_x17( struct work *work, uint32_t max_nonce,
uint64_t *hashes_done, struct thr_info *mythr)
{
uint32_t endiandata[20] __attribute__((aligned(64)));
@@ -191,7 +191,7 @@ int scanhash_x17( int thr_id, struct work *work, uint32_t max_nonce,
uint32_t n = pdata[19] - 1;
const uint32_t first_nonce = pdata[19];
const uint32_t Htarg = ptarget[7];
/* int */ thr_id = mythr->id; // thr_id arg is deprecated
int thr_id = mythr->id; // thr_id arg is deprecated
uint64_t htmax[] =
{
@@ -219,10 +219,6 @@ int scanhash_x17( int thr_id, struct work *work, uint32_t max_nonce,
casti_m128i( endiandata, 3 ) = mm128_bswap_32( casti_m128i( pdata, 3 ) );
casti_m128i( endiandata, 4 ) = mm128_bswap_32( casti_m128i( pdata, 4 ) );
#ifdef DEBUG_ALGO
if ( Htarg != 0 )
printf( "[%d] Htarg=%X\n", thr_id, Htarg );
#endif
for ( int m = 0; m < 6; m++ )
{
if ( Htarg <= htmax[m] )
@@ -233,32 +229,10 @@ int scanhash_x17( int thr_id, struct work *work, uint32_t max_nonce,
pdata[19] = ++n;
be32enc( &endiandata[19], n );
x17_hash( hash64, endiandata );
#ifndef DEBUG_ALGO
if ( !( hash64[7] & mask ) )
{
if ( fulltest( hash64, ptarget ) )
{
*hashes_done = n - first_nonce + 1;
return true;
}
// else
// applog(LOG_INFO, "Result does not validate on CPU!");
}
#else
if ( !( n % 0x1000 ) && !thr_id ) printf(".");
if ( !( hash64[7] & mask ) )
{
printf("[%d]",thr_id);
if ( fulltest( hash64, ptarget ) )
{
work_set_target_ratio( work, hash64 );
*hashes_done = n - first_nonce + 1;
return true;
}
}
#endif
} while (n < max_nonce && !work_restart[thr_id].restart);
// see blake.c if else to understand the loop on htmax => mask
if ( fulltest( hash64, ptarget ) && !opt_benchmark )
submit_solution( work, hash64, mythr );
} while ( n < max_nonce && !work_restart[thr_id].restart);
break;
}
}

6
algo/x17/xevan-4way.c
View File

@@ -329,7 +329,7 @@ void xevan_4way_hash( void *output, const void *input )
haval256_5_4way_close( &ctx.haval, output );
}
int scanhash_xevan_4way( int thr_id, struct work *work, uint32_t max_nonce,
int scanhash_xevan_4way( struct work *work, uint32_t max_nonce,
uint64_t *hashes_done, struct thr_info *mythr )
{
uint32_t hash[4*8] __attribute__ ((aligned (64)));
@@ -338,7 +338,7 @@ int scanhash_xevan_4way( int thr_id, struct work *work, uint32_t max_nonce,
uint32_t vdata[24*4] __attribute__ ((aligned (64)));
uint32_t *pdata = work->data;
uint32_t *ptarget = work->target;
/* int */ thr_id = mythr->id; // thr_id arg is deprecated
int thr_id = mythr->id; // thr_id arg is deprecated
__m256i *noncev = (__m256i*)vdata + 9; // aligned
const uint32_t Htarg = ptarget[7];
@@ -361,7 +361,7 @@ int scanhash_xevan_4way( int thr_id, struct work *work, uint32_t max_nonce,
if ( fulltest( lane_hash, ptarget ) && !opt_benchmark )
{
pdata[19] = n + lane;
submit_solution( work, lane_hash, mythr, lane );
submit_lane_solution( work, lane_hash, mythr, lane );
}
}
n += 4;

4
algo/x17/xevan-gate.h
View File

@@ -14,7 +14,7 @@ bool register_xevan_algo( algo_gate_t* gate );
void xevan_4way_hash( void *state, const void *input );
int scanhash_xevan_4way( int thr_id, struct work *work, uint32_t max_nonce,
int scanhash_xevan_4way( struct work *work, uint32_t max_nonce,
uint64_t *hashes_done, struct thr_info *mythr );
//void init_xevan_4way_ctx();
@@ -23,7 +23,7 @@ int scanhash_xevan_4way( int thr_id, struct work *work, uint32_t max_nonce,
void xevan_hash( void *state, const void *input );
int scanhash_xevan( int thr_id, struct work *work, uint32_t max_nonce,
int scanhash_xevan( struct work *work, uint32_t max_nonce,
uint64_t *hashes_done, struct thr_info *mythr );
void init_xevan_ctx();

17
algo/x17/xevan.c
View File

@@ -230,14 +230,14 @@ void xevan_hash(void *output, const void *input)
memcpy(output, hash, 32);
}
int scanhash_xevan( int thr_id, struct work *work, uint32_t max_nonce,
int scanhash_xevan( struct work *work, uint32_t max_nonce,
uint64_t *hashes_done, struct thr_info *mythr )
{
uint32_t _ALIGN(64) hash[8];
uint32_t _ALIGN(64) endiandata[20];
uint32_t *pdata = work->data;
uint32_t *ptarget = work->target;
/* int */ thr_id = mythr->id; // thr_id arg is deprecated
int thr_id = mythr->id; // thr_id arg is deprecated
const uint32_t Htarg = ptarget[7];
const uint32_t first_nonce = pdata[19];
uint32_t nonce = first_nonce;
@@ -254,15 +254,14 @@ int scanhash_xevan( int thr_id, struct work *work, uint32_t max_nonce,
be32enc(&endiandata[19], nonce);
xevan_hash(hash, endiandata);
if (hash[7] <= Htarg && fulltest(hash, ptarget)) {
work_set_target_ratio(work, hash);
pdata[19] = nonce;
*hashes_done = pdata[19] - first_nonce;
return 1;
if (hash[7] <= Htarg )
if ( fulltest( hash, ptarget ) && !opt_benchmark )
{
pdata[19] = nonce;
submit_solution( work, hash, mythr );
}
nonce++;
} while (nonce < max_nonce && !(*restart));
} while ( nonce < max_nonce && !(*restart) );
pdata[19] = nonce;
*hashes_done = pdata[19] - first_nonce + 1;