From 65d47dbb313d7126c588e1e86adac6c7e8803cf2 Mon Sep 17 00:00:00 2001 From: psychocrypt Date: Thu, 5 Oct 2017 22:22:30 +0200 Subject: add assert to validate number of global threads check that the number of global threads is a multiple of the workgroup size --- xmrstak/backend/amd/amd_gpu/gpu.cpp | 5 +++++ 1 file changed, 5 insertions(+) (limited to 'xmrstak/backend/amd') diff --git a/xmrstak/backend/amd/amd_gpu/gpu.cpp b/xmrstak/backend/amd/amd_gpu/gpu.cpp index 2bd3688..d43961e 100644 --- a/xmrstak/backend/amd/amd_gpu/gpu.cpp +++ b/xmrstak/backend/amd/amd_gpu/gpu.cpp @@ -20,6 +20,7 @@ #include #include #include +#include #ifdef _WIN32 #include @@ -807,6 +808,8 @@ size_t XMRRunJob(GpuContext* ctx, cl_uint* HashOutput) size_t w_size = ctx->workSize; // round up to next multiple of w_size size_t g_thd = ((g_intensity + w_size - 1u) / w_size) * w_size; + // number of global threads must be a multiple of the work group size (w_size) + assert(g_thd%w_size == 0); for(int i = 2; i < 6; ++i) { @@ -892,6 +895,8 @@ size_t XMRRunJob(GpuContext* ctx, cl_uint* HashOutput) // round up to next multiple of w_size BranchNonces[i] = ((BranchNonces[i] + w_size - 1u) / w_size) * w_size; + // number of global threads must be a multiple of the work group size (w_size) + assert(BranchNonces%w_size == 0); if((ret = clEnqueueNDRangeKernel(ctx->CommandQueues, ctx->Kernels[i + 3], 1, &ctx->Nonce, BranchNonces + i, &w_size, 0, NULL, NULL)) != CL_SUCCESS) { printer::inst()->print_msg(L1,"Error %s when calling clEnqueueNDRangeKernel for kernel %d.", err_to_str(ret), i + 3); -- cgit v1.1