arno/hashcat
1
0
Fork 0
mirror of https://github.com/hashcat/hashcat.git synced 2024-11-16 04:49:24 +00:00
Code Issues Releases Wiki Activity
8,304 Commits 1 Branch 27 Tags 191 MiB
95405ddc2a
Commit Graph

39 Commits

Author SHA1 Message Date
Gabriele Gristina
b3d3b31c3e Metal: added support for vectors up to 4 2022-02-10 21:53: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
Jens Steube
cb7f99ef79 Renamed macro INLINE to HC_INLINE to avoid naming conflict with INLINE on MacOS 2021-12-24 16:40:43 +01:00
Jens Steube
89cd5bd78b Remove inline static keyword in inc_vendor.h for HIP platform since it's the default setting with HIP 4.4 2021-11-02 08:12:13 +01:00
Jens Steube
aee8e559c4 PDF Kernel (10700): Improved performance on AMD GPU by using shared memory for the scratch buffer and disable inlining to save spilling 
Inspired by https://github.com/reger-men/hashcat/blob/6.2.4/OpenCL/m10700-optimized.cl
 2021-10-31 10:05:58 +01:00
Jens Steube
3d4e2aec43 Work around segmentation fault in Intel JiT 2021.12.6.0.19_160000 compiling hc_enc_next()/hc_enc_next_global() 2021-08-03 08:34:37 +02:00
Jens Steube
3f6c5a0042 Update module_unstable_warning() for -m 172xx on HIP 2021-07-23 21:09:55 +02:00
Jens Steube
257098a301 Get rid of hip/hip_runtime.h dependancy 2021-07-18 21:14:45 +02:00
Jens Steube
45e65dd05a Backport more ROCm based optimizations to HIP 2021-07-15 23:34:27 +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
7faf6859d6 Backport hand-optimized compiler settings in modules from ROCM to HIP 
Backport DECLSPEC settings from ROCM to HIP
 2021-07-13 20:45:01 +02:00
Jens Steube
1b84a9e53b Add missing backports from code base v6.2.2 
Fix context to thread management
Fix missing code in selftest.c, autotune.c, hashes.c, dispatch.c and backend.c
Use IS_HIP depending code makes it easier for future optimization related to inline assembly calls - instead of using IS_CUDA || IS_HIP
See TODO markers for more optimizations / next steps
 2021-07-11 12:38:59 +02:00
Jens Steube
a22f8149fc
Merge branch 'HIP' into hip 2021-07-10 21:34:09 +02:00
reger-men
ea7b74389f First draft HIP Version 2021-07-09 03:50:40 +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
ddb641b843 Add option to force disable real SHM access to be used from within the module 2020-03-20 16:20:22 +01:00
Jens Steube
61fe90bacb Use oldschool SHA1 kernel for CPU it's slightly faster 2020-03-03 12:36:55 +01:00
Jens Steube
b4bac70bd6 Remove inline keyword in DECLSPEC for CPU 2020-03-03 08:52:26 +01:00
Jens Steube
e53bff0fb0 Reenable bitselect() and rotate() on Intel SDK 2020-03-02 16:07:13 +01:00
Jens Steube
c90d83c3eb Prepare for UNROLL whitelisting 2020-02-15 12:44:12 +01:00
Jens Steube
3561e7b8d7 Add special ROCM detection in OpenCL/inc_vendor.h 2020-01-25 12:09:39 +01:00
Jens Steube
3a5544a554 Help some compiler with 64 bit constants 2020-01-21 22:09:56 +01:00
Jens Steube
cf4cee2f2f Update selection of API to make use of bitselect and rotate 2020-01-20 09:20:12 +01:00
Jens Steube
89f9ef45b6 Whitelist some OpenCL specific functions 2020-01-12 13:32:02 +01:00
Jens Steube
8ff8c5d536 Add LOCAL_VK to make use of __shared__ 2019-05-07 09:01:32 +02:00
Jens Steube
d0bd33c9d1 Rename CONSTANT_AS to CONSTANT_VK 2019-05-06 14:34:16 +02:00
Jens Steube
5ee033673c Disable name mangling in NVRTC's PTX output and more 2019-05-03 15:50:07 +02:00
Jens Steube
9faba41848 Use nvrtc to compile PTX (resulting PTX not yet used) 2019-04-26 13:28:44 +02:00
Jens Steube
4b986de5fb Prepare native CUDA hybrid integration 2019-04-25 14:45:17 +02:00
Jens Steube
38c1029f2e Need volatile for IRIS GPU on Mac OSX for -m 2500 and -m 2501 2019-04-17 13:21:35 +02:00
jsteube
7c6970dbdd Remove hard-coded static keyword from OpenCL kernels 2019-04-13 18:46:19 +02:00
jsteube
b7cdca09c4 OpenCL Runtime: Workaround JiT compiler error on ROCM 2.3 driver if the 'inline' keyword is used in function declaration 2019-04-13 13:46:55 +02:00
jsteube
d7d716f3ab Make it easier to include OpenCL kernels into modules 2019-04-04 20:01:37 +02:00
jsteube
9ced13cc94 Get rid of CONSTSPEC macro in OpenCL kernels 2019-04-04 10:15:34 +02:00
Jens Steube
0fb3b3c83e Declare internal functions in OpenCL kernels as static 2019-03-26 11:03:25 +01:00
jsteube
66d94b06e4 Get rid of src/rp_kernel_on_cpu.c and src/rp_kernel_on_cpu_optimized.c and use OpenCL emulated kernel version 2019-03-25 12:24:04 +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
7d4bea41a0 Get rid of OpenCL/inc_hash_constants.h and OpenCL/inc_hash_functions.cl 2019-03-21 23:00:38 +01:00