Follow @Openwall on Twitter for new release announcements and other news
[<prev] [next>] [<thread-prev] [thread-next>] [day] [month] [year] [list]
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[0](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.