Merge branch 'dev' of github.com:xmrig/xmrig-cuda into dev

diantinoto · Nov 13, 2020 · 0c43fd2 · 0c43fd2
2 parents 3064c25 + 6ddb273
commit 0c43fd2
Show file tree

Hide file tree

Showing 2 changed files with 6 additions and 0 deletions.
diff --git a/src/AstroBWT/dero/AstroBWT.cu b/src/AstroBWT/dero/AstroBWT.cu
@@ -165,6 +165,8 @@ void hash(nvid_ctx *ctx, uint32_t nonce, uint64_t target, uint32_t *rescount, ui
     CUDA_CHECK_KERNEL(ctx->device_id, sha3<<<global_work_size, 32>>>((uint8_t*)ctx->astrobwt_tmp_indices, (uint32_t*)ctx->astrobwt_bwt_data_sizes, STAGE1_DATA_STRIDE * 8, (uint64_t*)ctx->astrobwt_salsa20_keys));
     CUDA_CHECK_KERNEL(ctx->device_id, filter<<<global_work_size / 32, 32>>>(nonce, BWT_DATA_MAX_SIZE, (uint32_t*)ctx->astrobwt_salsa20_keys, (uint32_t*)ctx->astrobwt_filtered_hashes));
 
+    CUDA_CHECK(ctx->device_id, cudaDeviceSynchronize());
+
     uint32_t num_filtered_hashes;
     CUDA_CHECK(ctx->device_id, cudaMemcpy(&num_filtered_hashes, ctx->astrobwt_filtered_hashes, sizeof(num_filtered_hashes), cudaMemcpyDeviceToHost));
 
@@ -183,6 +185,8 @@ void hash(nvid_ctx *ctx, uint32_t nonce, uint64_t target, uint32_t *rescount, ui
         CUDA_CHECK_KERNEL(ctx->device_id, find_shares<<<global_work_size / 32, 32>>>((uint64_t*)ctx->astrobwt_salsa20_keys, (uint32_t*)ctx->astrobwt_filtered_hashes, target, (uint32_t*)ctx->astrobwt_shares));
     }
 
+    CUDA_CHECK(ctx->device_id, cudaDeviceSynchronize());
+
     uint32_t shares[11];
     CUDA_CHECK(ctx->device_id, cudaMemcpy(shares, ctx->astrobwt_shares, sizeof(shares), cudaMemcpyDeviceToHost));
 

diff --git a/src/cuda_extra.cu b/src/cuda_extra.cu
@@ -437,6 +437,8 @@ void cryptonight_extra_cpu_final(nvid_ctx *ctx, uint32_t startNonce, uint64_t ta
         CUDA_CHECK_KERNEL(ctx->device_id, cryptonight_extra_gpu_final<Algorithm::CN_0> << <grid, block >> > (wsize, target, ctx->d_result_count, ctx->d_result_nonce, ctx->d_ctx_state, ctx->d_ctx_key2));
     }
 
+    CUDA_CHECK(ctx->device_id, cudaDeviceSynchronize());
+
     CUDA_CHECK(ctx->device_id, cudaMemcpy(rescount, ctx->d_result_count, sizeof(uint32_t), cudaMemcpyDeviceToHost));
     CUDA_CHECK(ctx->device_id, cudaMemcpy(resnonce, ctx->d_result_nonce, 16 * sizeof(uint32_t), cudaMemcpyDeviceToHost));