Date: Thu, 8 Aug 2013 18:53:08 +0530 From: Sayantan Datta <std2048@...il.com> To: john-dev@...ts.openwall.com Subject: Re: Busted OpenCL formats Hi magnum, On Thu, Aug 8, 2013 at 4:39 PM, magnum <john.magnum@...hmail.com> wrote: > Sayantan, > > What's with the bleeding changes to nt-opencl and raw-md4/5-opencl that > you started in July? All three seem to be b0rken right now. > > raw-md4-opencl now fails on OSX CPU but it passes Test Suite using GPU. > > $ ../run/john test -form:raw-md4-opencl -dev=0 -inc:digits -max-len=9 > -prog=10 > Device 0: Intel(R) Core(TM) i7-3615QM CPU @ 2.30GHz > Local worksize (LWS) 1, global worksize (GWS) 16777216 > Loaded 1 password hash (Raw-MD4-opencl [MD4 OpenCL (inefficient, > development use only)]) > New local worksize (LWS) 64 > Press 'q' or Ctrl-C to abort, almost any other key for status > OpenCL error (CL_INVALID_WORK_GROUP_SIZE) in file (opencl_rawmd4_fmt.c) at > line (495) - (failed in clEnqueueNDRangeKernel) > > Can you please wait a while. I'm trying to optimize these formats as best as I can without compromising any performance for gpus, which makes them harder to run on cpu devices with 1 or 2 work item per work group. I'll probably need to write new kernels for cpu or try some #ifdefs to fix them. Also why are we trying to run them on cpu devices when we have well optimized code specifically meant for cpus? Should I add some code that cleanly ejects when a cpu device is detected for now? Later on I'll add a generic kernels for the above formats that will run on any device provided. > > raw-md5-opencl doesn't report candidates, it just says "CHECK". And it > fails Test Suite. > > $ ../run/john test -form:raw-md5-opencl -dev=0 -inc:digits -max-len=9 > -prog=10 > Device 0: Intel(R) Core(TM) i7-3615QM CPU @ 2.30GHz > Local worksize (LWS) 1, global worksize (GWS) 4194304 > Loaded 1 password hash (Raw-MD5-opencl [MD5 OpenCL (inefficient, > development use only)]) > Press 'q' or Ctrl-C to abort, almost any other key for status > 0g 0:00:00:12 26.80% (ETA: 12:55:21) 0g/s 22978Kp/s 22978Kc/s 22978KC/s > ..CHECK > 0g 0:00:00:22 47.94% (ETA: 12:55:22) 0g/s 23190Kp/s 23190Kc/s 23190KC/s > ..CHECK > 0g 0:00:00:32 69.08% (ETA: 12:55:23) 0g/s 23287Kp/s 23287Kc/s 23287KC/s > ..CHECK > 0g 0:00:00:42 90.21% (ETA: 12:55:23) 0g/s 23361Kp/s 23361Kc/s 23361KC/s > ..CHECK > 0g 0:00:00:47 DONE (2013-08-08 12:55) 0g/s 23426Kp/s 23426Kc/s 23426KC/s > ..CHECK > Session completed > > I'm working on these issues please wait. It is difficult to recognize which call to get_key is meant for status checks and which are meant for procuring the cracked password. Although I have fixed this issue to some extent on my private repo but they are still not perfect. To get accurate status checks I'll probably need another function meant for status checks alone. > > NT-opencl segfaults (CPU) or fails (GPU). For some strange reason it > passes Test Suite using GPU. I can't see why it passes self-test under TS > but not from command line. > > > $ ../run/john test -form:nt-opencl -dev=1 -inc:digits -max-len=9 -prog=10 > Device 1: GeForce GT 650M > Local worksize (LWS) 64, Global worksize (GWS) 524288 > Loaded 1 password hash (nt-opencl, NT [MD4 OpenCL (inefficient, > development use only)]) > Self test failed (get_hash(6)) > > I'll try to reproduce the above error on 7970 fix it ASAP. Regards, Sayantan Content of type "text/html" skipped
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.