Follow @Openwall on Twitter for new release announcements and other news
[<prev] [next>] [<thread-prev] [day] [month] [year] [list]
Message-ID: <0c2958f0e82b0a8b3e2efcfc1ec81b93@smtp.hushmail.com>
Date: Wed, 31 Oct 2012 18:54:43 +0100
From: magnum <john.magnum@...hmail.com>
To: john-dev@...ts.openwall.com
Subject: Re: Serious bug in md5crypt-cuda - or in driver?

On 30 Oct, 2012, at 22:08 , Lukas Odzioba <lukas.odzioba@...il.com> wrote:

> 2012/10/30 magnum <john.magnum@...hmail.com>:
>> Is there any more error checking that can be made in CUDA than already is the case?
> 
> Hmm, please try this:
> 
> HANDLE_ERROR(cudaGetLastError());
> 
> Just after the:
> 
> kernel_crypt_r <<< BLOCKS, THREADS >>> (cuda_inbuffer, cuda_outbuffer);

Problem solved. The added error handling gave this:
"too many resources requested for launch in cryptmd5.cu at line 241"

I decided to add a similar check to all CUDA kernels and this quickly solved the three other problems I had too (mscash, wpapsk and pwsafe). All four problematic formats just needed a lower THREADS figure in their respective header file.

I did not commit these lower figures though we might consider doing that. I did commit the added error handling to all formats.

I also added some info in README-CUDA and a specific hint from HANDLE_ERROR() when this particular error comes up.

Thanks,
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.