[<prev] [next>] [thread-next>] [day] [month] [year] [list]
Date: Tue, 30 Oct 2012 22:09:47 +0100
From: magnum <john.magnum@...hmail.com>
To: john-dev@...ts.openwall.com
Subject: mscash2-opencl problems with GT650M
Sayantan,
On some weaker cards (GT650M for one) I get this from mscash2-opencl:
OpenCL error (UNKNOWN ERROR :() in file (common_opencl_pbkdf2.c) at line
(313) - (SYNC FAILED)
That's not really a good error message. After some looks at the code I
came up with this patch (not applied to git yet):
diff --git a/src/common_opencl_pbkdf2.c b/src/common_opencl_pbkdf2.c
index aa43382..07c749c 100644
--- a/src/common_opencl_pbkdf2.c
+++ b/src/common_opencl_pbkdf2.c
@@ -291,7 +291,8 @@ static gpu_mem_buffer exec_pbkdf2(cl_uint
*pass_api,cl_uint *salt_api,cl_uint sa
HANDLE_CLERROR(clSetKernelArg(krnl[platform_no][dev_no],3,sizeof(cl_uint),&num),"Set
Kernel Arg FAILED arg3");
-
err=clEnqueueNDRangeKernel(cmdq[platform_no][dev_no],krnl[platform_no][dev_no],1,NULL,&N,&M,0,NULL,&evnt);
+
clEnqueueNDRangeKernel(cmdq[platform_no][dev_no],krnl[platform_no][dev_no],1,NULL,&N,&M,0,NULL,&evnt);
+ err=clFinish(cmdq[platform_no][dev_no]);
if(err){
if(PROFILE){
This properly sets 'err' with any failure from the kernel and lets the
subsequent if clause halve the LWS. However, the problem moves. I now
get this:
Optimal Work Group Size:8
Kernel Execution Speed (Higher is better):-1.000000
Benchmarking: M$ Cache Hash 2 (DCC2) PBKDF2-HMAC-SHA-1 [OpenCL]...
OpenCL error (CL_INVALID_COMMAND_QUEUE) in file (common_opencl_pbkdf2.c)
at line (286) - (Copy data to gpu)
I could look further into this but it may be quicker for you. I believe
my patch above should be applied but some more seem to be needed to
adopt for available resources. Do you have any immediate ideas or should
I go on trying to fix this?
magnum
Powered by blists - more mailing lists
Confused about mailing lists and their use?
Read about mailing lists on Wikipedia
and check out these
guidelines on proper formatting of your messages.
