diff --git a/run/opencl/opencl_DES_hst_dev_shared.h b/run/opencl/opencl_DES_hst_dev_shared.h index 346b3e7544..7501c2722d 100644 --- a/run/opencl/opencl_DES_hst_dev_shared.h +++ b/run/opencl/opencl_DES_hst_dev_shared.h @@ -18,7 +18,6 @@ #define OVERRIDE_AUTO_CONFIG 0 #define HARDCODE_SALT 0 #define FULL_UNROLL 0 -#define PARALLEL_BUILD 0 #define WORD int #define DES_bs_vector WORD diff --git a/src/opencl_DES_bs_f_plug.c b/src/opencl_DES_bs_f_plug.c index 1db996186a..3e8b6ee2d3 100644 --- a/src/opencl_DES_bs_f_plug.c +++ b/src/opencl_DES_bs_f_plug.c @@ -10,10 +10,6 @@ #include #include -#if _OPENMP -#include -#endif - #include "options.h" #include "opencl_DES_bs.h" #include "../run/opencl/opencl_DES_hst_dev_shared.h" @@ -701,18 +697,11 @@ static void reset(struct db_main *db) if (num_salts > 10 && !ocl_any_test_running && john_main_process) fprintf(stderr, "Building %d per-salt kernels, one dot per three salts done: ", num_salts); -#if _OPENMP && PARALLEL_BUILD -#pragma omp parallel for -#endif for (i = 0; i < num_salts; i++) { init_kernel(salt_list[i], gpu_id, 1, 0, forced_global_keys ? 0 :local_work_size); -#if _OPENMP && PARALLEL_BUILD - if (omp_get_thread_num() == 0) -#endif - { - opencl_process_event(); - } + opencl_process_event(); + if (num_salts > 10 && (i % 3) == 2 && !ocl_any_test_running && john_main_process) fprintf(stderr, "."); } diff --git a/src/opencl_DES_bs_h_plug.c b/src/opencl_DES_bs_h_plug.c index 33caae396a..a24dc8a19c 100644 --- a/src/opencl_DES_bs_h_plug.c +++ b/src/opencl_DES_bs_h_plug.c @@ -10,10 +10,6 @@ #include #include -#if _OPENMP -#include -#endif - #include "options.h" #include "opencl_DES_bs.h" #include "../run/opencl/opencl_DES_hst_dev_shared.h" @@ -253,12 +249,7 @@ static void init_kernel(WORD salt_val, int id_gpu, int save_binary, int force_bu kernels[id_gpu][salt_val] = clCreateKernel(program, "DES_bs_25", &err_code); HANDLE_CLERROR(err_code, "Create Kernel DES_bs_25 failed.\n"); -#if _OPENMP -#pragma omp critical -#endif -{ HANDLE_CLERROR(clSetKernelArg(kernels[id_gpu][salt_val], 0, sizeof(cl_mem), &buffer_map), "Failed setting kernel argument buffer_map, kernel DES_bs_25.\n"); -} marked_salts[salt_val] = salt_val; HANDLE_CLERROR(clReleaseProgram(program), "Error releasing Program"); @@ -715,18 +706,11 @@ static void reset(struct db_main *db) if (num_salts > 10 && !ocl_any_test_running && john_main_process) fprintf(stderr, "Building %d per-salt kernels, one dot per three salts done: ", num_salts); -#if _OPENMP && PARALLEL_BUILD -#pragma omp parallel for -#endif for (i = 0; i < num_salts; i++) { init_kernel(salt_list[i], gpu_id, 1, 0, forced_global_keys ? 0 :local_work_size); -#if _OPENMP && PARALLEL_BUILD - if (omp_get_thread_num() == 0) -#endif - { - opencl_process_event(); - } + opencl_process_event(); + if (num_salts > 10 && (i % 3) == 2 && !ocl_any_test_running && john_main_process) fprintf(stderr, "."); } diff --git a/src/opencl_common.c b/src/opencl_common.c index af2c0b08ff..7ac8168f22 100644 --- a/src/opencl_common.c +++ b/src/opencl_common.c @@ -1268,10 +1268,10 @@ void opencl_build(int sequential_id, const char *opts, int save, const char *fil * pathnames differently than the OpenCL backend would for the includes). * * Saving and restoring of the current directory here is incompatible with - * concurrent kernel builds by multiple threads, like we'd do with the - * PARALLEL_BUILD setting in descrypt-opencl (currently disabled and considered - * unsupported). We'd probably need to save and restore the directory - * before/after all kernel builds, not before/after each. + * concurrent kernel builds by multiple threads, like we did for a while with + * the PARALLEL_BUILD setting in descrypt-opencl (which was first disabled and + * later removed altogether). We'd probably need to save and restore the + * directory before/after all kernel builds, not before/after each. * * We primarily use open()/fchdir(), falling back to getcwd()/chdir() when * open() or/and fchdir() fails.