v3.12.4.3

README.md

@@ -12,14 +12,23 @@ a false positive, they are flagged simply because they are cryptocurrency
 miners. The source code is open for anyone to inspect. If you don't trust 
 the software, don't use it.
 
+
+New thread:
+
+https://bitcointalk.org/index.php?topic=5226770.msg53865575#msg53865575
+
+Old thread:
+
 https://bitcointalk.org/index.php?topic=1326803.0
 
 mailto://jayddee246@gmail.com
 
+This note is to confirm that bitcointalk users JayDDee and joblo are the
+same person.
 
 I created a new BCT user JayDDee to match my github user id.
-Over timme all discussion will be migrated using the new user.
+The old thread has been locked but still contains useful information for
-This note is to confirm JayDDee is the sae person as joblo.
+reading.
 
 See file RELEASE_NOTES for change log and INSTALL_LINUX or INSTALL_WINDOWS
 for compile instructions.

RELEASE_NOTES

@@ -65,9 +65,34 @@ If not what makes it happen or not happen?
 Change Log
 ----------
 
+v3.12.4.3
+
+Fixed segfault in new block log for getwork.
+
+Disabled silent discarding of stale work after the submit is logged.
+
+v3.12.4.2
+
+Issue #245: fixed getwork stale shares, solo mining with getwork now works.
+
+Issue #246: implemented block and summary logs for getwork.
+
+v3.12.4.1
+
+Issue #245: fix scantime when mining solo with getwork.
+
+Added debug logs for creation of stratum and longpoll threads, use -D to
+enable.
+
+v3.12.4
+
+Issue #244: Change longpoll to ignore job id.
+
+Lyra2rev2 AVX2 +3%, AVX512 +6%.
+
 v3.12.3.1
 
-Issue #241: Fixed regression that broke GBT in v3.11.2.
+Issue #241: Fixed regression that broke coinbase address in v3.11.7.
 
 v3.12.3
 

algo/lyra2/lyra2-gate.c

@@ -94,12 +94,12 @@ bool lyra2rev2_thread_init()
    const int64_t ROW_LEN_BYTES = ROW_LEN_INT64 * 8;
 
    int size = (int64_t)ROW_LEN_BYTES * 4; // nRows;
-#if defined (LYRA2REV2_8WAY)
+#if defined (LYRA2REV2_16WAY)
    l2v2_wholeMatrix = _mm_malloc( 2 * size, 64 );   // 2 way
-   init_lyra2rev2_8way_ctx();;
+   init_lyra2rev2_16way_ctx();;
-#elif defined (LYRA2REV2_4WAY)
+#elif defined (LYRA2REV2_8WAY)
    l2v2_wholeMatrix = _mm_malloc( size, 64 );
-   init_lyra2rev2_4way_ctx();;
+   init_lyra2rev2_8way_ctx();;
 #else
    l2v2_wholeMatrix = _mm_malloc( size, 64 );
    init_lyra2rev2_ctx();
