Move tuningdb specific functions into their own source file
This commit is contained in:
jsteube
@@ -29,6 +29,9 @@
|
||||
|
||||
#define DEVICES_MAX 128
|
||||
|
||||
#define PARAMCNT 64
|
||||
|
||||
|
||||
#define CEIL(a) ((a - (int) (a)) > 0 ? a + 1 : a)
|
||||
|
||||
#ifdef __APPLE__
|
||||
|
||||
+195
-14
@@ -8,6 +8,27 @@
|
||||
#ifndef _EXT_OPENCL_H
|
||||
#define _EXT_OPENCL_H
|
||||
|
||||
#include <stdio.h>
|
||||
|
||||
#define CL_USE_DEPRECATED_OPENCL_1_2_APIS
|
||||
#define CL_USE_DEPRECATED_OPENCL_2_0_APIS
|
||||
|
||||
#ifdef __APPLE__
|
||||
#include <OpenCL/cl.h>
|
||||
#endif
|
||||
|
||||
#ifdef WIN
|
||||
#include <CL/cl.h>
|
||||
#endif
|
||||
|
||||
#ifdef __linux__
|
||||
#include <CL/cl.h>
|
||||
#endif
|
||||
|
||||
#ifdef __FreeBSD__
|
||||
#include <CL/cl.h>
|
||||
#endif
|
||||
|
||||
#define CL_PLATFORMS_MAX 16
|
||||
|
||||
static const char CL_VENDOR_AMD[] = "Advanced Micro Devices, Inc.";
|
||||
@@ -33,24 +54,184 @@ typedef enum vendor_id
|
||||
|
||||
} vendor_id_t;
|
||||
|
||||
#define CL_USE_DEPRECATED_OPENCL_1_2_APIS
|
||||
#define CL_USE_DEPRECATED_OPENCL_2_0_APIS
|
||||
typedef struct __hc_device_param hc_device_param_t;
|
||||
|
||||
#ifdef __APPLE__
|
||||
#include <OpenCL/cl.h>
|
||||
#endif
|
||||
struct __hc_device_param
|
||||
{
|
||||
cl_device_id device;
|
||||
cl_device_type device_type;
|
||||
|
||||
#ifdef WIN
|
||||
#include <CL/cl.h>
|
||||
#endif
|
||||
uint device_id;
|
||||
uint platform_devices_id; // for mapping with hms devices
|
||||
|
||||
#ifdef __linux__
|
||||
#include <CL/cl.h>
|
||||
#endif
|
||||
bool skipped;
|
||||
|
||||
#ifdef __FreeBSD__
|
||||
#include <CL/cl.h>
|
||||
#endif
|
||||
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 *);
|
||||
|
||||
@@ -0,0 +1,19 @@
|
||||
/**
|
||||
* Author......: Jens Steube <jens.steube@gmail.com>
|
||||
* License.....: MIT
|
||||
*/
|
||||
|
||||
#ifndef _FILEHANDLING_H
|
||||
#define _FILEHANDLING_H
|
||||
|
||||
#include <stdio.h>
|
||||
#include <string.h>
|
||||
#include <errno.h>
|
||||
|
||||
uint count_lines (FILE *fd);
|
||||
|
||||
int fgetl (FILE *fp, char *line_buf);
|
||||
|
||||
int in_superchop (char *buf);
|
||||
|
||||
#endif // _FILEHANDLING_H
|
||||
+1
-8
@@ -126,8 +126,6 @@ static const char ST_0010[] = "Autotuning";
|
||||
* functions
|
||||
*/
|
||||
|
||||
uint count_lines (FILE *fd);
|
||||
|
||||
void *rulefind (const void *key, void *base, int nmemb, size_t size, int (*compar) (const void *, const void *));
|
||||
|
||||
int sort_by_u32 (const void *p1, const void *p2);
|
||||
@@ -179,8 +177,7 @@ u64 mydivc64 (const u64 dividend, const u64 divisor);
|
||||
|
||||
void format_speed_display (double val, char *buf, size_t len);
|
||||
void format_timer_display (struct tm *tm, char *buf, size_t len);
|
||||
int fgetl (FILE *fp, char *line_buf);
|
||||
int in_superchop (char *buf);
|
||||
|
||||
char **scan_directory (const char *path);
|
||||
int count_dictionaries (char **dictionary_files);
|
||||
|
||||
@@ -198,10 +195,6 @@ void myquit (void);
|
||||
|
||||
void set_cpu_affinity (char *cpu_affinity);
|
||||
|
||||
void tuning_db_destroy (tuning_db_t *tuning_db);
|
||||
tuning_db_t *tuning_db_alloc (FILE *fp);
|
||||
tuning_db_t *tuning_db_init (const char *tuning_db_file);
|
||||
tuning_db_entry_t *tuning_db_search (tuning_db_t *tuning_db, hc_device_param_t *device_param, int attack_mode, int hash_type);
|
||||
|
||||
void naive_replace (char *s, const u8 key_char, const u8 replace_char);
|
||||
void naive_escape (char *s, size_t s_max, const u8 key_char, const u8 escape_char);
|
||||
|
||||
@@ -0,0 +1,46 @@
|
||||
/**
|
||||
* Author......: Jens Steube <jens.steube@gmail.com>
|
||||
* License.....: MIT
|
||||
*/
|
||||
|
||||
#ifndef _TUNINGDB_H
|
||||
#define _TUNINGDB_H
|
||||
|
||||
#include <stdio.h>
|
||||
#include <errno.h>
|
||||
|
||||
typedef struct
|
||||
{
|
||||
char *device_name;
|
||||
char *alias_name;
|
||||
|
||||
} tuning_db_alias_t;
|
||||
|
||||
typedef struct
|
||||
{
|
||||
char *device_name;
|
||||
int attack_mode;
|
||||
int hash_type;
|
||||
int workload_profile;
|
||||
int vector_width;
|
||||
int kernel_accel;
|
||||
int kernel_loops;
|
||||
|
||||
} tuning_db_entry_t;
|
||||
|
||||
typedef struct
|
||||
{
|
||||
tuning_db_alias_t *alias_buf;
|
||||
int alias_cnt;
|
||||
|
||||
tuning_db_entry_t *entry_buf;
|
||||
int entry_cnt;
|
||||
|
||||
} tuning_db_t;
|
||||
|
||||
void tuning_db_destroy (tuning_db_t *tuning_db);
|
||||
tuning_db_t *tuning_db_alloc (FILE *fp);
|
||||
tuning_db_t *tuning_db_init (const char *tuning_db_file);
|
||||
tuning_db_entry_t *tuning_db_search (tuning_db_t *tuning_db, hc_device_param_t *device_param, int attack_mode, int hash_type);
|
||||
|
||||
#endif // _TUNINGDB_H
|
||||
-209
@@ -330,220 +330,11 @@ typedef struct
|
||||
|
||||
} wordr_t;
|
||||
|
||||
typedef struct
|
||||
{
|
||||
char *device_name;
|
||||
char *alias_name;
|
||||
|
||||
} tuning_db_alias_t;
|
||||
|
||||
typedef struct
|
||||
{
|
||||
char *device_name;
|
||||
int attack_mode;
|
||||
int hash_type;
|
||||
int workload_profile;
|
||||
int vector_width;
|
||||
int kernel_accel;
|
||||
int kernel_loops;
|
||||
|
||||
} tuning_db_entry_t;
|
||||
|
||||
typedef struct
|
||||
{
|
||||
tuning_db_alias_t *alias_buf;
|
||||
int alias_cnt;
|
||||
|
||||
tuning_db_entry_t *entry_buf;
|
||||
int entry_cnt;
|
||||
|
||||
} tuning_db_t;
|
||||
|
||||
#define RULES_MAX 256
|
||||
#define PW_MIN 0
|
||||
#define PW_MAX 54
|
||||
#define PW_MAX1 (PW_MAX + 1)
|
||||
#define PW_DICTMAX 31
|
||||
#define PW_DICTMAX1 (PW_DICTMAX + 1)
|
||||
#define PARAMCNT 64
|
||||
|
||||
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 struct __hc_device_param hc_device_param_t;
|
||||
|
||||
#endif // _TYPES_H
|
||||
|
||||
Reference in New Issue
Block a user