arno/hashcat
1
0
Fork 0
mirror of https://github.com/hashcat/hashcat.git synced 2024-11-23 00:28:11 +00:00
Code Issues Releases Wiki Activity

Merge pull request #212 from gm4tr1x/buildProgramMod

Browse Source 
Skip device if build process fail
This commit is contained in:
Jens Steube 2016-02-08 08:20:34 +01:00
commit bea43f9e73
4 changed files with 69 additions and 21 deletions
Show Stats Download Patch File Download Diff File Expand all files Collapse all files

2
include/ext_OpenCL.h
View File

@ -109,7 +109,7 @@ cl_context hc_clCreateContext (OCL_PTR *ocl, cl_context_properties *properties,
cl_kernel hc_clCreateKernel (OCL_PTR *ocl, cl_program program, const char *kernel_name);
cl_program hc_clCreateProgramWithSource (OCL_PTR *ocl, cl_context context, cl_uint count, const char **strings, const size_t *lengths);
cl_program hc_clCreateProgramWithBinary (OCL_PTR *ocl, cl_context context, cl_uint num_devices, const cl_device_id *device_list, const size_t *lengths, const unsigned char **binaries, cl_int *binary_status);
void hc_clBuildProgram (OCL_PTR *ocl, cl_program program, cl_uint num_devices, const cl_device_id *device_list, const char *options, void (CL_CALLBACK *pfn_notify) (cl_program program, void *user_data), void *user_data);
cl_int hc_clBuildProgram (OCL_PTR *ocl, cl_program program, cl_uint num_devices, const cl_device_id *device_list, const char *options, void (CL_CALLBACK *pfn_notify) (cl_program program, void *user_data), void *user_data, bool exitOnFail);
cl_int hc_clEnqueueNDRangeKernel (OCL_PTR *ocl, cl_command_queue command_queue, cl_kernel kernel, cl_uint work_dim, const size_t *global_work_offset, const size_t *global_work_size, const size_t *local_work_size, cl_uint num_events_in_wait_list, const cl_event *event_wait_list, cl_event *event, bool exitOnFail);
void hc_clEnqueueReadBuffer (OCL_PTR *ocl, cl_command_queue command_queue, cl_mem buffer, cl_bool blocking_read, size_t offset, size_t cb, void *ptr, cl_uint num_events_in_wait_list, const cl_event *event_wait_list, cl_event *event);
void hc_clEnqueueWriteBuffer (OCL_PTR *ocl, cl_command_queue command_queue, cl_mem buffer, cl_bool blocking_write, size_t offset, size_t cb, const void *ptr, cl_uint num_events_in_wait_list, const cl_event *event_wait_list, cl_event *event);

2
include/types.h
View File

@ -834,7 +834,7 @@ struct __hc_device_param
uint device_id;
uint platform_devices_id; // for mapping with hms devices
uint skipped;
bool skipped;
uint sm_major;
uint sm_minor;

45
src/ext_OpenCL.c
View File

@ -376,33 +376,54 @@ cl_program hc_clCreateProgramWithBinary (OCL_PTR *ocl, cl_context context, cl_ui
return (program);
}
void hc_clBuildProgram (OCL_PTR *ocl, cl_program program, cl_uint num_devices, const cl_device_id *device_list, const char *options, void (CL_CALLBACK *pfn_notify) (cl_program program, void *user_data), void *user_data)
cl_int hc_clBuildProgram (OCL_PTR *ocl, cl_program program, cl_uint num_devices, const cl_device_id *device_list, const char *options, void (CL_CALLBACK *pfn_notify) (cl_program program, void *user_data), void *user_data, bool exitOnFail)
{
cl_int CL_err = ocl->clBuildProgram (program, num_devices, device_list, options, pfn_notify, user_data);
if (CL_err != CL_SUCCESS)
{
log_error ("ERROR: %s : %d : %s\n", "clBuildProgram()", CL_err, val2cstr_cl (CL_err));
size_t len = strlen (options) + 1 + 15;
char *buf = NULL;
size_t len = 0;
char *options_update = (char *) mymalloc (len + 1);
cl_int err = hc_clGetProgramBuildInfo (ocl, program, *device_list, CL_PROGRAM_BUILD_LOG, 0, NULL, &len);
options_update = strncat (options_update, options, len - 1 - 15);
options_update = strncat (options_update, " -cl-opt-disable", 1 + 15);
if (err == CL_SUCCESS && len > 0)
if (data.quiet == 0) log_error ("\n=== Build failed, retry with optimization disabled ===\n");
CL_err = ocl->clBuildProgram (program, num_devices, device_list, options_update, pfn_notify, user_data);
myfree (options_update);
if (CL_err != CL_SUCCESS)
{
buf = (char *) mymalloc (len + 1);
log_error ("ERROR: %s : %d : %s\n", "clBuildProgram()", CL_err, val2cstr_cl (CL_err));
if (hc_clGetProgramBuildInfo (ocl, program, *device_list, CL_PROGRAM_BUILD_LOG, len, buf, NULL) == CL_SUCCESS)
log_error ("\n=== Build Options : %s ===\n", options);
size_t len = 0;
cl_int err = hc_clGetProgramBuildInfo (ocl, program, *device_list, CL_PROGRAM_BUILD_LOG, 0, NULL, &len);
if (err == CL_SUCCESS && len > 0)
{
fprintf (stderr, "\n=== Build Log (start) ===\n%s\n=== Build Log (end) ===\n", buf);
char *buf = (char *) mymalloc (len + 1);
if (hc_clGetProgramBuildInfo (ocl, program, *device_list, CL_PROGRAM_BUILD_LOG, len, buf, NULL) == CL_SUCCESS)
{
fprintf (stderr, "\n=== Build Log (start) ===\n%s\n=== Build Log (end) ===\n", buf);
}
myfree (buf);
}
myfree (buf);
}
if (exitOnFail) exit (-1);
exit (-1);
return (-1);
}
}
return 0;
}
cl_kernel hc_clCreateKernel (OCL_PTR *ocl, cl_program program, const char *kernel_name)

