From cd08fa5f8c25570481aa5550b7c192aca3921028 Mon Sep 17 00:00:00 2001 From: jsteube Date: Wed, 4 May 2016 10:32:54 +0200 Subject: [PATCH] Limit kernel_threads on CPU --- src/oclHashcat.c | 17 ++++++++++++----- 1 file changed, 12 insertions(+), 5 deletions(-) diff --git a/src/oclHashcat.c b/src/oclHashcat.c index 97b09a9b1..6e8bc749a 100644 --- a/src/oclHashcat.c +++ b/src/oclHashcat.c @@ -84,6 +84,8 @@ double TARGET_MS_PROFILE[3] = { 8, 16, 96 }; #define KERNEL_RULES 1024 #define KERNEL_COMBS 1024 #define KERNEL_BFS 1024 +#define KERNEL_THREADS_MAX 256 +#define KERNEL_THREADS_MAX_CPU 16 #define POWERTUNE_ENABLE 0 #define LOGFILE_DISABLE 0 #define SCRYPT_TMTO 0 @@ -3024,9 +3026,8 @@ static void autotune (hc_device_param_t *device_param) } // because of the balance we may have some free space left! - // at this point, allow a small variance to overdrive the limit - const int exec_left = (target_ms * 1.2) / exec_best; + const int exec_left = target_ms / exec_best; const int accel_left = kernel_accel_max / kernel_accel_best; @@ -13583,15 +13584,21 @@ int main (int argc, char **argv) /** * kernel threads: some algorithms need a fixed kernel-threads count * because of shared memory usage or bitslice + * there needs to be some upper limit, otherwise there's too much overhead */ - uint kernel_threads = device_param->device_maxworkgroup_size; + uint kernel_threads = MIN (KERNEL_THREADS_MAX, device_param->device_maxworkgroup_size); + + if (device_param->device_type & CL_DEVICE_TYPE_CPU) + { + kernel_threads = KERNEL_THREADS_MAX_CPU; + } if (hash_mode == 1500) kernel_threads = 64; // DES if (hash_mode == 3000) kernel_threads = 64; // DES - if (hash_mode == 3200) kernel_threads = 8; // blowfish + if (hash_mode == 3200) kernel_threads = 8; // Blowfish if (hash_mode == 7500) kernel_threads = 64; // RC4 - if (hash_mode == 9000) kernel_threads = 8; // blowfish + if (hash_mode == 9000) kernel_threads = 8; // Blowfish if (hash_mode == 9700) kernel_threads = 64; // RC4 if (hash_mode == 9710) kernel_threads = 64; // RC4 if (hash_mode == 9800) kernel_threads = 64; // RC4