From 8dd57b715b093caaafcfa5579ba503ea0df0e9d2 Mon Sep 17 00:00:00 2001 From: jsteube Date: Wed, 7 Sep 2016 15:13:50 +0200 Subject: [PATCH] Create real opencl.c to better distinguish between opencl helper functions and library handling functions --- include/ext_OpenCL.h | 204 ---------------------------------------- include/opencl.h | 219 +++++++++++++++++++++++++++++++++++++++++++ include/shared.h | 4 - src/Makefile | 10 +- src/data.c | 1 + src/ext_OpenCL.c | 2 - src/hashcat.c | 1 + src/hwmon.c | 1 + src/interface.c | 1 + src/logfile.c | 1 + src/mpsp.c | 1 + src/opencl.c | 123 ++++++++++++++++++++++++ src/rp_cpu.c | 1 + src/shared.c | 108 +-------------------- src/tuningdb.c | 1 + 15 files changed, 356 insertions(+), 322 deletions(-) create mode 100644 include/opencl.h create mode 100644 src/opencl.c diff --git a/include/ext_OpenCL.h b/include/ext_OpenCL.h index d61133728..504516356 100644 --- a/include/ext_OpenCL.h +++ b/include/ext_OpenCL.h @@ -8,8 +8,6 @@ #ifndef _EXT_OPENCL_H #define _EXT_OPENCL_H -#include - #define CL_USE_DEPRECATED_OPENCL_1_2_APIS #define CL_USE_DEPRECATED_OPENCL_2_0_APIS @@ -31,208 +29,6 @@ #define CL_PLATFORMS_MAX 16 -static const char CL_VENDOR_AMD[] = "Advanced Micro Devices, Inc."; -static const char CL_VENDOR_AMD_USE_INTEL[] = "GenuineIntel"; -static const char CL_VENDOR_APPLE[] = "Apple"; -static const char CL_VENDOR_INTEL_BEIGNET[] = "Intel"; -static const char CL_VENDOR_INTEL_SDK[] = "Intel(R) Corporation"; -static const char CL_VENDOR_MESA[] = "Mesa"; -static const char CL_VENDOR_NV[] = "NVIDIA Corporation"; -static const char CL_VENDOR_POCL[] = "The pocl project"; - -typedef enum vendor_id -{ - VENDOR_ID_AMD = (1 << 0), - VENDOR_ID_APPLE = (1 << 1), - VENDOR_ID_INTEL_BEIGNET = (1 << 2), - VENDOR_ID_INTEL_SDK = (1 << 3), - VENDOR_ID_MESA = (1 << 4), - VENDOR_ID_NV = (1 << 5), - VENDOR_ID_POCL = (1 << 6), - VENDOR_ID_AMD_USE_INTEL = (1 << 7), - VENDOR_ID_GENERIC = (1 << 31) - -} vendor_id_t; - -typedef struct __hc_device_param hc_device_param_t; - -struct __hc_device_param -{ - cl_device_id device; - cl_device_type device_type; - - uint device_id; - uint platform_devices_id; // for mapping with hms devices - - bool skipped; - - uint sm_major; - uint sm_minor; - uint kernel_exec_timeout; - - uint device_processors; - u64 device_maxmem_alloc; - u64 device_global_mem; - u32 device_maxclock_frequency; - size_t device_maxworkgroup_size; - - uint vector_width; - - uint kernel_threads; - uint kernel_loops; - uint kernel_accel; - uint kernel_loops_min; - uint kernel_loops_max; - uint kernel_accel_min; - uint kernel_accel_max; - uint kernel_power; - uint hardware_power; - - size_t size_pws; - size_t size_tmps; - size_t size_hooks; - size_t size_bfs; - size_t size_combs; - size_t size_rules; - size_t size_rules_c; - size_t size_root_css; - size_t size_markov_css; - size_t size_digests; - size_t size_salts; - size_t size_shown; - size_t size_results; - size_t size_plains; - - FILE *combs_fp; - comb_t *combs_buf; - - void *hooks_buf; - - pw_t *pws_buf; - uint pws_cnt; - - u64 words_off; - u64 words_done; - - uint outerloop_pos; - uint outerloop_left; - - uint innerloop_pos; - uint innerloop_left; - - uint exec_pos; - double exec_ms[EXEC_CACHE]; - - // workaround cpu spinning - - double exec_us_prev1[EXPECTED_ITERATIONS]; - double exec_us_prev2[EXPECTED_ITERATIONS]; - double exec_us_prev3[EXPECTED_ITERATIONS]; - - // this is "current" speed - - uint speed_pos; - u64 speed_cnt[SPEED_CACHE]; - double speed_ms[SPEED_CACHE]; - - hc_timer_t timer_speed; - - // device specific attributes starting - - char *device_name; - char *device_vendor; - char *device_name_chksum; - char *device_version; - char *driver_version; - - bool opencl_v12; - - double nvidia_spin_damp; - - cl_platform_id platform; - - cl_uint device_vendor_id; - cl_uint platform_vendor_id; - - cl_kernel kernel1; - cl_kernel kernel12; - cl_kernel kernel2; - cl_kernel kernel23; - cl_kernel kernel3; - cl_kernel kernel_mp; - cl_kernel kernel_mp_l; - cl_kernel kernel_mp_r; - cl_kernel kernel_amp; - cl_kernel kernel_tm; - cl_kernel kernel_weak; - cl_kernel kernel_memset; - - cl_context context; - - cl_program program; - cl_program program_mp; - cl_program program_amp; - - cl_command_queue command_queue; - - cl_mem d_pws_buf; - cl_mem d_pws_amp_buf; - cl_mem d_words_buf_l; - cl_mem d_words_buf_r; - cl_mem d_rules; - cl_mem d_rules_c; - cl_mem d_combs; - cl_mem d_combs_c; - cl_mem d_bfs; - cl_mem d_bfs_c; - cl_mem d_tm_c; - cl_mem d_bitmap_s1_a; - cl_mem d_bitmap_s1_b; - cl_mem d_bitmap_s1_c; - cl_mem d_bitmap_s1_d; - cl_mem d_bitmap_s2_a; - cl_mem d_bitmap_s2_b; - cl_mem d_bitmap_s2_c; - cl_mem d_bitmap_s2_d; - cl_mem d_plain_bufs; - cl_mem d_digests_buf; - cl_mem d_digests_shown; - cl_mem d_salt_bufs; - cl_mem d_esalt_bufs; - cl_mem d_bcrypt_bufs; - cl_mem d_tmps; - cl_mem d_hooks; - cl_mem d_result; - cl_mem d_scryptV0_buf; - cl_mem d_scryptV1_buf; - cl_mem d_scryptV2_buf; - cl_mem d_scryptV3_buf; - cl_mem d_root_css_buf; - cl_mem d_markov_css_buf; - - void *kernel_params[PARAMCNT]; - void *kernel_params_mp[PARAMCNT]; - void *kernel_params_mp_r[PARAMCNT]; - void *kernel_params_mp_l[PARAMCNT]; - void *kernel_params_amp[PARAMCNT]; - void *kernel_params_tm[PARAMCNT]; - void *kernel_params_memset[PARAMCNT]; - - u32 kernel_params_buf32[PARAMCNT]; - - u32 kernel_params_mp_buf32[PARAMCNT]; - u64 kernel_params_mp_buf64[PARAMCNT]; - - u32 kernel_params_mp_r_buf32[PARAMCNT]; - u64 kernel_params_mp_r_buf64[PARAMCNT]; - - u32 kernel_params_mp_l_buf32[PARAMCNT]; - u64 kernel_params_mp_l_buf64[PARAMCNT]; - - u32 kernel_params_amp_buf32[PARAMCNT]; - u32 kernel_params_memset_buf32[PARAMCNT]; -}; - typedef cl_int (CL_API_CALL *OCL_CLBUILDPROGRAM) (cl_program, cl_uint, const cl_device_id *, const char *, void (CL_CALLBACK *)(cl_program, void *), void *); typedef cl_mem (CL_API_CALL *OCL_CLCREATEBUFFER) (cl_context, cl_mem_flags, size_t, void *, cl_int *); typedef cl_command_queue (CL_API_CALL *OCL_CLCREATECOMMANDQUEUE) (cl_context, cl_device_id, cl_command_queue_properties, cl_int *); diff --git a/include/opencl.h b/include/opencl.h new file mode 100644 index 000000000..804341338 --- /dev/null +++ b/include/opencl.h @@ -0,0 +1,219 @@ +/** + * Authors.....: Jens Steube + * Gabriele Gristina + * + * License.....: MIT + */ + +#ifndef _OPENCL_H +#define _OPENCL_H + +#include + +static const char CL_VENDOR_AMD[] = "Advanced Micro Devices, Inc."; +static const char CL_VENDOR_AMD_USE_INTEL[] = "GenuineIntel"; +static const char CL_VENDOR_APPLE[] = "Apple"; +static const char CL_VENDOR_INTEL_BEIGNET[] = "Intel"; +static const char CL_VENDOR_INTEL_SDK[] = "Intel(R) Corporation"; +static const char CL_VENDOR_MESA[] = "Mesa"; +static const char CL_VENDOR_NV[] = "NVIDIA Corporation"; +static const char CL_VENDOR_POCL[] = "The pocl project"; + +typedef enum vendor_id +{ + VENDOR_ID_AMD = (1 << 0), + VENDOR_ID_APPLE = (1 << 1), + VENDOR_ID_INTEL_BEIGNET = (1 << 2), + VENDOR_ID_INTEL_SDK = (1 << 3), + VENDOR_ID_MESA = (1 << 4), + VENDOR_ID_NV = (1 << 5), + VENDOR_ID_POCL = (1 << 6), + VENDOR_ID_AMD_USE_INTEL = (1 << 7), + VENDOR_ID_GENERIC = (1 << 31) + +} vendor_id_t; + +typedef struct __hc_device_param hc_device_param_t; + +struct __hc_device_param +{ + cl_device_id device; + cl_device_type device_type; + + uint device_id; + uint platform_devices_id; // for mapping with hms devices + + bool skipped; + + uint sm_major; + uint sm_minor; + uint kernel_exec_timeout; + + uint device_processors; + u64 device_maxmem_alloc; + u64 device_global_mem; + u32 device_maxclock_frequency; + size_t device_maxworkgroup_size; + + uint vector_width; + + uint kernel_threads; + uint kernel_loops; + uint kernel_accel; + uint kernel_loops_min; + uint kernel_loops_max; + uint kernel_accel_min; + uint kernel_accel_max; + uint kernel_power; + uint hardware_power; + + size_t size_pws; + size_t size_tmps; + size_t size_hooks; + size_t size_bfs; + size_t size_combs; + size_t size_rules; + size_t size_rules_c; + size_t size_root_css; + size_t size_markov_css; + size_t size_digests; + size_t size_salts; + size_t size_shown; + size_t size_results; + size_t size_plains; + + FILE *combs_fp; + comb_t *combs_buf; + + void *hooks_buf; + + pw_t *pws_buf; + uint pws_cnt; + + u64 words_off; + u64 words_done; + + uint outerloop_pos; + uint outerloop_left; + + uint innerloop_pos; + uint innerloop_left; + + uint exec_pos; + double exec_ms[EXEC_CACHE]; + + // workaround cpu spinning + + double exec_us_prev1[EXPECTED_ITERATIONS]; + double exec_us_prev2[EXPECTED_ITERATIONS]; + double exec_us_prev3[EXPECTED_ITERATIONS]; + + // this is "current" speed + + uint speed_pos; + u64 speed_cnt[SPEED_CACHE]; + double speed_ms[SPEED_CACHE]; + + hc_timer_t timer_speed; + + // device specific attributes starting + + char *device_name; + char *device_vendor; + char *device_name_chksum; + char *device_version; + char *driver_version; + + bool opencl_v12; + + double nvidia_spin_damp; + + cl_platform_id platform; + + cl_uint device_vendor_id; + cl_uint platform_vendor_id; + + cl_kernel kernel1; + cl_kernel kernel12; + cl_kernel kernel2; + cl_kernel kernel23; + cl_kernel kernel3; + cl_kernel kernel_mp; + cl_kernel kernel_mp_l; + cl_kernel kernel_mp_r; + cl_kernel kernel_amp; + cl_kernel kernel_tm; + cl_kernel kernel_weak; + cl_kernel kernel_memset; + + cl_context context; + + cl_program program; + cl_program program_mp; + cl_program program_amp; + + cl_command_queue command_queue; + + cl_mem d_pws_buf; + cl_mem d_pws_amp_buf; + cl_mem d_words_buf_l; + cl_mem d_words_buf_r; + cl_mem d_rules; + cl_mem d_rules_c; + cl_mem d_combs; + cl_mem d_combs_c; + cl_mem d_bfs; + cl_mem d_bfs_c; + cl_mem d_tm_c; + cl_mem d_bitmap_s1_a; + cl_mem d_bitmap_s1_b; + cl_mem d_bitmap_s1_c; + cl_mem d_bitmap_s1_d; + cl_mem d_bitmap_s2_a; + cl_mem d_bitmap_s2_b; + cl_mem d_bitmap_s2_c; + cl_mem d_bitmap_s2_d; + cl_mem d_plain_bufs; + cl_mem d_digests_buf; + cl_mem d_digests_shown; + cl_mem d_salt_bufs; + cl_mem d_esalt_bufs; + cl_mem d_bcrypt_bufs; + cl_mem d_tmps; + cl_mem d_hooks; + cl_mem d_result; + cl_mem d_scryptV0_buf; + cl_mem d_scryptV1_buf; + cl_mem d_scryptV2_buf; + cl_mem d_scryptV3_buf; + cl_mem d_root_css_buf; + cl_mem d_markov_css_buf; + + void *kernel_params[PARAMCNT]; + void *kernel_params_mp[PARAMCNT]; + void *kernel_params_mp_r[PARAMCNT]; + void *kernel_params_mp_l[PARAMCNT]; + void *kernel_params_amp[PARAMCNT]; + void *kernel_params_tm[PARAMCNT]; + void *kernel_params_memset[PARAMCNT]; + + u32 kernel_params_buf32[PARAMCNT]; + + u32 kernel_params_mp_buf32[PARAMCNT]; + u64 kernel_params_mp_buf64[PARAMCNT]; + + u32 kernel_params_mp_r_buf32[PARAMCNT]; + u64 kernel_params_mp_r_buf64[PARAMCNT]; + + u32 kernel_params_mp_l_buf32[PARAMCNT]; + u64 kernel_params_mp_l_buf64[PARAMCNT]; + + u32 kernel_params_amp_buf32[PARAMCNT]; + u32 kernel_params_memset_buf32[PARAMCNT]; +}; + +#endif // _OPENCL_H + +uint setup_opencl_platforms_filter (char *opencl_platforms); +u32 setup_devices_filter (char *opencl_devices); +cl_device_type setup_device_types_filter (char *opencl_device_types); diff --git a/include/shared.h b/include/shared.h index 3cc43b89a..3e3328aba 100644 --- a/include/shared.h +++ b/include/shared.h @@ -146,10 +146,6 @@ void handle_left_request (pot_t *pot, uint pot_cnt, char *input_buf, int input_l void handle_show_request_lm (pot_t *pot, uint pot_cnt, char *input_buf, int input_len, hash_t *hash_left, hash_t *hash_right, int (*sort_by_pot) (const void *, const void *), FILE *out_fp); void handle_left_request_lm (pot_t *pot, uint pot_cnt, char *input_buf, int input_len, hash_t *hash_left, hash_t *hash_right, int (*sort_by_pot) (const void *, const void *), FILE *out_fp); -u32 setup_opencl_platforms_filter (char *opencl_platforms); -u32 setup_devices_filter (char *opencl_devices); -cl_device_type setup_device_types_filter (char *opencl_device_types); - u32 get_random_num (const u32 min, const u32 max); u32 mydivc32 (const u32 dividend, const u32 divisor); diff --git a/src/Makefile b/src/Makefile index bd577a49f..a33c1b55b 100644 --- a/src/Makefile +++ b/src/Makefile @@ -165,7 +165,7 @@ LFLAGS_CROSS_WIN := -lpsapi ## Objects ## -NATIVE_OBJS := obj/affinity.NATIVE.o obj/filehandling.NATIVE.o obj/tuningdb.NATIVE.o obj/locking.NATIVE.o obj/folder.NATIVE.o obj/bitops.NATIVE.o obj/convert.NATIVE.o obj/cpu_aes.NATIVE.o obj/cpu_crc32.NATIVE.o obj/cpu_des.NATIVE.o obj/cpu_md5.NATIVE.o obj/cpu_sha1.NATIVE.o obj/cpu_sha256.NATIVE.o obj/data.NATIVE.o obj/ext_OpenCL.NATIVE.o obj/hwmon.NATIVE.o obj/interface.NATIVE.o obj/logfile.NATIVE.o obj/logging.NATIVE.o obj/memory.NATIVE.o obj/mpsp.NATIVE.o obj/rp_cpu.NATIVE.o obj/rp_kernel_on_cpu.NATIVE.o obj/shared.NATIVE.o obj/terminal.NATIVE.o obj/usage.NATIVE.o +NATIVE_OBJS := obj/opencl.NATIVE.o obj/affinity.NATIVE.o obj/filehandling.NATIVE.o obj/tuningdb.NATIVE.o obj/locking.NATIVE.o obj/folder.NATIVE.o obj/bitops.NATIVE.o obj/convert.NATIVE.o obj/cpu_aes.NATIVE.o obj/cpu_crc32.NATIVE.o obj/cpu_des.NATIVE.o obj/cpu_md5.NATIVE.o obj/cpu_sha1.NATIVE.o obj/cpu_sha256.NATIVE.o obj/data.NATIVE.o obj/ext_OpenCL.NATIVE.o obj/hwmon.NATIVE.o obj/interface.NATIVE.o obj/logfile.NATIVE.o obj/logging.NATIVE.o obj/memory.NATIVE.o obj/mpsp.NATIVE.o obj/rp_cpu.NATIVE.o obj/rp_kernel_on_cpu.NATIVE.o obj/shared.NATIVE.o obj/terminal.NATIVE.o obj/usage.NATIVE.o ifeq ($(UNAME),Linux) NATIVE_OBJS += obj/ext_ADL.NATIVE.o @@ -174,8 +174,8 @@ NATIVE_OBJS += obj/ext_nvml.NATIVE.o NATIVE_OBJS += obj/ext_xnvctrl.NATIVE.o endif -LINUX_32_OBJS := obj/affinity.LINUX.32.o obj/filehandling.LINUX.32.o obj/tuningdb.LINUX.32.o obj/locking.LINUX.32.o obj/folder.LINUX.32.o obj/bitops.LINUX.32.o obj/convert.LINUX.32.o obj/cpu_aes.LINUX.32.o obj/cpu_crc32.LINUX.32.o obj/cpu_des.LINUX.32.o obj/cpu_md5.LINUX.32.o obj/cpu_sha1.LINUX.32.o obj/cpu_sha256.LINUX.32.o obj/data.LINUX.32.o obj/ext_ADL.LINUX.32.o obj/ext_nvapi.LINUX.32.o obj/ext_nvml.LINUX.32.o obj/ext_OpenCL.LINUX.32.o obj/ext_xnvctrl.LINUX.32.o obj/hwmon.LINUX.32.o obj/interface.LINUX.32.o obj/logfile.LINUX.32.o obj/logging.LINUX.32.o obj/memory.LINUX.32.o obj/mpsp.LINUX.32.o obj/rp_cpu.LINUX.32.o obj/rp_kernel_on_cpu.LINUX.32.o obj/shared.LINUX.32.o obj/terminal.LINUX.32.o obj/usage.LINUX.32.o -LINUX_64_OBJS := obj/affinity.LINUX.64.o obj/filehandling.LINUX.64.o obj/tuningdb.LINUX.64.o obj/locking.LINUX.64.o obj/folder.LINUX.64.o obj/bitops.LINUX.64.o obj/convert.LINUX.64.o obj/cpu_aes.LINUX.64.o obj/cpu_crc32.LINUX.64.o obj/cpu_des.LINUX.64.o obj/cpu_md5.LINUX.64.o obj/cpu_sha1.LINUX.64.o obj/cpu_sha256.LINUX.64.o obj/data.LINUX.64.o obj/ext_ADL.LINUX.64.o obj/ext_nvapi.LINUX.64.o obj/ext_nvml.LINUX.64.o obj/ext_OpenCL.LINUX.64.o obj/ext_xnvctrl.LINUX.64.o obj/hwmon.LINUX.64.o obj/interface.LINUX.64.o obj/logfile.LINUX.64.o obj/logging.LINUX.64.o obj/memory.LINUX.64.o obj/mpsp.LINUX.64.o obj/rp_cpu.LINUX.64.o obj/rp_kernel_on_cpu.LINUX.64.o obj/shared.LINUX.64.o obj/terminal.LINUX.64.o obj/usage.LINUX.64.o +LINUX_32_OBJS := obj/opencl.LINUX.32.o obj/affinity.LINUX.32.o obj/filehandling.LINUX.32.o obj/tuningdb.LINUX.32.o obj/locking.LINUX.32.o obj/folder.LINUX.32.o obj/bitops.LINUX.32.o obj/convert.LINUX.32.o obj/cpu_aes.LINUX.32.o obj/cpu_crc32.LINUX.32.o obj/cpu_des.LINUX.32.o obj/cpu_md5.LINUX.32.o obj/cpu_sha1.LINUX.32.o obj/cpu_sha256.LINUX.32.o obj/data.LINUX.32.o obj/ext_ADL.LINUX.32.o obj/ext_nvapi.LINUX.32.o obj/ext_nvml.LINUX.32.o obj/ext_OpenCL.LINUX.32.o obj/ext_xnvctrl.LINUX.32.o obj/hwmon.LINUX.32.o obj/interface.LINUX.32.o obj/logfile.LINUX.32.o obj/logging.LINUX.32.o obj/memory.LINUX.32.o obj/mpsp.LINUX.32.o obj/rp_cpu.LINUX.32.o obj/rp_kernel_on_cpu.LINUX.32.o obj/shared.LINUX.32.o obj/terminal.LINUX.32.o obj/usage.LINUX.32.o +LINUX_64_OBJS := obj/opencl.LINUX.64.o obj/affinity.LINUX.64.o obj/filehandling.LINUX.64.o obj/tuningdb.LINUX.64.o obj/locking.LINUX.64.o obj/folder.LINUX.64.o obj/bitops.LINUX.64.o obj/convert.LINUX.64.o obj/cpu_aes.LINUX.64.o obj/cpu_crc32.LINUX.64.o obj/cpu_des.LINUX.64.o obj/cpu_md5.LINUX.64.o obj/cpu_sha1.LINUX.64.o obj/cpu_sha256.LINUX.64.o obj/data.LINUX.64.o obj/ext_ADL.LINUX.64.o obj/ext_nvapi.LINUX.64.o obj/ext_nvml.LINUX.64.o obj/ext_OpenCL.LINUX.64.o obj/ext_xnvctrl.LINUX.64.o obj/hwmon.LINUX.64.o obj/interface.LINUX.64.o obj/logfile.LINUX.64.o obj/logging.LINUX.64.o obj/memory.LINUX.64.o obj/mpsp.LINUX.64.o obj/rp_cpu.LINUX.64.o obj/rp_kernel_on_cpu.LINUX.64.o obj/shared.LINUX.64.o obj/terminal.LINUX.64.o obj/usage.LINUX.64.o # Windows CRT file globbing: @@ -185,8 +185,8 @@ CRT_GLOB_INCLUDE_FOLDER := $(dir $(lastword $(MAKEFILE_LIST))) include $(CRT_GLOB_INCLUDE_FOLDER)/win_file_globbing.mk -WIN_32_OBJS := obj/affinity.WIN.32.o obj/filehandling.WIN.32.o obj/tuningdb.WIN.32.o obj/locking.WIN.32.o obj/folder.WIN.32.o obj/bitops.WIN.32.o obj/convert.WIN.32.o obj/cpu_aes.WIN.32.o obj/cpu_crc32.WIN.32.o obj/cpu_des.WIN.32.o obj/cpu_md5.WIN.32.o obj/cpu_sha1.WIN.32.o obj/cpu_sha256.WIN.32.o obj/data.WIN.32.o obj/ext_ADL.WIN.32.o obj/ext_nvapi.WIN.32.o obj/ext_nvml.WIN.32.o obj/ext_OpenCL.WIN.32.o obj/ext_xnvctrl.WIN.32.o obj/hwmon.WIN.32.o obj/interface.WIN.32.o obj/logfile.WIN.32.o obj/logging.WIN.32.o obj/memory.WIN.32.o obj/mpsp.WIN.32.o obj/rp_cpu.WIN.32.o obj/rp_kernel_on_cpu.WIN.32.o obj/shared.WIN.32.o obj/terminal.WIN.32.o obj/usage.WIN.32.o $(CRT_GLOB_32) -WIN_64_OBJS := obj/affinity.WIN.64.o obj/filehandling.WIN.64.o obj/tuningdb.WIN.64.o obj/locking.WIN.64.o obj/folder.WIN.64.o obj/bitops.WIN.64.o obj/convert.WIN.64.o obj/cpu_aes.WIN.64.o obj/cpu_crc32.WIN.64.o obj/cpu_des.WIN.64.o obj/cpu_md5.WIN.64.o obj/cpu_sha1.WIN.64.o obj/cpu_sha256.WIN.64.o obj/data.WIN.64.o obj/ext_ADL.WIN.64.o obj/ext_nvapi.WIN.64.o obj/ext_nvml.WIN.64.o obj/ext_OpenCL.WIN.64.o obj/ext_xnvctrl.WIN.64.o obj/hwmon.WIN.64.o obj/interface.WIN.64.o obj/logfile.WIN.64.o obj/logging.WIN.64.o obj/memory.WIN.64.o obj/mpsp.WIN.64.o obj/rp_cpu.WIN.64.o obj/rp_kernel_on_cpu.WIN.64.o obj/shared.WIN.64.o obj/terminal.WIN.64.o obj/usage.WIN.64.o $(CRT_GLOB_64) +WIN_32_OBJS := obj/opencl.WIN.32.o obj/affinity.WIN.32.o obj/filehandling.WIN.32.o obj/tuningdb.WIN.32.o obj/locking.WIN.32.o obj/folder.WIN.32.o obj/bitops.WIN.32.o obj/convert.WIN.32.o obj/cpu_aes.WIN.32.o obj/cpu_crc32.WIN.32.o obj/cpu_des.WIN.32.o obj/cpu_md5.WIN.32.o obj/cpu_sha1.WIN.32.o obj/cpu_sha256.WIN.32.o obj/data.WIN.32.o obj/ext_ADL.WIN.32.o obj/ext_nvapi.WIN.32.o obj/ext_nvml.WIN.32.o obj/ext_OpenCL.WIN.32.o obj/ext_xnvctrl.WIN.32.o obj/hwmon.WIN.32.o obj/interface.WIN.32.o obj/logfile.WIN.32.o obj/logging.WIN.32.o obj/memory.WIN.32.o obj/mpsp.WIN.32.o obj/rp_cpu.WIN.32.o obj/rp_kernel_on_cpu.WIN.32.o obj/shared.WIN.32.o obj/terminal.WIN.32.o obj/usage.WIN.32.o $(CRT_GLOB_32) +WIN_64_OBJS := obj/opencl.WIN.64.o obj/affinity.WIN.64.o obj/filehandling.WIN.64.o obj/tuningdb.WIN.64.o obj/locking.WIN.64.o obj/folder.WIN.64.o obj/bitops.WIN.64.o obj/convert.WIN.64.o obj/cpu_aes.WIN.64.o obj/cpu_crc32.WIN.64.o obj/cpu_des.WIN.64.o obj/cpu_md5.WIN.64.o obj/cpu_sha1.WIN.64.o obj/cpu_sha256.WIN.64.o obj/data.WIN.64.o obj/ext_ADL.WIN.64.o obj/ext_nvapi.WIN.64.o obj/ext_nvml.WIN.64.o obj/ext_OpenCL.WIN.64.o obj/ext_xnvctrl.WIN.64.o obj/hwmon.WIN.64.o obj/interface.WIN.64.o obj/logfile.WIN.64.o obj/logging.WIN.64.o obj/memory.WIN.64.o obj/mpsp.WIN.64.o obj/rp_cpu.WIN.64.o obj/rp_kernel_on_cpu.WIN.64.o obj/shared.WIN.64.o obj/terminal.WIN.64.o obj/usage.WIN.64.o $(CRT_GLOB_64) ## ## Targets: Global diff --git a/src/data.c b/src/data.c index 968ca56c4..824b6d8ff 100644 --- a/src/data.c +++ b/src/data.c @@ -15,6 +15,7 @@ #include "hwmon.h" #include "mpsp.h" #include "rp_cpu.h" +#include "opencl.h" #include "data.h" hc_global_data_t data; diff --git a/src/ext_OpenCL.c b/src/ext_OpenCL.c index bf50aa928..fd1185416 100644 --- a/src/ext_OpenCL.c +++ b/src/ext_OpenCL.c @@ -7,10 +7,8 @@ #include "common.h" #include "types_int.h" -#include "types.h" #include "memory.h" #include "logging.h" -#include "timer.h" #include "dynloader.h" #include "ext_OpenCL.h" diff --git a/src/hashcat.c b/src/hashcat.c index b80b1dec7..d4bd582d1 100644 --- a/src/hashcat.c +++ b/src/hashcat.c @@ -45,6 +45,7 @@ #include "cpu_sha1.h" #include "cpu_sha256.h" #include "filehandling.h" +#include "opencl.h" #include "tuningdb.h" #include "thread.h" #include "locking.h" diff --git a/src/hwmon.c b/src/hwmon.c index 4fe4587bf..c1da9c808 100644 --- a/src/hwmon.c +++ b/src/hwmon.c @@ -19,6 +19,7 @@ #include "hwmon.h" #include "mpsp.h" #include "rp_cpu.h" +#include "opencl.h" #include "data.h" #ifdef HAVE_HWMON diff --git a/src/interface.c b/src/interface.c index 1c76ca549..17e7f9d90 100644 --- a/src/interface.c +++ b/src/interface.c @@ -33,6 +33,7 @@ #include "hwmon.h" #include "mpsp.h" #include "rp_cpu.h" +#include "opencl.h" #include "data.h" #include "interface.h" diff --git a/src/logfile.c b/src/logfile.c index bd881ce53..79abc0658 100644 --- a/src/logfile.c +++ b/src/logfile.c @@ -23,6 +23,7 @@ #include "hwmon.h" #include "rp_cpu.h" #include "mpsp.h" +#include "opencl.h" #include "data.h" #include "logfile.h" diff --git a/src/mpsp.c b/src/mpsp.c index fa19aff5c..60de9e2a1 100644 --- a/src/mpsp.c +++ b/src/mpsp.c @@ -26,6 +26,7 @@ #include "hwmon.h" #include "mpsp.h" #include "rp_cpu.h" +#include "opencl.h" #include "data.h" #include "shared.h" diff --git a/src/opencl.c b/src/opencl.c new file mode 100644 index 000000000..595437a2e --- /dev/null +++ b/src/opencl.c @@ -0,0 +1,123 @@ +/** + * Authors.....: Jens Steube + * Gabriele Gristina + * + * License.....: MIT + */ + +#include "common.h" +#include "types_int.h" +#include "types.h" +#include "memory.h" +#include "logging.h" +#include "ext_OpenCL.h" +#include "timer.h" +#include "opencl.h" + +uint setup_opencl_platforms_filter (char *opencl_platforms) +{ + uint opencl_platforms_filter = 0; + + if (opencl_platforms) + { + char *platforms = mystrdup (opencl_platforms); + + char *next = strtok (platforms, ","); + + do + { + int platform = atoi (next); + + if (platform < 1 || platform > 32) + { + log_error ("ERROR: Invalid OpenCL platform %u specified", platform); + + exit (-1); + } + + opencl_platforms_filter |= 1u << (platform - 1); + + } while ((next = strtok (NULL, ",")) != NULL); + + myfree (platforms); + } + else + { + opencl_platforms_filter = -1u; + } + + return opencl_platforms_filter; +} + +u32 setup_devices_filter (char *opencl_devices) +{ + u32 devices_filter = 0; + + if (opencl_devices) + { + char *devices = mystrdup (opencl_devices); + + char *next = strtok (devices, ","); + + do + { + int device_id = atoi (next); + + if (device_id < 1 || device_id > 32) + { + log_error ("ERROR: Invalid device_id %u specified", device_id); + + exit (-1); + } + + devices_filter |= 1u << (device_id - 1); + + } while ((next = strtok (NULL, ",")) != NULL); + + myfree (devices); + } + else + { + devices_filter = -1u; + } + + return devices_filter; +} + +cl_device_type setup_device_types_filter (char *opencl_device_types) +{ + cl_device_type device_types_filter = 0; + + if (opencl_device_types) + { + char *device_types = mystrdup (opencl_device_types); + + char *next = strtok (device_types, ","); + + do + { + int device_type = atoi (next); + + if (device_type < 1 || device_type > 3) + { + log_error ("ERROR: Invalid device_type %u specified", device_type); + + exit (-1); + } + + device_types_filter |= 1u << device_type; + + } while ((next = strtok (NULL, ",")) != NULL); + + myfree (device_types); + } + else + { + // Do not use CPU by default, this often reduces GPU performance because + // the CPU is too busy to handle GPU synchronization + + device_types_filter = CL_DEVICE_TYPE_ALL & ~CL_DEVICE_TYPE_CPU; + } + + return device_types_filter; +} diff --git a/src/rp_cpu.c b/src/rp_cpu.c index 6fb09671b..1fa64ab60 100644 --- a/src/rp_cpu.c +++ b/src/rp_cpu.c @@ -25,6 +25,7 @@ #include "hwmon.h" #include "mpsp.h" #include "rp_cpu.h" +#include "opencl.h" #include "data.h" #include "shared.h" #include "rp_cpu.h" diff --git a/src/shared.c b/src/shared.c index 86bce3fea..7258fd547 100644 --- a/src/shared.c +++ b/src/shared.c @@ -29,6 +29,7 @@ #include "hwmon.h" #include "mpsp.h" #include "rp_cpu.h" +#include "opencl.h" #include "data.h" #include "shared.h" @@ -903,113 +904,6 @@ void handle_left_request_lm (pot_t *pot, uint pot_cnt, char *input_buf, int inpu if (weak_hash_found == 1) myfree (pot_right_ptr); } -uint setup_opencl_platforms_filter (char *opencl_platforms) -{ - uint opencl_platforms_filter = 0; - - if (opencl_platforms) - { - char *platforms = mystrdup (opencl_platforms); - - char *next = strtok (platforms, ","); - - do - { - int platform = atoi (next); - - if (platform < 1 || platform > 32) - { - log_error ("ERROR: Invalid OpenCL platform %u specified", platform); - - exit (-1); - } - - opencl_platforms_filter |= 1u << (platform - 1); - - } while ((next = strtok (NULL, ",")) != NULL); - - myfree (platforms); - } - else - { - opencl_platforms_filter = -1u; - } - - return opencl_platforms_filter; -} - -u32 setup_devices_filter (char *opencl_devices) -{ - u32 devices_filter = 0; - - if (opencl_devices) - { - char *devices = mystrdup (opencl_devices); - - char *next = strtok (devices, ","); - - do - { - int device_id = atoi (next); - - if (device_id < 1 || device_id > 32) - { - log_error ("ERROR: Invalid device_id %u specified", device_id); - - exit (-1); - } - - devices_filter |= 1u << (device_id - 1); - - } while ((next = strtok (NULL, ",")) != NULL); - - myfree (devices); - } - else - { - devices_filter = -1u; - } - - return devices_filter; -} - -cl_device_type setup_device_types_filter (char *opencl_device_types) -{ - cl_device_type device_types_filter = 0; - - if (opencl_device_types) - { - char *device_types = mystrdup (opencl_device_types); - - char *next = strtok (device_types, ","); - - do - { - int device_type = atoi (next); - - if (device_type < 1 || device_type > 3) - { - log_error ("ERROR: Invalid device_type %u specified", device_type); - - exit (-1); - } - - device_types_filter |= 1u << device_type; - - } while ((next = strtok (NULL, ",")) != NULL); - - myfree (device_types); - } - else - { - // Do not use CPU by default, this often reduces GPU performance because - // the CPU is too busy to handle GPU synchronization - - device_types_filter = CL_DEVICE_TYPE_ALL & ~CL_DEVICE_TYPE_CPU; - } - - return device_types_filter; -} u32 get_random_num (const u32 min, const u32 max) { diff --git a/src/tuningdb.c b/src/tuningdb.c index a3ac051a0..072aefa0b 100644 --- a/src/tuningdb.c +++ b/src/tuningdb.c @@ -11,6 +11,7 @@ #include "memory.h" #include "filehandling.h" #include "ext_OpenCL.h" +#include "opencl.h" #include "tuningdb.h" static int sort_by_tuning_db_alias (const void *v1, const void *v2)