@@ -109,12 +109,12 @@ bool lyra2rev2_thread_init()
 
 bool register_lyra2rev2_algo( algo_gate_t* gate )
 {
-#if defined (LYRA2REV2_8WAY)
+#if defined (LYRA2REV2_16WAY)
+  gate->scanhash  = (void*)&scanhash_lyra2rev2_16way;
+  gate->hash      = (void*)&lyra2rev2_16way_hash;
+#elif defined (LYRA2REV2_8WAY)
   gate->scanhash  = (void*)&scanhash_lyra2rev2_8way;
   gate->hash      = (void*)&lyra2rev2_8way_hash;
-#elif defined (LYRA2REV2_4WAY)
-  gate->scanhash  = (void*)&scanhash_lyra2rev2_4way;
-  gate->hash      = (void*)&lyra2rev2_4way_hash;
 #else
   gate->scanhash  = (void*)&scanhash_lyra2rev2;
   gate->hash      = (void*)&lyra2rev2_hash;

algo/lyra2/lyra2-gate.h

@@ -51,30 +51,32 @@ bool init_lyra2rev3_ctx();
 //////////////////////////////////
 
 #if defined(__AVX512F__) && defined(__AVX512VL__) && defined(__AVX512DQ__) && defined(__AVX512BW__)
-  #define LYRA2REV2_8WAY 1
+  #define LYRA2REV2_16WAY 1
 #elif defined(__AVX2__)
-  #define LYRA2REV2_4WAY 1
+  #define LYRA2REV2_8WAY 1
 #endif
 
 extern __thread uint64_t* l2v2_wholeMatrix;
 
 bool register_lyra2rev2_algo( algo_gate_t* gate );
 
-#if defined(LYRA2REV2_8WAY)
+#if defined(LYRA2REV2_16WAY)
+
+void lyra2rev2_16way_hash( void *state, const void *input );
+int scanhash_lyra2rev2_16way( struct work *work, uint32_t max_nonce,
+                             uint64_t *hashes_done, struct thr_info *mythr );
+bool init_lyra2rev2_16way_ctx();
+
+#elif defined(LYRA2REV2_8WAY)
 
 void lyra2rev2_8way_hash( void *state, const void *input );
 int scanhash_lyra2rev2_8way( struct work *work, uint32_t max_nonce,
                              uint64_t *hashes_done, struct thr_info *mythr );
 bool init_lyra2rev2_8way_ctx();
 
-#elif defined(LYRA2REV2_4WAY)
-
-void lyra2rev2_4way_hash( void *state, const void *input );
-int scanhash_lyra2rev2_4way( struct work *work, uint32_t max_nonce,
-                             uint64_t *hashes_done, struct thr_info *mythr );
-bool init_lyra2rev2_4way_ctx();
 
 #else
+
 void lyra2rev2_hash( void *state, const void *input );
 int scanhash_lyra2rev2( struct work *work, uint32_t max_nonce,
                         uint64_t *hashes_done, struct thr_info *mythr );

algo/lyra2/lyra2rev2-4way.c

@@ -8,12 +8,30 @@
 #include "algo/cubehash/cube-hash-2way.h"
 
 
-#if 0
+#if defined (LYRA2REV2_16WAY)
-void lyra2rev2_8way_hash( void *state, const void *input )
+
+typedef struct {
+   blake256_16way_context    blake;
+   keccak256_8way_context    keccak;
+   cubehashParam             cube;
+   skein256_8way_context     skein;
+   bmw256_16way_context      bmw;
+} lyra2v2_16way_ctx_holder __attribute__ ((aligned (64)));
+
+static lyra2v2_16way_ctx_holder l2v2_16way_ctx;
+
+bool init_lyra2rev2_16way_ctx()
 {
-   uint32_t vhash[8*8] __attribute__ ((aligned (128)));
+   keccak256_8way_init( &l2v2_16way_ctx.keccak );
-   uint32_t vhashA[8*8] __attribute__ ((aligned (64)));
+   cubehashInit( &l2v2_16way_ctx.cube, 256, 16, 32 );
-   uint32_t vhashB[8*8] __attribute__ ((aligned (64)));
+   skein256_8way_init( &l2v2_16way_ctx.skein );
+   bmw256_16way_init( &l2v2_16way_ctx.bmw );
+   return true;
+}
+
+void lyra2rev2_16way_hash( void *state, const void *input )
+{
+   uint32_t vhash[8*16] __attribute__ ((aligned (128)));
    uint32_t hash0[8] __attribute__ ((aligned (64)));
    uint32_t hash1[8] __attribute__ ((aligned (64)));
    uint32_t hash2[8] __attribute__ ((aligned (64)));
@@ -22,35 +40,60 @@ void lyra2rev2_8way_hash( void *state, const void *input )
    uint32_t hash5[8] __attribute__ ((aligned (64)));
    uint32_t hash6[8] __attribute__ ((aligned (64)));
    uint32_t hash7[8] __attribute__ ((aligned (64)));
-   lyra2v2_8way_ctx_holder ctx __attribute__ ((aligned (64)));
+   uint32_t hash8[8] __attribute__ ((aligned (64)));
-   memcpy( &ctx, &l2v2_8way_ctx, sizeof(l2v2_8way_ctx) );
+   uint32_t hash9[8] __attribute__ ((aligned (64)));
+   uint32_t hash10[8] __attribute__ ((aligned (64)));
+   uint32_t hash11[8] __attribute__ ((aligned (64)));
+   uint32_t hash12[8] __attribute__ ((aligned (64)));
+   uint32_t hash13[8] __attribute__ ((aligned (64)));
+   uint32_t hash14[8] __attribute__ ((aligned (64)));
+   uint32_t hash15[8] __attribute__ ((aligned (64)));
+   lyra2v2_16way_ctx_holder ctx __attribute__ ((aligned (64)));
+   memcpy( &ctx, &l2v2_16way_ctx, sizeof(l2v2_16way_ctx) );
 
-   blake256_8way_update( &ctx.blake, input + (64<<3), 16 );
+   blake256_16way_update( &ctx.blake, input + (64<<4), 16 );
-   blake256_8way_close( &ctx.blake, vhash );
+   blake256_16way_close( &ctx.blake, vhash );
 
-   rintrlv_8x32_8x64( vhashA, vhash, 256 );
+   dintrlv_16x32( hash0,  hash1,  hash2,  hash3,
+                  hash4,  hash5,  hash6,  hash7,
+                  hash8,  hash9,  hash10, hash11,
+                  hash12, hash13, hash14, hash15, vhash, 256 );
 
-   keccak256_8way_update( &ctx.keccak, vhashA, 32 );
+   intrlv_8x64( vhash, hash0, hash1, hash2, hash3,
+                       hash4, hash5, hash6, hash7, 256 );
+
+   keccak256_8way_update( &ctx.keccak, vhash, 32 );
    keccak256_8way_close( &ctx.keccak, vhash );
 
    dintrlv_8x64( hash0, hash1, hash2, hash3,
                  hash4, hash5, hash6, hash7, vhash, 256 );
+   intrlv_8x64( vhash, hash8,  hash9,  hash10, hash11,
+                       hash12, hash13, hash14, hash15, 256 );
 
-   cubehash_full( &ctx.cube, (byte*) hash0, 256, (const byte*) hash0, 32 );
+   keccak256_8way_init( &ctx.keccak );
-   cubehash_full( &ctx.cube, (byte*) hash1, 256, (const byte*) hash1, 32 );
+   keccak256_8way_update( &ctx.keccak, vhash, 32 );
-   cubehash_full( &ctx.cube, (byte*) hash2, 256, (const byte*) hash2, 32 );
+   keccak256_8way_close( &ctx.keccak, vhash );
-   cubehash_full( &ctx.cube, (byte*) hash3, 256, (const byte*) hash3, 32 );
+
-   cubehash_full( &ctx.cube, (byte*) hash4, 256, (const byte*) hash4, 32 );
+   dintrlv_8x64( hash8,  hash9,  hash10,  hash11,
-   cubehash_full( &ctx.cube, (byte*) hash5, 256, (const byte*) hash5, 32 );
+                 hash12, hash13, hash14, hash5, vhash, 256 );
-   cubehash_full( &ctx.cube, (byte*) hash6, 256, (const byte*) hash6, 32 );
+
-   cubehash_full( &ctx.cube, (byte*) hash7, 256, (const byte*) hash7, 32 );
+   cubehash_full( &ctx.cube, (byte*) hash0,  256, (const byte*) hash0,  32 );
+   cubehash_full( &ctx.cube, (byte*) hash1,  256, (const byte*) hash1,  32 );
+   cubehash_full( &ctx.cube, (byte*) hash2,  256, (const byte*) hash2,  32 );
+   cubehash_full( &ctx.cube, (byte*) hash3,  256, (const byte*) hash3,  32 );
+   cubehash_full( &ctx.cube, (byte*) hash4,  256, (const byte*) hash4,  32 );
+   cubehash_full( &ctx.cube, (byte*) hash5,  256, (const byte*) hash5,  32 );
+   cubehash_full( &ctx.cube, (byte*) hash6,  256, (const byte*) hash6,  32 );
+   cubehash_full( &ctx.cube, (byte*) hash7,  256, (const byte*) hash7,  32 );
+   cubehash_full( &ctx.cube, (byte*) hash8,  256, (const byte*) hash8,  32 );
+   cubehash_full( &ctx.cube, (byte*) hash9,  256, (const byte*) hash9,  32 );
+   cubehash_full( &ctx.cube, (byte*) hash10, 256, (const byte*) hash10, 32 );
+   cubehash_full( &ctx.cube, (byte*) hash11, 256, (const byte*) hash11, 32 );
+   cubehash_full( &ctx.cube, (byte*) hash12, 256, (const byte*) hash12, 32 );
+   cubehash_full( &ctx.cube, (byte*) hash13, 256, (const byte*) hash13, 32 );
+   cubehash_full( &ctx.cube, (byte*) hash14, 256, (const byte*) hash14, 32 );
+   cubehash_full( &ctx.cube, (byte*) hash15, 256, (const byte*) hash15, 32 );
 
-//   cube_4way_update_close( &ctx.cube, vhashA, vhashA, 32 );
-//   cube_4way_init( &ctx.cube, 256, 16, 32 );
-//   cube_4way_update_close( &ctx.cube, vhashB, vhashB, 32 );
-//
-//   dintrlv_4x128( hash0, hash1, hash2, hash3, vhashA, 256 );
-//   dintrlv_4x128( hash4, hash5, hash6, hash7, vhashB, 256 );
 
    intrlv_2x256( vhash, hash0, hash1, 256 );
    LYRA2REV2_2WAY( l2v2_wholeMatrix, vhash, 32, vhash, 32, 1, 4, 4 );
@@ -64,61 +107,127 @@ void lyra2rev2_8way_hash( void *state, const void *input )
    intrlv_2x256( vhash, hash6, hash7, 256 );
    LYRA2REV2_2WAY( l2v2_wholeMatrix, vhash, 32, vhash, 32, 1, 4, 4 );
    dintrlv_2x256( hash6, hash7, vhash, 256 );
+   intrlv_2x256( vhash, hash8, hash9, 256 );
+   LYRA2REV2_2WAY( l2v2_wholeMatrix, vhash, 32, vhash, 32, 1, 4, 4 );
+   dintrlv_2x256( hash8, hash9, vhash, 256 );
+   intrlv_2x256( vhash, hash10, hash11, 256 );
+   LYRA2REV2_2WAY( l2v2_wholeMatrix, vhash, 32, vhash, 32, 1, 4, 4 );
+   dintrlv_2x256( hash10, hash11, vhash, 256 );
+   intrlv_2x256( vhash, hash12, hash13, 256 );
+   LYRA2REV2_2WAY( l2v2_wholeMatrix, vhash, 32, vhash, 32, 1, 4, 4 );
+   dintrlv_2x256( hash12, hash13, vhash, 256 );
+   intrlv_2x256( vhash, hash14, hash15, 256 );
+   LYRA2REV2_2WAY( l2v2_wholeMatrix, vhash, 32, vhash, 32, 1, 4, 4 );
+   dintrlv_2x256( hash14, hash15, vhash, 256 );
 
-   intrlv_8x64( vhash, hash0, hash1, hash2, hash3, hash4, hash5, hash6,
+   intrlv_8x64( vhash, hash0, hash1, hash2, hash3,
-                hash7, 256 );
+                       hash4, hash5, hash6, hash7, 256 );
-
    skein256_8way_update( &ctx.skein, vhash, 32 );
    skein256_8way_close( &ctx.skein, vhash );
 
    dintrlv_8x64( hash0, hash1, hash2, hash3,
                  hash4, hash5, hash6, hash7, vhash, 256 );
+   intrlv_8x64( vhash, hash8,  hash9,  hash10, hash11, hash12,
+                       hash13, hash14, hash15, 256 );
 
-   cubehash_full( &ctx.cube, (byte*) hash0, 256, (const byte*) hash0, 32 );
+   skein256_8way_init( &ctx.skein );
-   cubehash_full( &ctx.cube, (byte*) hash1, 256, (const byte*) hash1, 32 );
+   skein256_8way_update( &ctx.skein, vhash, 32 );
-   cubehash_full( &ctx.cube, (byte*) hash2, 256, (const byte*) hash2, 32 );
+   skein256_8way_close( &ctx.skein, vhash );
-   cubehash_full( &ctx.cube, (byte*) hash3, 256, (const byte*) hash3, 32 );
-   cubehash_full( &ctx.cube, (byte*) hash4, 256, (const byte*) hash4, 32 );
-   cubehash_full( &ctx.cube, (byte*) hash5, 256, (const byte*) hash5, 32 );
-   cubehash_full( &ctx.cube, (byte*) hash6, 256, (const byte*) hash6, 32 );
-   cubehash_full( &ctx.cube, (byte*) hash7, 256, (const byte*) hash7, 32 );
 
-//   cube_4way_init( &ctx.cube, 256, 16, 32 );
+   dintrlv_8x64( hash8,  hash9,  hash10, hash11,
-//   cube_4way_update_close( &ctx.cube, vhashA, vhashA, 32 );
+                 hash12, hash13, hash14, hash15, vhash, 256 );
-//   cube_4way_init( &ctx.cube, 256, 16, 32 );
-//   cube_4way_update_close( &ctx.cube, vhashB, vhashB, 32 );
-//   
-//   dintrlv_4x128( hash0, hash1, hash2, hash3, vhashA, 256 );
-//   dintrlv_4x128( hash4, hash5, hash6, hash7, vhashB, 256 );
 
-   intrlv_8x32( vhash, hash0, hash1, hash2, hash3, hash4, hash5, hash6,
+   
-                hash7, 256 );
+   cubehash_full( &ctx.cube, (byte*) hash0,  256, (const byte*) hash0, 32 );
+   cubehash_full( &ctx.cube, (byte*) hash1,  256, (const byte*) hash1, 32 );
+   cubehash_full( &ctx.cube, (byte*) hash2,  256, (const byte*) hash2, 32 );
+   cubehash_full( &ctx.cube, (byte*) hash3,  256, (const byte*) hash3, 32 );
+   cubehash_full( &ctx.cube, (byte*) hash4,  256, (const byte*) hash4, 32 );
+   cubehash_full( &ctx.cube, (byte*) hash5,  256, (const byte*) hash5, 32 );
+   cubehash_full( &ctx.cube, (byte*) hash6,  256, (const byte*) hash6, 32 );
+   cubehash_full( &ctx.cube, (byte*) hash7,  256, (const byte*) hash7, 32 );
+   cubehash_full( &ctx.cube, (byte*) hash8,  256, (const byte*) hash8,  32 );
+   cubehash_full( &ctx.cube, (byte*) hash9,  256, (const byte*) hash9,  32 );
+   cubehash_full( &ctx.cube, (byte*) hash10, 256, (const byte*) hash10, 32 );
+   cubehash_full( &ctx.cube, (byte*) hash11, 256, (const byte*) hash11, 32 );
+   cubehash_full( &ctx.cube, (byte*) hash12, 256, (const byte*) hash12, 32 );
+   cubehash_full( &ctx.cube, (byte*) hash13, 256, (const byte*) hash13, 32 );
+   cubehash_full( &ctx.cube, (byte*) hash14, 256, (const byte*) hash14, 32 );
+   cubehash_full( &ctx.cube, (byte*) hash15, 256, (const byte*) hash15, 32 );
 
-   bmw256_8way_update( &ctx.bmw, vhash, 32 );
+   intrlv_16x32( vhash, hash0,  hash1,  hash2,  hash3,
-   bmw256_8way_close( &ctx.bmw, state );
+                        hash4,  hash5,  hash6,  hash7,
+                        hash8,  hash9,  hash10, hash11,
+                        hash12, hash13, hash14, hash15, 256 );
+
+   bmw256_16way_update( &ctx.bmw, vhash, 32 );
+   bmw256_16way_close( &ctx.bmw, state );
 }
-#endif
 
+int scanhash_lyra2rev2_16way( struct work *work, const uint32_t max_nonce,
+                             uint64_t *hashes_done, struct thr_info *mythr )
+{
+   uint32_t hash[8*16] __attribute__ ((aligned (128)));
+   uint32_t vdata[20*16] __attribute__ ((aligned (64)));
+   uint32_t *hashd7 = &hash[7*16];
+   uint32_t lane_hash[8] __attribute__ ((aligned (32)));
+   uint32_t *pdata = work->data;
+   uint32_t *ptarget = work->target;
+   const uint32_t first_nonce = pdata[19];
+   const uint32_t last_nonce = max_nonce - 16;
+   uint32_t n = first_nonce;
+   const uint32_t targ32 = ptarget[7];
+   __m512i  *noncev = (__m512i*)vdata + 19;
+   const int thr_id = mythr->id;
+   const bool bench = opt_benchmark;
 
+   if ( bench )  ptarget[7] = 0x0000ff;
 
+   mm512_bswap32_intrlv80_16x32( vdata, pdata );
+   *noncev = _mm512_set_epi32( n+15, n+14, n+13, n+12, n+11, n+10, n+ 9, n+ 8,
+                               n+ 7, n+ 6, n+ 5, n+ 4, n+ 3, n+ 2, n+ 1, n );
+   blake256_16way_init( &l2v2_16way_ctx.blake );
+   blake256_16way_update( &l2v2_16way_ctx.blake, vdata, 64 );
 
-#if defined (LYRA2REV2_8WAY)
+   do
+   {
+      lyra2rev2_16way_hash( hash, vdata );
+
+      for ( int lane = 0; lane < 16; lane++ )
+      if ( unlikely( hashd7[lane] <= targ32 ) )
+      {
+         extr_lane_16x32( lane_hash, hash, lane, 256 );
+         if ( likely( valid_hash( lane_hash, ptarget ) && !bench ) )
+         {
+             pdata[19] = bswap_32( n + lane );
+             submit_lane_solution( work, lane_hash, mythr, lane );
+         }
+      }
+      *noncev = _mm512_add_epi32( *noncev, m512_const1_32( 16 ) );
+      n += 16;
+   } while ( likely( (n < last_nonce) && !work_restart[thr_id].restart ) );
+   pdata[19] = n;
+   *hashes_done = n - first_nonce;
+   return 0;
+}
+
+#elif defined (LYRA2REV2_8WAY)
 
 typedef struct {
    blake256_8way_context     blake;
-   keccak256_8way_context    keccak;
+   keccak256_4way_context    keccak;
    cubehashParam             cube;
-   skein256_8way_context     skein;
+   skein256_4way_context     skein;
-   bmw256_8way_context          bmw;
+   bmw256_8way_context       bmw;
 } lyra2v2_8way_ctx_holder __attribute__ ((aligned (64)));
 
 static lyra2v2_8way_ctx_holder l2v2_8way_ctx;
 
 bool init_lyra2rev2_8way_ctx()
 {
-   keccak256_8way_init( &l2v2_8way_ctx.keccak );
+   keccak256_4way_init( &l2v2_8way_ctx.keccak );
    cubehashInit( &l2v2_8way_ctx.cube, 256, 16, 32 );
-   skein256_8way_init( &l2v2_8way_ctx.skein );
+   skein256_4way_init( &l2v2_8way_ctx.skein );
    bmw256_8way_init( &l2v2_8way_ctx.bmw );
    return true;
 }
@@ -126,7 +235,6 @@ bool init_lyra2rev2_8way_ctx()
 void lyra2rev2_8way_hash( void *state, const void *input )
 {
    uint32_t vhash[8*8] __attribute__ ((aligned (128)));
-   uint32_t vhashA[8*8] __attribute__ ((aligned (64)));
    uint32_t hash0[8] __attribute__ ((aligned (64)));
    uint32_t hash1[8] __attribute__ ((aligned (64)));
    uint32_t hash2[8] __attribute__ ((aligned (64)));
@@ -141,14 +249,19 @@ void lyra2rev2_8way_hash( void *state, const void *input )
    blake256_8way_update( &ctx.blake, input + (64<<3), 16 );
    blake256_8way_close( &ctx.blake, vhash );
 
-   rintrlv_8x32_8x64( vhashA, vhash, 256 );
+   dintrlv_8x32( hash0, hash1, hash2, hash3,
-
-   keccak256_8way_update( &ctx.keccak, vhashA, 32 );
-   keccak256_8way_close( &ctx.keccak, vhash );
-
-   dintrlv_8x64( hash0, hash1, hash2, hash3,
                  hash4, hash5, hash6, hash7, vhash, 256 );
 
+   intrlv_4x64( vhash, hash0, hash1, hash2, hash3, 256 );
+   keccak256_4way_update( &ctx.keccak, vhash, 32 );
+   keccak256_4way_close( &ctx.keccak, vhash );
+   dintrlv_4x64( hash0, hash1, hash2, hash3, vhash, 256 );
+   intrlv_4x64( vhash, hash4, hash5, hash6, hash7, 256 );
+   keccak256_4way_init( &ctx.keccak );
+   keccak256_4way_update( &ctx.keccak, vhash, 32 );
+   keccak256_4way_close( &ctx.keccak, vhash );
+   dintrlv_4x64( hash4, hash5, hash6, hash7, vhash, 256 );
+
    cubehash_full( &ctx.cube, (byte*) hash0, 256, (const byte*) hash0, 32 );
    cubehash_full( &ctx.cube, (byte*) hash1, 256, (const byte*) hash1, 32 );
    cubehash_full( &ctx.cube, (byte*) hash2, 256, (const byte*) hash2, 32 );
@@ -158,27 +271,25 @@ void lyra2rev2_8way_hash( void *state, const void *input )
    cubehash_full( &ctx.cube, (byte*) hash6, 256, (const byte*) hash6, 32 );
    cubehash_full( &ctx.cube, (byte*) hash7, 256, (const byte*) hash7, 32 );
 
-   intrlv_2x256( vhash, hash0, hash1, 256 );
-   LYRA2REV2_2WAY( l2v2_wholeMatrix, vhash, 32, vhash, 32, 1, 4, 4 );
-   dintrlv_2x256( hash0, hash1, vhash, 256 );
-   intrlv_2x256( vhash, hash2, hash3, 256 );
-   LYRA2REV2_2WAY( l2v2_wholeMatrix, vhash, 32, vhash, 32, 1, 4, 4 );
-   dintrlv_2x256( hash2, hash3, vhash, 256 );
-   intrlv_2x256( vhash, hash4, hash5, 256 );
-   LYRA2REV2_2WAY( l2v2_wholeMatrix, vhash, 32, vhash, 32, 1, 4, 4 );
-   dintrlv_2x256( hash4, hash5, vhash, 256 );
-   intrlv_2x256( vhash, hash6, hash7, 256 );
-   LYRA2REV2_2WAY( l2v2_wholeMatrix, vhash, 32, vhash, 32, 1, 4, 4 );
-   dintrlv_2x256( hash6, hash7, vhash, 256 );
 
-   intrlv_8x64( vhash, hash0, hash1, hash2, hash3, hash4, hash5, hash6,
+   LYRA2REV2( l2v2_wholeMatrix, hash0, 32, hash0, 32, hash0, 32, 1, 4, 4 );
-                hash7, 256 );
+   LYRA2REV2( l2v2_wholeMatrix, hash1, 32, hash1, 32, hash1, 32, 1, 4, 4 );
-
+   LYRA2REV2( l2v2_wholeMatrix, hash2, 32, hash2, 32, hash2, 32, 1, 4, 4 );
-   skein256_8way_update( &ctx.skein, vhash, 32 );
+   LYRA2REV2( l2v2_wholeMatrix, hash3, 32, hash3, 32, hash3, 32, 1, 4, 4 );
-   skein256_8way_close( &ctx.skein, vhash );
+   LYRA2REV2( l2v2_wholeMatrix, hash4, 32, hash4, 32, hash4, 32, 1, 4, 4 );
-
+   LYRA2REV2( l2v2_wholeMatrix, hash5, 32, hash5, 32, hash5, 32, 1, 4, 4 );
-   dintrlv_8x64( hash0, hash1, hash2, hash3,
+   LYRA2REV2( l2v2_wholeMatrix, hash6, 32, hash6, 32, hash6, 32, 1, 4, 4 );
-                 hash4, hash5, hash6, hash7, vhash, 256 );
+   LYRA2REV2( l2v2_wholeMatrix, hash7, 32, hash7, 32, hash7, 32, 1, 4, 4 );
+   
+   intrlv_4x64( vhash, hash0, hash1, hash2, hash3, 256 );
+   skein256_4way_update( &ctx.skein, vhash, 32 );
+   skein256_4way_close( &ctx.skein, vhash );
+   dintrlv_4x64( hash0, hash1, hash2, hash3, vhash, 256 );
+   intrlv_4x64( vhash, hash4, hash5, hash6, hash7, 256 );
+   skein256_4way_init( &ctx.skein );
+   skein256_4way_update( &ctx.skein, vhash, 32 );
+   skein256_4way_close( &ctx.skein, vhash );
+   dintrlv_4x64( hash4, hash5, hash6, hash7, vhash, 256 );
 
    cubehash_full( &ctx.cube, (byte*) hash0, 256, (const byte*) hash0, 32 );
    cubehash_full( &ctx.cube, (byte*) hash1, 256, (const byte*) hash1, 32 );
@@ -189,8 +300,8 @@ void lyra2rev2_8way_hash( void *state, const void *input )
    cubehash_full( &ctx.cube, (byte*) hash6, 256, (const byte*) hash6, 32 );
    cubehash_full( &ctx.cube, (byte*) hash7, 256, (const byte*) hash7, 32 );
 
-   intrlv_8x32( vhash, hash0, hash1, hash2, hash3, hash4, hash5, hash6, 
+   intrlv_8x32( vhash, hash0, hash1, hash2, hash3,
-                hash7, 256 );
+                       hash4, hash5, hash6, hash7, 256 );
 
    bmw256_8way_update( &ctx.bmw, vhash, 32 );
    bmw256_8way_close( &ctx.bmw, state );
@@ -223,7 +334,6 @@ int scanhash_lyra2rev2_8way( struct work *work, const uint32_t max_nonce,
    do
    {
       lyra2rev2_8way_hash( hash, vdata );
-      pdata[19] = n;
 
       for ( int lane = 0; lane < 8; lane++ )
       if ( unlikely( hashd7[lane] <= targ32 ) )
@@ -243,6 +353,9 @@ int scanhash_lyra2rev2_8way( struct work *work, const uint32_t max_nonce,
    return 0;
 }
 
+#endif
+
+/*
 #elif defined (LYRA2REV2_4WAY)
 
 typedef struct {
@@ -367,3 +480,4 @@ int scanhash_lyra2rev2_4way( struct work *work, uint32_t max_nonce,
 }
 
 #endif
+*/

configure

@@ -1,6 +1,6 @@
 #! /bin/sh
 # Guess values for system-dependent variables and create Makefiles.
-# Generated by GNU Autoconf 2.69 for cpuminer-opt 3.12.3.1.
+# Generated by GNU Autoconf 2.69 for cpuminer-opt 3.12.4.3.
 #
 #
 # Copyright (C) 1992-1996, 1998-2012 Free Software Foundation, Inc.
@@ -577,8 +577,8 @@ MAKEFLAGS=
 # Identity of this package.
 PACKAGE_NAME='cpuminer-opt'
 PACKAGE_TARNAME='cpuminer-opt'
-PACKAGE_VERSION='3.12.3.1'
+PACKAGE_VERSION='3.12.4.3'
-PACKAGE_STRING='cpuminer-opt 3.12.3.1'
+PACKAGE_STRING='cpuminer-opt 3.12.4.3'
 PACKAGE_BUGREPORT=''
 PACKAGE_URL=''
 
@@ -1332,7 +1332,7 @@ if test "$ac_init_help" = "long"; then
   # Omit some internal or obsolete options to make the list less imposing.
   # This message is too long to be a string in the A/UX 3.1 sh.
   cat <<_ACEOF
-\`configure' configures cpuminer-opt 3.12.3.1 to adapt to many kinds of systems.
+\`configure' configures cpuminer-opt 3.12.4.3 to adapt to many kinds of systems.
 
 Usage: $0 [OPTION]... [VAR=VALUE]...
 
@@ -1404,7 +1404,7 @@ fi
 
 if test -n "$ac_init_help"; then
   case $ac_init_help in
-     short | recursive ) echo "Configuration of cpuminer-opt 3.12.3.1:";;
+     short | recursive ) echo "Configuration of cpuminer-opt 3.12.4.3:";;
    esac
   cat <<\_ACEOF
 
@@ -1509,7 +1509,7 @@ fi
 test -n "$ac_init_help" && exit $ac_status
 if $ac_init_version; then
   cat <<\_ACEOF
-cpuminer-opt configure 3.12.3.1
+cpuminer-opt configure 3.12.4.3
 generated by GNU Autoconf 2.69
 
 Copyright (C) 2012 Free Software Foundation, Inc.
@@ -2012,7 +2012,7 @@ cat >config.log <<_ACEOF
 This file contains any messages produced by compilers while
 running configure, to aid debugging if configure makes a mistake.
 
-It was created by cpuminer-opt $as_me 3.12.3.1, which was
+It was created by cpuminer-opt $as_me 3.12.4.3, which was
 generated by GNU Autoconf 2.69.  Invocation command line was
 
   $ $0 $@
@@ -2993,7 +2993,7 @@ fi
 
 # Define the identity of the package.
  PACKAGE='cpuminer-opt'
- VERSION='3.12.3.1'
+ VERSION='3.12.4.3'
 
 
 cat >>confdefs.h <<_ACEOF
@@ -6690,7 +6690,7 @@ cat >>$CONFIG_STATUS <<\_ACEOF || ac_write_fail=1
 # report actual input values of CONFIG_FILES etc. instead of their
 # values after options handling.
 ac_log="
-This file was extended by cpuminer-opt $as_me 3.12.3.1, which was
+This file was extended by cpuminer-opt $as_me 3.12.4.3, which was
 generated by GNU Autoconf 2.69.  Invocation command line was
 
   CONFIG_FILES    = $CONFIG_FILES
@@ -6756,7 +6756,7 @@ _ACEOF
 cat >>$CONFIG_STATUS <<_ACEOF || ac_write_fail=1
 ac_cs_config="`$as_echo "$ac_configure_args" | sed 's/^ //; s/[\\""\`\$]/\\\\&/g'`"
 ac_cs_version="\\
-cpuminer-opt config.status 3.12.3.1
+cpuminer-opt config.status 3.12.4.3
 configured by $0, generated by GNU Autoconf 2.69,
   with options \\"\$ac_cs_config\\"
 

configure.ac

@@ -1,4 +1,4 @@
-AC_INIT([cpuminer-opt], [3.12.3.1])
+AC_INIT([cpuminer-opt], [3.12.4.3])
 
 AC_PREREQ([2.59c])
 AC_CANONICAL_SYSTEM

cpu-miner.c

@@ -431,68 +431,71 @@ static bool work_decode( const json_t *val, struct work *work )
 static const char *info_req =
 "{\"method\": \"getmininginfo\", \"params\": [], \"id\":8}\r\n";
 
-static bool get_mininginfo(CURL *curl, struct work *work)
+static bool get_mininginfo( CURL *curl, struct work *work )
 {
-	if (have_stratum || !allow_mininginfo)
+	if ( have_stratum || !allow_mininginfo )
 		return false;
 
 	int curl_err = 0;
-	json_t *val = json_rpc_call(curl, rpc_url, rpc_userpass, info_req, &curl_err, 0);
+	json_t *val = json_rpc_call( curl, rpc_url, rpc_userpass, info_req,
+                                &curl_err, 0 );
 
-	if (!val && curl_err == -1) {
+	if ( !val && curl_err == -1 )
+   {
 		allow_mininginfo = false;
-		if (opt_debug) {
+		if ( opt_debug )
-			applog(LOG_DEBUG, "getmininginfo not supported");
+			applog( LOG_DEBUG, "getmininginfo not supported" );
-		}
 		return false;
 	}
-	else
+
-        {
+   json_t *res = json_object_get( val, "result" );
-	   json_t *res = json_object_get(val, "result");
+   // "blocks": 491493 (= current work height - 1)
-	   // "blocks": 491493 (= current work height - 1)
+   // "difficulty": 0.99607860999999998
-	   // "difficulty": 0.99607860999999998
+   // "networkhashps": 56475980
-	   // "networkhashps": 56475980
+   if ( res )
-	   if (res)
+   {
-           {
+  		json_t *key = json_object_get( res, "difficulty" );
-		json_t *key = json_object_get(res, "difficulty");
+   	if ( key )
-		if (key) {
+      {
-			if (json_is_object(key))
+	   	if ( json_is_object( key ) )
-				key = json_object_get(key, "proof-of-work");
+		   	key = json_object_get( key, "proof-of-work" );
-			if (json_is_real(key))
+		   if ( json_is_real( key ) )
-				net_diff = json_real_value(key);
+			   net_diff = json_real_value( key );
-		}
-		key = json_object_get(res, "networkhashps");
-		if (key && json_is_integer(key)) {
-			net_hashrate = (double) json_integer_value(key);
-		}
-		key = json_object_get(res, "blocks");
-		if (key && json_is_integer(key)) {
-			net_blocks = json_integer_value(key);
-		}
-		if (!work->height)
-                {
-		   // complete missing data from getwork
-		   work->height = (uint32_t) net_blocks + 1;
-		   if (work->height > g_work.height)
-                   {
-			restart_threads();
-			if (!opt_quiet) {
-			   char netinfo[64] = { 0 };
-			   char srate[32] = { 0 };
-			   sprintf(netinfo, "diff %.2f", net_diff);
-			   if (net_hashrate) {
-				format_hashrate(net_hashrate, srate);
-				strcat(netinfo, ", net ");
-				strcat(netinfo, srate);
-			   }
-			   applog(LOG_BLUE, "%s block %d, %s",
-				algo_names[opt_algo], work->height, netinfo);
-			}
-		   }
-		}
 	   }
+
+      key = json_object_get( res, "networkhashps" );
+	   if ( key && json_is_integer( key ) )
+		   net_hashrate = (double) json_integer_value( key );
+
+      key = json_object_get( res, "blocks" );
+	   if ( key && json_is_integer( key ) )
+		  	net_blocks = json_integer_value( key );
+
+      if ( !work->height )
+      {
+	      // complete missing data from getwork
+	      work->height = (uint32_t) net_blocks + 1;
+	      if ( work->height > g_work.height )
+         {
+            restart_threads();
+		      if ( !opt_quiet )
+            {
+		         char netinfo[64] = { 0 };
+		         char srate[32] = { 0 };
+		         sprintf( netinfo, "diff %.2f", net_diff );
+		         if ( net_hashrate )
+               {
+	               format_hashrate( net_hashrate, srate );
+                  strcat( netinfo, ", net " );
+			         strcat( netinfo, srate );
+		         }
+		         applog( LOG_BLUE, "%s block %d, %s",
+			                algo_names[opt_algo], work->height, netinfo );
+		      }
+		   } 
+	   }  // res
 	}
-	json_decref(val);
+	json_decref( val );
 	return true;
 }
 
@@ -1059,25 +1062,39 @@ static int share_result( int result, struct work *null_work,
    if ( likely( result ) )
    {
       accepted_share_count++;
+      sprintf( sres, "S%d", stale_share_count );
+      sprintf( rres, "R%d", rejected_share_count );
       if unlikely( ( my_stats.net_diff > 0. )
                 && ( my_stats.share_diff >= net_diff ) )
       {
          solved = true;
          solved_block_count++;
+         sprintf( bres, "BLOCK SOLVED %d", solved_block_count );
+         sprintf( ares, "A%d", accepted_share_count );
+      }
+      else
+      {
+         sprintf( bres, "B%d", solved_block_count );
+         sprintf( ares, "Accepted %d", accepted_share_count );
       }
    }
    else
    {
+     sprintf( ares, "A%d", accepted_share_count );
+     sprintf( bres, "B%d", solved_block_count );
      if ( reason && strstr( reason, "Invalid job id" ) )
      {
         stale = true;
         stale_share_count++;
+        sprintf( sres, "Stale %d", stale_share_count );
+        sprintf( rres, "R%d", rejected_share_count );
      }
      else
      {
         rejected_share_count++;
+        sprintf( sres, "S%d", stale_share_count );
+        sprintf( rres, "Rejected %d" , rejected_share_count );
         lowdiff_debug = true;
-  
      }
    }
 
@@ -1103,6 +1120,7 @@ static int share_result( int result, struct work *null_work,
 
    pthread_mutex_unlock( &stats_lock );
 
+/*
    if ( likely( result ) )
    {
      if ( unlikely( solved ) )
@@ -1133,6 +1151,7 @@ static int share_result( int result, struct work *null_work,
         sprintf( rres, "Rejected %d" , rejected_share_count );
      }
    } 
+*/
 
    if ( use_colors )
    {
@@ -1152,10 +1171,16 @@ static int share_result( int result, struct work *null_work,
            my_stats.share_count, acol, ares, scol, sres, rcol, rres, bcol,
            bres, share_time, latency );
 
-   if ( have_stratum && !opt_quiet )
+   if ( !opt_quiet )
-      applog2( LOG_NOTICE, "Diff %.5g (%.3g%), %sBlock %d, %sJob %s" CL_WHT,
+   {
+      if ( have_stratum )
+         applog2( LOG_NOTICE, "Diff %.5g (%.3g%), %sBlock %d, %sJob %s" CL_WHT,
                my_stats.share_diff, share_ratio, bcol, stratum.block_height,
                scol, my_stats.job_id );
+      else
+         applog2( LOG_NOTICE, "Diff %.5g (%.3g%), %sBlock %d" CL_WHT,
+               my_stats.share_diff, share_ratio, bcol, stratum.block_height );
+   }
 
    if ( unlikely( reason && !result ) )
    {
@@ -1188,6 +1213,9 @@ static int share_result( int result, struct work *null_work,
    return 1;
 }
 
+static const char *json_submit_req =
+   "{\"method\": \"mining.submit\", \"params\": [\"%s\", \"%s\", \"%s\", \"%s\", \"%s\"], \"id\":4}";
+
 void std_le_build_stratum_request( char *req, struct work *work )
 {
    unsigned char *xnonce2str;
@@ -1198,9 +1226,8 @@ void std_le_build_stratum_request( char *req, struct work *work )
    bin2hex( ntimestr, (char*)(&ntime), sizeof(uint32_t) );
    bin2hex( noncestr, (char*)(&nonce), sizeof(uint32_t) );
    xnonce2str = abin2hex( work->xnonce2, work->xnonce2_len );
-   snprintf( req, JSON_BUF_LEN,
+   snprintf( req, JSON_BUF_LEN, json_submit_req, rpc_user, work->job_id,
-        "{\"method\": \"mining.submit\", \"params\": [\"%s\", \"%s\", \"%s\", \"%s\", \"%s\"], \"id\":4}",
+             xnonce2str, ntimestr, noncestr );
-         rpc_user, work->job_id, xnonce2str, ntimestr, noncestr );
    free( xnonce2str );
 }
 
@@ -1215,12 +1242,14 @@ void std_be_build_stratum_request( char *req, struct work *work )
    bin2hex( ntimestr, (char*)(&ntime), sizeof(uint32_t) );
    bin2hex( noncestr, (char*)(&nonce), sizeof(uint32_t) );
    xnonce2str = abin2hex( work->xnonce2, work->xnonce2_len );
-   snprintf( req, JSON_BUF_LEN,
+   snprintf( req, JSON_BUF_LEN, json_submit_req, rpc_user, work->job_id,
-        "{\"method\": \"mining.submit\", \"params\": [\"%s\", \"%s\", \"%s\", \"%s\", \"%s\"], \"id\":4}",
+             xnonce2str, ntimestr, noncestr );
-         rpc_user, work->job_id, xnonce2str, ntimestr, noncestr );
    free( xnonce2str );
 }
 
+static const char *json_getwork_req = 
+  "{\"method\": \"getwork\", \"params\": [\"%s\"], \"id\":4}\r\n";
+
 bool std_le_submit_getwork_result( CURL *curl, struct work *work )
 {
    char req[JSON_BUF_LEN];
@@ -1237,8 +1266,7 @@ bool std_le_submit_getwork_result( CURL *curl, struct work *work )
       return false;
    }
    // build JSON-RPC request 
-   snprintf( req, JSON_BUF_LEN,
+   snprintf( req, JSON_BUF_LEN, json_getwork_req, gw_str );
-     "{\"method\": \"getwork\", \"params\": [\"%s\"], \"id\":4}\r\n", gw_str );
    free( gw_str );
    // issue JSON-RPC request 
    val = json_rpc_call( curl, rpc_url, rpc_userpass, req, NULL, 0 );
@@ -1271,8 +1299,7 @@ bool std_be_submit_getwork_result( CURL *curl, struct work *work )
       return false;
    }
    // build JSON-RPC request 
-   snprintf( req, JSON_BUF_LEN,
+   snprintf( req, JSON_BUF_LEN, json_getwork_req, gw_str );
-     "{\"method\": \"getwork\", \"params\": [\"%s\"], \"id\":4}\r\n", gw_str );
    free( gw_str );
    // issue JSON-RPC request 
    val = json_rpc_call( curl, rpc_url, rpc_userpass, req, NULL, 0 );
@@ -1327,25 +1354,27 @@ char* std_malloc_txs_request( struct work *work )
 
 static bool submit_upstream_work( CURL *curl, struct work *work )
 {
+
    /* pass if the previous hash is not the current previous hash */
+/* Submit anyway, discardring here messes up the stats
    if ( !submit_old && memcmp( &work->data[1], &g_work.data[1], 32 ) )
    {
-      if (opt_debug)
+      applog( LOG_WARNING, "Stale work detected, discarding" );
-         applog(LOG_DEBUG, "DEBUG: stale work detected, discarding");
       return true;
    }
 
    if ( !have_stratum && allow_mininginfo )
    {
-      struct work wheight;
+      struct work mining_info;
-      get_mininginfo( curl, &wheight );
+      get_mininginfo( curl, &mining_info );
-      if ( work->height && work->height <= net_blocks )
+      if ( work->height < mining_info.height )
       {
-         if (opt_debug)
+ 	      applog( LOG_WARNING, "Block %u was already solved, current block %d",
- 	        applog(LOG_WARNING, "block %u was already solved", work->height);
+                               work->height, mining_info.height );
 	      return true;
       }
    }
+*/
 
    if ( have_stratum )
    {
@@ -1477,6 +1506,37 @@ start:
    json_decref( val );
    // store work height in solo
    get_mininginfo(curl, work);
+
+   applog( LOG_BLUE, "%s %s block %d, diff %.5g", algo_names[ opt_algo ],
+                      short_url, work->height, net_diff );
+
+   if ( !opt_quiet && net_diff && net_hashrate )
+   {
+      double miner_hr = 0.;
+      pthread_mutex_lock( &stats_lock );
+
+      for ( int i = 0; i < opt_n_threads; i++ )
+         miner_hr += thr_hashrates[i];
+      global_hashrate = miner_hr;
+
+      pthread_mutex_unlock( &stats_lock );
+
+      if ( miner_hr )
+      {
+         char net_hr_units[4] = {0};
+         char miner_hr_units[4] = {0};
+         char net_ttf[32];
+         char miner_ttf[32];
+
+         sprintf_et( net_ttf, net_diff * diff_to_hash / net_hashrate );
+         sprintf_et( miner_ttf, net_diff * diff_to_hash / miner_hr );
+         scale_hash_for_display ( &miner_hr, miner_hr_units );
+         scale_hash_for_display ( &net_hashrate, net_hr_units );
+         applog2(LOG_INFO, "Miner TTF @ %.2f %sh/s %s, net TTF @ %.2f %sh/s %s",
+                             miner_hr, miner_hr_units, miner_ttf,
+                             net_hashrate, net_hr_units, net_ttf );
+      }
+   }
    return rc;
 }
 
@@ -1523,6 +1583,8 @@ static bool workio_get_work(struct workio_cmd *wc, CURL *curl)
 	sleep(opt_fail_pause);
    }
 
+   report_summary_log( false );
+
    /* send work to requesting thread */
    if (!tq_push(wc->thr->q, ret_work))
 	free(ret_work);
@@ -1698,8 +1760,8 @@ void work_set_target_ratio( struct work* work, const void *hash )
    share_stats[ s_put_ptr ].net_diff = net_diff;
    share_stats[ s_put_ptr ].stratum_diff = stratum_diff;
    share_stats[ s_put_ptr ].target_diff = work->targetdiff;
-   ( (uint64_t*)share_stats[ s_put_ptr ].job_id )[3] = 0;
+   if ( have_stratum )
-   strncpy( share_stats[ s_put_ptr ].job_id, work->job_id, 30 );
+      strncpy( share_stats[ s_put_ptr ].job_id, work->job_id, 30 );
    s_put_ptr = stats_ptr_incr( s_put_ptr );
 
    pthread_mutex_unlock( &stats_lock );
@@ -1712,24 +1774,31 @@ bool submit_solution( struct work *work, const void *hash,
   {
      submitted_share_count++;
      work_set_target_ratio( work, hash );
-     if ( !opt_quiet )
-        applog( LOG_NOTICE, "%d submitted by thread %d, job %s",
-            submitted_share_count, thr->id, work->job_id );
 
-if ( lowdiff_debug )
+     if ( !opt_quiet )
-{
+     {
-   uint32_t* h = (uint32_t*)hash;
+        if ( have_stratum )
-   uint32_t* t = (uint32_t*)work->target;
+           applog( LOG_NOTICE, "%d submitted by thread %d, job %s",
-   applog(LOG_INFO,"Hash[7:0]: %08x %08x %08x %08x %08x %08x %08x %08x",
+               submitted_share_count, thr->id, work->job_id );
-                              h[7],h[6],h[5],h[4],h[3],h[2],h[1],h[0]);
+        else
-   applog(LOG_INFO,"Targ[7:0]: %08x %08x %08x %08x %08x %08x %08x %08x",
+           applog( LOG_NOTICE, "%d submitted by thread %d",
-                              t[7],t[6],t[5],t[4],t[3],t[2],t[1],t[0]);
+               submitted_share_count, thr->id );
-}
+     }
-    return true;
+
+     if ( lowdiff_debug )
+     {
+        uint32_t* h = (uint32_t*)hash;
+        uint32_t* t = (uint32_t*)work->target;
+        applog(LOG_INFO,"Hash[7:0]: %08x %08x %08x %08x %08x %08x %08x %08x",
+                                    h[7],h[6],h[5],h[4],h[3],h[2],h[1],h[0]);
+        applog(LOG_INFO,"Targ[7:0]: %08x %08x %08x %08x %08x %08x %08x %08x",
+                                    t[7],t[6],t[5],t[4],t[3],t[2],t[1],t[0]);
+     }
+     return true;
   }
   else
-     applog( LOG_WARNING, "%d failed to submit share.",
+     applog( LOG_WARNING, "%d failed to submit share thread %d.",
-             submitted_share_count );
+             submitted_share_count, thr->id );
   return false;
 }
 
@@ -1740,26 +1809,31 @@ bool submit_lane_solution( struct work *work, const void *hash,
   {
      submitted_share_count++;
      work_set_target_ratio( work, hash );
+
      if ( !opt_quiet )
-        applog( LOG_NOTICE, "%d submitted by thread %d, lane %d, job %s",
+     {
-            submitted_share_count, thr->id, lane, work->job_id );
+        if ( have_stratum )
+           applog( LOG_NOTICE, "%d submitted by thread %d, lane %d, job %s",
+               submitted_share_count, thr->id, lane, work->job_id );
+        else
+           applog( LOG_NOTICE, "%d submitted by thread %d, lane %d",
+               submitted_share_count, thr->id, lane );
+     }
 
-if ( lowdiff_debug )
+     if ( lowdiff_debug )
-{
+     {
-   uint32_t* h = (uint32_t*)hash;
+        uint32_t* h = (uint32_t*)hash;
-   uint32_t* t = (uint32_t*)work->target;
+        uint32_t* t = (uint32_t*)work->target;
-   applog(LOG_INFO,"Hash[7:0]: %08x %08x %08x %08x %08x %08x %08x %08x",
+        applog(LOG_INFO,"Hash[7:0]: %08x %08x %08x %08x %08x %08x %08x %08x",
-                              h[7],h[6],h[5],h[4],h[3],h[2],h[1],h[0]);
+                                    h[7],h[6],h[5],h[4],h[3],h[2],h[1],h[0]);
-   applog(LOG_INFO,"Targ[7:0]: %08x %08x %08x %08x %08x %08x %08x %08x",
+         applog(LOG_INFO,"Targ[7:0]: %08x %08x %08x %08x %08x %08x %08x %08x",
-                              t[7],t[6],t[5],t[4],t[3],t[2],t[1],t[0]);
+                                     t[7],t[6],t[5],t[4],t[3],t[2],t[1],t[0]);
-}
+    }
-
+    return true;
-
-     return true;
   }
   else
-     applog( LOG_WARNING, "%d failed to submit share.",
+     applog( LOG_WARNING, "%d failed to submit share, thread %d, lane %d.",
-          submitted_share_count );
+          submitted_share_count, thr->id, lane );
   return false;
 }
 
@@ -1853,12 +1927,15 @@ void std_get_new_work( struct work* work, struct work* g_work, int thr_id,
                      uint32_t *end_nonce_ptr )
 {
    uint32_t *nonceptr = work->data + algo_gate.nonce_index;
+   bool force_new_work = false; 
 
-   bool force_new_work = work->job_id ? strtoul(   work->job_id, NULL, 16 ) !=
+   if ( have_stratum ) 
-                                        strtoul( g_work->job_id, NULL, 16 )
+      force_new_work = work->job_id ?    strtoul(   work->job_id, NULL, 16 )
-                                      : true;
+                                      != strtoul( g_work->job_id, NULL, 16 )
+                                     : false;
 
-   if ( force_new_work || *nonceptr >= *end_nonce_ptr )
+   if ( force_new_work || ( *nonceptr >= *end_nonce_ptr )
+     || memcmp( work->data, g_work->data, algo_gate.work_cmp_size ) )
    {
      work_free( work );
      work_copy( work, g_work );
@@ -2036,12 +2113,28 @@ static void *miner_thread( void *userdata )
           sleep(5);
 	       continue;
        }
-       // adjust max_nonce to meet target scan time
+
+// LP_SCANTIME overrides opt_scantime option, is this right?
+
+       // adjust max_nonce to meet target scan time. Startum and longpoll
+       // can go longer because they can rely on restart_threads to signal
+       // an early abort. get_work on the other hand can't rely on
+       // restart_threads so need a much shorter scantime
+       if ( have_stratum )
+          max64 = 60 * thr_hashrates[thr_id];
+       else if ( have_longpoll )
+          max64 = LP_SCANTIME * thr_hashrates[thr_id];
+       else  // getwork inline
+          max64 = opt_scantime * thr_hashrates[thr_id];   
+
+/*       
        if ( have_stratum )
           max64 = LP_SCANTIME;
        else
           max64 = g_work_time + ( have_longpoll ? LP_SCANTIME : opt_scantime )
 	                      - time(NULL);
+*/
+
        // time limit
        if ( unlikely( opt_time_limit && firstwork_time ) )
        {
@@ -2068,17 +2161,20 @@ static void *miner_thread( void *userdata )
           }
           if ( remain < max64 ) max64 = remain;
        }
-       // Select nonce range for approx 1 min duration based
+
-       // on hashrate, initial value arbitrarilly set to 1000 just to get
+       // Select nonce range based on max64, the estimated number of hashes
+       // to meet the desired scan time.
+       // Initial value arbitrarilly set to 1000 just to get
        // a sample hashrate for the next time.
        uint32_t work_nonce = *nonceptr;
-       max64 = 60 * thr_hashrates[thr_id];
+//       max64 = 60 * thr_hashrates[thr_id];
        if ( max64 <= 0)
           max64 = 1000;
        if ( work_nonce + max64 > end_nonce )
           max_nonce = end_nonce;
        else
           max_nonce = work_nonce + (uint32_t)max64;
+
        // init time
        if ( firstwork_time == 0 )
           firstwork_time = time(NULL);
@@ -2209,6 +2305,8 @@ void restart_threads(void)
 {
 	for ( int i = 0; i < opt_n_threads; i++)
 		work_restart[i].restart = 1;
+   if ( opt_debug )
+      applog( LOG_INFO, "Threads restarted for new work."); 
 }
 
 json_t *std_longpoll_rpc_call( CURL *curl, int *err, char* lp_url )
@@ -2292,13 +2390,18 @@ start:
       soval = json_object_get(res, "submitold");
       submit_old = soval ? json_is_true(soval) : false;
 	   pthread_mutex_lock(&g_work_lock);
-	   start_job_id = g_work.job_id ? strdup(g_work.job_id) : NULL;
+
+// This code has been here for a long time even though job_id isn't used.
+// This needs to be changed eventually to test the block height properly
+// using g_work.block_height .     
+      start_job_id = g_work.job_id ? strdup(g_work.job_id) : NULL;
 	   if (have_gbt)
 	      rc = gbt_work_decode(res, &g_work);
 	   else
 	      rc = work_decode(res, &g_work);
 	   if (rc)
       {
+// purge job id from solo mining
         bool newblock = g_work.job_id && strcmp(start_job_id, g_work.job_id);
 	     newblock |= (start_diff != net_diff); // the best is the height but... longpoll...
         if (newblock)
@@ -3634,7 +3737,10 @@ int main(int argc, char *argv[])
 	/* ESET-NOD32 Detects these 2 thread_create... */
 	if (want_longpoll && !have_stratum)
    {
-		/* init longpoll thread info */
+      if ( opt_debug )
+         applog(LOG_INFO,"Creating long poll thread");
+
+      /* init longpoll thread info */
 		longpoll_thr_id = opt_n_threads + 1;
 		thr = &thr_info[longpoll_thr_id];
 		thr->id = longpoll_thr_id;
@@ -3650,7 +3756,10 @@ int main(int argc, char *argv[])
 	}
 	if (want_stratum)
    {
-		/* init stratum thread info */
+      if ( opt_debug )
+         applog(LOG_INFO,"Creating stratum thread");
+
+      /* init stratum thread info */
 		stratum_thr_id = opt_n_threads + 2;
 		thr = &thr_info[stratum_thr_id];
 		thr->id = stratum_thr_id;
@@ -3670,7 +3779,10 @@ int main(int argc, char *argv[])
 
 	if ( opt_api_enabled )
    {
-		/* api thread */
+      if ( opt_debug )
+         applog(LOG_INFO,"Creating API thread");
+
+      /* api thread */
 		api_thr_id = opt_n_threads + 3;
 		thr = &thr_info[api_thr_id];
 		thr->id = api_thr_id;
@@ -3680,7 +3792,7 @@ int main(int argc, char *argv[])
 		err = thread_create( thr, api_thread );
 		if ( err )
       {
-			applog( LOG_ERR, "api thread create failed" );
+			applog( LOG_ERR, "API thread create failed" );
 			return 1;
 		}
       if ( !opt_quiet )