diff options
| author | David Phillips <david@sighup.nz> | 2017-11-19 23:12:13 +1300 |
|---|---|---|
| committer | David Phillips <david@sighup.nz> | 2018-04-24 17:11:30 +1200 |
| commit | 27e9f33ed36c10b1a4e34c0ebc86d87ab2cfb5ff (patch) | |
| tree | d1d5ea55e9875d83c41b1fbb572a420850102fc3 /trampoline.c | |
| parent | 036c470e94c278aeddcd9528f857eb66809ac7d6 (diff) | |
| download | sand-leek-27e9f33ed36c10b1a4e34c0ebc86d87ab2cfb5ff.tar.xz | |
Initial dump of SHA CL port
Diffstat (limited to 'trampoline.c')
| -rw-r--r-- | trampoline.c | 18 |
1 files changed, 15 insertions, 3 deletions
diff --git a/trampoline.c b/trampoline.c index bfbc508..bebb160 100644 --- a/trampoline.c +++ b/trampoline.c @@ -193,9 +193,15 @@ void tramp_destroy() { clReleaseKernel(kernel); clReleaseProgram(program); + clFlush(command_queue); + clFinish(command_queue); clReleaseCommandQueue(command_queue); clReleaseContext(context); + clReleaseMemObject(device_result); + clReleaseMemObject(device_sha); + clReleaseMemObject(device_search); + if (devices) { free(devices); devices = NULL; @@ -322,11 +328,11 @@ int tramp_compile_kernel() * * FIXME investigate using something more flexible? */ -int tramp_set_kernel_args(unsigned int raw_len) +int tramp_set_kernel_args(unsigned int raw_len, unsigned int bitmask) { cl_int ret = 0; - device_result = clCreateBuffer(context, CL_MEM_WRITE_ONLY, 65536, NULL, &ret); + device_result = clCreateBuffer(context, CL_MEM_WRITE_ONLY, 65536*4, NULL, &ret); if (ret != CL_SUCCESS) { fprintf(stderr, "Failed to create buffer for slave device: %s ", get_cl_error_string(ret)); return 1; @@ -368,6 +374,12 @@ int tramp_set_kernel_args(unsigned int raw_len) return 1; } + ret = clSetKernelArg(kernel, 4, sizeof(cl_int), &bitmask); + if (ret != CL_SUCCESS) { + fprintf(stderr, "Error on bitmask argument: %s ", get_cl_error_string(ret)); + return 1; + } + return 0; } @@ -383,7 +395,7 @@ int tramp_run_kernel() cl_event event; cl_int ret = 0; size_t workgroup_sizes[2]; - workgroup_sizes[0] = 65536; + workgroup_sizes[0] = 63356; workgroup_sizes[1] = 1; ret = clEnqueueNDRangeKernel(command_queue, kernel, 2, NULL, workgroup_sizes, NULL, 0, NULL, &event); |