mirror of
https://github.com/hashcat/hashcat.git
synced 2024-12-22 22:58:30 +00:00
Fixed CL_INVALID_WORK_GROUP_SIZE error with Apple CPU
This commit is contained in:
parent
94b5cef2ab
commit
f9598f691d
@ -110,7 +110,7 @@ cl_kernel hc_clCreateKernel (OCL_PTR *ocl, cl_program program, const char *kerne
|
|||||||
cl_program hc_clCreateProgramWithSource (OCL_PTR *ocl, cl_context context, cl_uint count, const char **strings, const size_t *lengths);
|
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);
|
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);
|
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);
|
||||||
void 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);
|
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_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);
|
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);
|
||||||
void hc_clEnqueueCopyBuffer (OCL_PTR *ocl, cl_command_queue command_queue, cl_mem src_buffer, cl_mem dst_buffer, size_t src_offset, size_t dst_offset, size_t cb, cl_uint num_events_in_wait_list, const cl_event *event_wait_list, cl_event *event);
|
void hc_clEnqueueCopyBuffer (OCL_PTR *ocl, cl_command_queue command_queue, cl_mem src_buffer, cl_mem dst_buffer, size_t src_offset, size_t dst_offset, size_t cb, cl_uint num_events_in_wait_list, const cl_event *event_wait_list, cl_event *event);
|
||||||
|
@ -125,16 +125,23 @@ void ocl_close (OCL_PTR *ocl)
|
|||||||
}
|
}
|
||||||
}
|
}
|
||||||
|
|
||||||
void 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)
|
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)
|
||||||
{
|
{
|
||||||
cl_int CL_err = ocl->clEnqueueNDRangeKernel (command_queue, kernel, work_dim, global_work_offset, global_work_size, local_work_size, num_events_in_wait_list, event_wait_list, event);
|
cl_int CL_err = ocl->clEnqueueNDRangeKernel (command_queue, kernel, work_dim, global_work_offset, global_work_size, local_work_size, num_events_in_wait_list, event_wait_list, event);
|
||||||
|
|
||||||
if (CL_err != CL_SUCCESS)
|
if (CL_err != CL_SUCCESS)
|
||||||
{
|
{
|
||||||
log_error ("ERROR: %s : %d : %s\n", "clEnqueueNDRangeKernel()", CL_err, val2cstr_cl (CL_err));
|
if (exitOnFail)
|
||||||
|
{
|
||||||
|
log_error ("ERROR: %s : %d : %s\n", "clEnqueueNDRangeKernel()", CL_err, val2cstr_cl (CL_err));
|
||||||
|
|
||||||
exit (-1);
|
exit (-1);
|
||||||
|
}
|
||||||
|
|
||||||
|
return (-1);
|
||||||
}
|
}
|
||||||
|
|
||||||
|
return 0;
|
||||||
}
|
}
|
||||||
|
|
||||||
void hc_clGetEventInfo (OCL_PTR *ocl, cl_event event, cl_event_info param_name, size_t param_value_size, void *param_value, size_t *param_value_size_ret)
|
void hc_clGetEventInfo (OCL_PTR *ocl, cl_event event, cl_event_info param_name, size_t param_value_size, void *param_value, size_t *param_value_size_ret)
|
||||||
|
@ -2426,14 +2426,19 @@ static void run_kernel (const uint kern_run, hc_device_param_t *device_param, co
|
|||||||
const size_t global_work_size[3] = { num_elements, 32, 1 };
|
const size_t global_work_size[3] = { num_elements, 32, 1 };
|
||||||
const size_t local_work_size[3] = { kernel_threads / 32, 32, 1 };
|
const size_t local_work_size[3] = { kernel_threads / 32, 32, 1 };
|
||||||
|
|
||||||
hc_clEnqueueNDRangeKernel (data.ocl, device_param->command_queue, kernel, 2, NULL, global_work_size, local_work_size, 0, NULL, NULL);
|
hc_clEnqueueNDRangeKernel (data.ocl, device_param->command_queue, kernel, 2, NULL, global_work_size, local_work_size, 0, NULL, NULL, true);
|
||||||
}
|
}
|
||||||
else
|
else
|
||||||
{
|
{
|
||||||
const size_t global_work_size[3] = { num_elements, 1, 1 };
|
const size_t global_work_size[3] = { num_elements, 1, 1 };
|
||||||
const size_t local_work_size[3] = { kernel_threads, 1, 1 };
|
const size_t local_work_size[3] = { kernel_threads, 1, 1 };
|
||||||
|
|
||||||
hc_clEnqueueNDRangeKernel (data.ocl, device_param->command_queue, kernel, 1, NULL, global_work_size, local_work_size, 0, NULL, NULL);
|
if (hc_clEnqueueNDRangeKernel (data.ocl, device_param->command_queue, kernel, 1, NULL, global_work_size, local_work_size, 0, NULL, NULL, false) != CL_SUCCESS)
|
||||||
|
{
|
||||||
|
const size_t local_work_size_fallback[3] = { 1, 1, 1 };
|
||||||
|
|
||||||
|
hc_clEnqueueNDRangeKernel (data.ocl, device_param->command_queue, kernel, 1, NULL, global_work_size, local_work_size_fallback, 0, NULL, NULL, true);
|
||||||
|
}
|
||||||
}
|
}
|
||||||
|
|
||||||
hc_clFlush (data.ocl, device_param->command_queue);
|
hc_clFlush (data.ocl, device_param->command_queue);
|
||||||
@ -2495,9 +2500,14 @@ static void run_kernel_mp (const uint kern_run, hc_device_param_t *device_param,
|
|||||||
}
|
}
|
||||||
|
|
||||||
const size_t global_work_size[3] = { num_elements, 1, 1 };
|
const size_t global_work_size[3] = { num_elements, 1, 1 };
|
||||||
const size_t local_work_size[3] = { kernel_threads, 1, 1 };
|
const size_t local_work_size[3] = { kernel_threads, 1, 1 };
|
||||||
|
|
||||||
hc_clEnqueueNDRangeKernel (data.ocl, device_param->command_queue, kernel, 1, NULL, global_work_size, local_work_size, 0, NULL, NULL);
|
if (hc_clEnqueueNDRangeKernel (data.ocl, device_param->command_queue, kernel, 1, NULL, global_work_size, local_work_size, 0, NULL, NULL, false) != CL_SUCCESS)
|
||||||
|
{
|
||||||
|
const size_t local_work_size_fallback[3] = { 1, 1, 1 };
|
||||||
|
|
||||||
|
hc_clEnqueueNDRangeKernel (data.ocl, device_param->command_queue, kernel, 1, NULL, global_work_size, local_work_size_fallback, 0, NULL, NULL, true);
|
||||||
|
}
|
||||||
|
|
||||||
hc_clFlush (data.ocl, device_param->command_queue);
|
hc_clFlush (data.ocl, device_param->command_queue);
|
||||||
|
|
||||||
@ -2517,7 +2527,7 @@ static void run_kernel_tb (hc_device_param_t *device_param, const uint num)
|
|||||||
const size_t global_work_size[3] = { num_elements, 1, 1 };
|
const size_t global_work_size[3] = { num_elements, 1, 1 };
|
||||||
const size_t local_work_size[3] = { kernel_threads, 1, 1 };
|
const size_t local_work_size[3] = { kernel_threads, 1, 1 };
|
||||||
|
|
||||||
hc_clEnqueueNDRangeKernel (data.ocl, device_param->command_queue, kernel, 1, NULL, global_work_size, local_work_size, 0, NULL, NULL);
|
hc_clEnqueueNDRangeKernel (data.ocl, device_param->command_queue, kernel, 1, NULL, global_work_size, local_work_size, 0, NULL, NULL, true);
|
||||||
|
|
||||||
hc_clFlush (data.ocl, device_param->command_queue);
|
hc_clFlush (data.ocl, device_param->command_queue);
|
||||||
|
|
||||||
@ -2535,7 +2545,7 @@ static void run_kernel_tm (hc_device_param_t *device_param)
|
|||||||
const size_t global_work_size[3] = { num_elements, 1, 1 };
|
const size_t global_work_size[3] = { num_elements, 1, 1 };
|
||||||
const size_t local_work_size[3] = { kernel_threads, 1, 1 };
|
const size_t local_work_size[3] = { kernel_threads, 1, 1 };
|
||||||
|
|
||||||
hc_clEnqueueNDRangeKernel (data.ocl, device_param->command_queue, kernel, 1, NULL, global_work_size, local_work_size, 0, NULL, NULL);
|
hc_clEnqueueNDRangeKernel (data.ocl, device_param->command_queue, kernel, 1, NULL, global_work_size, local_work_size, 0, NULL, NULL, true);
|
||||||
|
|
||||||
hc_clFlush (data.ocl, device_param->command_queue);
|
hc_clFlush (data.ocl, device_param->command_queue);
|
||||||
|
|
||||||
@ -2564,7 +2574,7 @@ static void run_kernel_amp (hc_device_param_t *device_param, const uint num)
|
|||||||
const size_t global_work_size[3] = { num_elements, 1, 1 };
|
const size_t global_work_size[3] = { num_elements, 1, 1 };
|
||||||
const size_t local_work_size[3] = { kernel_threads, 1, 1 };
|
const size_t local_work_size[3] = { kernel_threads, 1, 1 };
|
||||||
|
|
||||||
hc_clEnqueueNDRangeKernel (data.ocl, device_param->command_queue, kernel, 1, NULL, global_work_size, local_work_size, 0, NULL, NULL);
|
hc_clEnqueueNDRangeKernel (data.ocl, device_param->command_queue, kernel, 1, NULL, global_work_size, local_work_size, 0, NULL, NULL, true);
|
||||||
|
|
||||||
hc_clFlush (data.ocl, device_param->command_queue);
|
hc_clFlush (data.ocl, device_param->command_queue);
|
||||||
|
|
||||||
|
Loading…
Reference in New Issue
Block a user