Gabriele Gristina
b0c6738289
Rename STR() to M2S(), part 2
2022-02-07 13:31:22 +01:00
Jens Steube
be75e4b4ea
Rename STR() to M2S() to avoid future collisions and move from kernel source to command line parameter
2022-02-07 09:57:08 +01:00
Gabriele Gristina
9d36245d51
Kernels: Set the default Address Space Qualifier for any pointer, refactored / updated KERN_ATTR macros and rc4 cipher functions, in order to support Apple Metal runtime
2022-02-04 19:54:00 +01:00
Gabriele Gristina
2e4a136758
Refactored standard kernel includes in order to support Apple Metal runtime, updated backend, test units and status code
2022-01-18 22:52:14 +01:00
Jens Steube
0abdcb1ae5
Rename GID_MAX to GID_CNT to avoid naming conflict with existing macro
2022-01-04 22:57:26 +01:00
Jens Steube
668d2179cd
Kernels: Refactored standard kernel declaration to use a structure holding u32/u64 attributes to reduce the number of attributes
2022-01-04 18:07:18 +01:00
Jens Steube
b2d1f42905
Fix self-test functionality if FIXED_LOCAL_SIZE_COMP is used
...
Fix -m 25700 datatype in -a 3 mode and maximum password length in pure kernel mode
Fix -m 12500, 23700 and 23800 if password is exactly length 128
2021-08-02 12:00:17 +02:00
Jens Steube
1e3bd2c8a0
AMD GPUs: Add inline assembly code for md5crypt/sha256crypt, PDF 1.7, 7-Zip, RAR3, Samsung Android and Windows Phone 8+
2021-07-26 07:59:12 +02:00
Jens Steube
cf512faa53
Update large switch() cases in inc_common.cl and some inline assembly common functions for devices managed with HIP backend
2021-07-14 17:06:20 +02:00
Jens Steube
23bbe457a0
RAR3-hp Plugin: Replaced naive with true UTF8 to UTF16 conversion in optimized and pure kernel
2021-05-21 11:24:43 +02:00
Jens Steube
78a053800f
Remove module_unstable_warning() entries for AMD (legacy) driver after workaround with zero check on context length is in use
2021-05-09 10:33:05 +00:00
Jens Steube
9bf0f36d0a
Get rid of MAYBE_VOLATILE for context position by replacing it with zero length check
2021-05-09 11:43:32 +02:00
Jens Steube
e4dab0f1bf
OpenCL Runtime: Workaround JiT compiler segfault on legacy AMDGPU driver compiling RAR3 OpenCL kernel
2021-05-09 07:38:22 +00:00
Jens Steube
04d5e5a119
New Attack-Mode: Association Attack. Like JtR's single mode. Very early
...
stage. See hashcat Forum for detailed writeup.
2020-09-29 15:56:32 +02:00
Jens Steube
36480b99c1
Remove unused esalt declaration in -m 12500
2020-08-04 14:49:22 +02:00
philsmd
df5564eee2
cosmetic: make pure kernel of rar3-hp easier to read
2020-06-05 08:59:20 +02:00
philsmd
edfe21b902
fixes #2390 : RAR3-hp cracking w/ pass > 28 (pure kernel) + tests added
2020-05-25 11:08:33 +02:00
Jens Steube
d325158e65
Fix functions used in m12500-pure.cl
2020-02-04 12:55:02 +01:00
Jens Steube
02466bf404
Add pure kernel for rar3-hp to support passwords with more than 20 characters
2020-02-04 12:44:54 +01:00
Jens Steube
8ff8c5d536
Add LOCAL_VK to make use of __shared__
2019-05-07 09:01:32 +02:00
Jens Steube
89119bf24a
Add missing inc_platform.h include
2019-04-26 13:59:43 +02:00
Jens Steube
00e1e32492
Replace barrier() with SYNC_THREADS()
2019-04-26 13:34:07 +02:00
jsteube
7c6970dbdd
Remove hard-coded static keyword from OpenCL kernels
2019-04-13 18:46:19 +02:00
jsteube
1c45de4b63
Remove more old GCC diagnostic options
2019-03-31 19:41:28 +02:00
Jens Steube
0fb3b3c83e
Declare internal functions in OpenCL kernels as static
2019-03-26 11:03:25 +01:00
jsteube
e80b1838e8
Rename some functions in inc_common.cl to avoid conflicts with bitops.c
2019-03-23 22:15:38 +01:00
jsteube
adeeaee84a
Replace __kernel, __constant, __global and __local qualifiers with macro for better control
2019-03-22 22:27:58 +01:00
jsteube
c9d60c079f
Prepare OpenCL kernels for non-static compilation
2019-03-22 15:16:25 +01:00
jsteube
7d4bea41a0
Get rid of OpenCL/inc_hash_constants.h and OpenCL/inc_hash_functions.cl
2019-03-21 23:00:38 +01:00
Jens Steube
30681e5151
Move remaining module specific typedefs and structures from inc_types.cl to kernels
2019-03-08 11:14:33 +01:00
jsteube
dc9279c95c
New Strategy: Instead of using volatile, mark the mode as unstable. Remove all volatiles
2019-03-03 19:18:56 +01:00
jsteube
5da1e4b872
Fixed maximum password length limit which was announced as 256 but actually was 255
2019-02-26 21:20:07 +01:00
Jens Steube
2a6444c05a
Give the compiler a hint for automatic optimizations based on password length
2018-11-20 15:26:46 +01:00
R. Yushaev
5de004103a
Replace kernel parameter lists with macros
...
Substitute long parameter lists in ~2900 kernel function declarations
with macros. This cleans up the code, reduces probability of copy-paste
errors and highlights the differences between kernel functions. Also
reduces the size of the OpenCL folder by ~3 MB.
2018-11-16 11:44:33 +01:00
R. Yushaev
b80ada1d65
Unify esalt_bufs parameter declarations
...
In preparation for the abstraction of long repetitive kernel function
declarations, rename the salt buffer pointers to *esalt_bufs. Also
declare them const where they are not.
2018-11-16 10:28:54 +01:00
jsteube
68bff94980
Workaround rocm OpenCL runtime bug when copy data from constant to local memory
2018-08-12 18:04:33 +02:00
Jens Steube
02a2495349
Switched array pointer types in function declarations in order to be compatible with OpenCL 2.0
2018-07-22 11:47:42 +02:00
jsteube
32d6b3e10e
OpenCL kernels: Add '-pure' prefix to kernel filenames to avoid problems caused by reusing existing hashcat installation folder
2018-06-20 14:18:17 +02:00