41
src/oclHashcat.c
View File

@ -13775,7 +13775,14 @@ int main (int argc, char **argv)
device_param->program = hc_clCreateProgramWithSource (data.ocl, device_param->context, 1, (const char **) kernel_sources, NULL);
hc_clBuildProgram (data.ocl, device_param->program, 1, &device_param->device, build_opts, NULL, NULL);
int rc = hc_clBuildProgram (data.ocl, device_param->program, 1, &device_param->device, build_opts, NULL, NULL, false);
if (rc != 0)
{
device_param->skipped = true;
log_info ("Device #%u: Kernel %s build failure. Proceed without this device.", device_id + 1, source_file);
continue;
}
size_t binary_size;
@ -13797,7 +13804,7 @@ int main (int argc, char **argv)
device_param->program = hc_clCreateProgramWithBinary (data.ocl, device_param->context, 1, &device_param->device, kernel_lengths, (const u8 **) kernel_sources, NULL);
hc_clBuildProgram (data.ocl, device_param->program, 1, &device_param->device, build_opts, NULL, NULL);
hc_clBuildProgram (data.ocl, device_param->program, 1, &device_param->device, build_opts, NULL, NULL, true);
}
}
else
@ -13819,7 +13826,13 @@ int main (int argc, char **argv)
snprintf (build_opts_update, sizeof (build_opts_update) - 1, "%s -DSCRYPT_N=%d -DSCRYPT_R=%d -DSCRYPT_P=%d -DSCRYPT_TMTO=%d", build_opts, data.salts_buf[0].scrypt_N, data.salts_buf[0].scrypt_r, data.salts_buf[0].scrypt_p, 1 << data.salts_buf[0].scrypt_tmto);
}
hc_clBuildProgram (data.ocl, device_param->program, 1, &device_param->device, build_opts_update, NULL, NULL);
int rc = hc_clBuildProgram (data.ocl, device_param->program, 1, &device_param->device, build_opts_update, NULL, NULL, false);
if (rc != 0)
{
device_param->skipped = true;
log_info ("Device #%u: Kernel %s build failure. Proceed without this device.", device_id + 1, source_file);
}
}
local_free (kernel_lengths);
@ -13883,7 +13896,14 @@ int main (int argc, char **argv)
device_param->program_mp = hc_clCreateProgramWithSource (data.ocl, device_param->context, 1, (const char **) kernel_sources, NULL);
hc_clBuildProgram (data.ocl, device_param->program_mp, 1, &device_param->device, build_opts, NULL, NULL);
int rc = hc_clBuildProgram (data.ocl, device_param->program_mp, 1, &device_param->device, build_opts, NULL, NULL, false);
if (rc != 0)
{
device_param->skipped = true;
log_info ("Device #%u: Kernel %s build failure. Proceed without this device.", device_id + 1, source_file);
continue;
}
size_t binary_size;
@ -13905,7 +13925,7 @@ int main (int argc, char **argv)
device_param->program_mp = hc_clCreateProgramWithBinary (data.ocl, device_param->context, 1, &device_param->device, kernel_lengths, (const u8 **) kernel_sources, NULL);
hc_clBuildProgram (data.ocl, device_param->program_mp, 1, &device_param->device, build_opts, NULL, NULL);
hc_clBuildProgram (data.ocl, device_param->program_mp, 1, &device_param->device, build_opts, NULL, NULL, true);
}
local_free (kernel_lengths);
@ -13973,7 +13993,14 @@ int main (int argc, char **argv)
device_param->program_amp = hc_clCreateProgramWithSource (data.ocl, device_param->context, 1, (const char **) kernel_sources, NULL);
hc_clBuildProgram (data.ocl, device_param->program_amp, 1, &device_param->device, build_opts, NULL, NULL);
int rc = hc_clBuildProgram (data.ocl, device_param->program_amp, 1, &device_param->device, build_opts, NULL, NULL, false);
if (rc != 0)
{
device_param->skipped = true;
log_info ("Device #%u: Kernel %s build failure. Proceed without this device.", device_id + 1, source_file);
continue;
}
size_t binary_size;
@ -13995,7 +14022,7 @@ int main (int argc, char **argv)
device_param->program_amp = hc_clCreateProgramWithBinary (data.ocl, device_param->context, 1, &device_param->device, kernel_lengths, (const u8 **) kernel_sources, NULL);
hc_clBuildProgram (data.ocl, device_param->program_amp, 1, &device_param->device, build_opts, NULL, NULL);
hc_clBuildProgram (data.ocl, device_param->program_amp, 1, &device_param->device, build_opts, NULL, NULL, true);
}
local_free (kernel_lengths);