From 3d701944b22fbd33fa7f0fb291a974c835542bfe Mon Sep 17 00:00:00 2001 From: Tanguy Pruvot Date: Fri, 5 Aug 2016 05:21:12 +0200 Subject: [PATCH] lbry: exit on card failure, dont loop --- lbry/cuda_sha256_lbry.cu | 3 ++- lbry/lbry.cu | 9 +++++++-- 2 files changed, 9 insertions(+), 3 deletions(-) diff --git a/lbry/cuda_sha256_lbry.cu b/lbry/cuda_sha256_lbry.cu index b6722c7..b7428d9 100644 --- a/lbry/cuda_sha256_lbry.cu +++ b/lbry/cuda_sha256_lbry.cu @@ -868,7 +868,7 @@ void lbry_sha256d_gpu_hash_final(const uint32_t threads, uint64_t *Hash512, uint } __host__ -void lbry_sha256d_hash_final(int thr_id, uint32_t threads, uint32_t *d_inputHash, uint32_t *d_resNonce) +int lbry_sha256d_hash_final(int thr_id, uint32_t threads, uint32_t *d_inputHash, uint32_t *d_resNonce) { int dev_id = device_map[thr_id]; const uint32_t threadsperblock = (device_sm[dev_id] > 500) ? 1024 : 768; @@ -877,4 +877,5 @@ void lbry_sha256d_hash_final(int thr_id, uint32_t threads, uint32_t *d_inputHash dim3 block(threadsperblock); lbry_sha256d_gpu_hash_final <<>> (threads, (uint64_t*) d_inputHash, d_resNonce); + return cudaGetLastError(); } diff --git a/lbry/lbry.cu b/lbry/lbry.cu index ac1e4ce..80eefdb 100644 --- a/lbry/lbry.cu +++ b/lbry/lbry.cu @@ -68,7 +68,7 @@ extern void lbry_sha256_setBlock_112(uint32_t *pdata, uint32_t *ptarget); extern void lbry_sha256d_hash_112(int thr_id, uint32_t threads, uint32_t startNonce, uint32_t *d_outputHash); extern void lbry_sha512_init(int thr_id); extern void lbry_sha512_hash_32(int thr_id, uint32_t threads, uint32_t *d_hash); -extern void lbry_sha256d_hash_final(int thr_id, uint32_t threads, uint32_t *d_inputHash, uint32_t *d_resNonce); +extern int lbry_sha256d_hash_final(int thr_id, uint32_t threads, uint32_t *d_inputHash, uint32_t *d_resNonce); static __inline uint32_t swab32_if(uint32_t val, bool iftrue) { return iftrue ? swab32(val) : val; @@ -136,7 +136,12 @@ extern "C" int scanhash_lbry(int thr_id, struct work *work, uint32_t max_nonce, lbry_sha512_hash_32(thr_id, throughput, d_hash[thr_id]); uint32_t resNonces[2] = { UINT32_MAX, UINT32_MAX }; - lbry_sha256d_hash_final(thr_id, throughput, d_hash[thr_id], d_resNonce[thr_id]); + int err = lbry_sha256d_hash_final(thr_id, throughput, d_hash[thr_id], d_resNonce[thr_id]); + if (err) { + // reinit + free_lbry(thr_id); + return -1; + } cudaMemcpy(resNonces, d_resNonce[thr_id], 2 * sizeof(uint32_t), cudaMemcpyDeviceToHost);