diff options
| author | psychocrypt <psychocryptHPC@gmail.com> | 2018-04-08 20:40:54 +0200 |
|---|---|---|
| committer | psychocrypt <psychocryptHPC@gmail.com> | 2018-04-08 20:52:30 +0200 |
| commit | 1b18f598aa1190a0e6126ed2c70e052e9403d180 (patch) | |
| tree | 626314a2add2f76649ca31c4b7e0a731f6d80f9b /xmrstak/backend/nvidia/nvcc_code | |
| parent | a57976439b1322eecb5e29198f3d9676d6f06909 (diff) | |
| download | xmr-stak-1b18f598aa1190a0e6126ed2c70e052e9403d180.zip xmr-stak-1b18f598aa1190a0e6126ed2c70e052e9403d180.tar.gz | |
refactor scratchpad creation
Use the maximum scratchpad size from before and after the fork.
Diffstat (limited to 'xmrstak/backend/nvidia/nvcc_code')
| -rw-r--r-- | xmrstak/backend/nvidia/nvcc_code/cuda_extra.cu | 10 |
1 files changed, 8 insertions, 2 deletions
diff --git a/xmrstak/backend/nvidia/nvcc_code/cuda_extra.cu b/xmrstak/backend/nvidia/nvcc_code/cuda_extra.cu index ead93c5..f016ef4 100644 --- a/xmrstak/backend/nvidia/nvcc_code/cuda_extra.cu +++ b/xmrstak/backend/nvidia/nvcc_code/cuda_extra.cu @@ -279,7 +279,10 @@ extern "C" int cryptonight_extra_cpu_init(nvid_ctx* ctx) if(gpuArch < 70) CUDA_CHECK(ctx->device_id, cudaDeviceSetCacheConfig(cudaFuncCachePreferL1)); - size_t hashMemSize = cn_select_memory(::jconf::inst()->GetMiningAlgo()); + size_t hashMemSize = std::max( + cn_select_memory(::jconf::inst()->GetMiningAlgo()), + cn_select_memory(::jconf::inst()->GetMiningAlgoRoot()) + ); size_t wsize = ctx->device_blocks * ctx->device_threads; CUDA_CHECK(ctx->device_id, cudaMalloc(&ctx->d_ctx_state, 50 * sizeof(uint32_t) * wsize)); @@ -576,7 +579,10 @@ extern "C" int cuda_get_deviceinfo(nvid_ctx* ctx) ctx->total_device_memory = totalMemory; ctx->free_device_memory = freeMemory; - size_t hashMemSize = cn_select_memory(::jconf::inst()->GetMiningAlgo()); + size_t hashMemSize = std::max( + cn_select_memory(::jconf::inst()->GetMiningAlgo()), + cn_select_memory(::jconf::inst()->GetMiningAlgoRoot()) + ); #ifdef WIN32 /* We use in windows bfactor (split slow kernel into smaller parts) to avoid |
