Jens Steube
5024865d87
Kernel Threads: Use warp size / wavefront size query instead of hardcoded values as base for kernel threads
...
Kernel Cache: Add kernel threads into hash computation which is later used in the kernel cache filename
Remove some unused function symbol lookups in HIP library
2021-07-22 11:46:47 +02:00
Jens Steube
e83611742d
Brain Session: Adds hashconfig specific opti_type and opts_type parameters to hashcat session computation to cover features like -O and -M
2021-07-21 15:47:26 +02:00
Jens Steube
c990e252d3
Added option --multiply-accel-disable (short: -M) to disable multiply the kernel-accel with the multiprocessor count automatism
2021-07-21 15:47:05 +02:00
Jens Steube
a7a899e5a4
Backport changes from #2888 to HIP backend
2021-07-21 14:01:28 +02:00
Jukka Ojanen
cb923d6e46
Replace CU_DEVICE_ATTRIBUTE_MAX_SHARED_MEMORY_PER_BLOCK with CU_DEVICE_ATTRIBUTE_MAX_SHARED_MEMORY_PER_BLOCK_OPTIN
2021-07-21 13:42:22 +03:00
Gabriele Gristina
7ea5da51a8
Added hash-mode: SNMPv3 HMAC-SHA1-96
2021-07-20 22:24:35 +02:00
Jukka Ojanen
55a475cb4b
Merge branch 'hashcat:master' into master
2021-07-20 21:27:51 +03:00
Jukka Ojanen
d23f2d6c2f
Calculation kernel dynamic memory size based on CU_DEVICE_ATTRIBUTE_MAX_SHARED_MEMORY_PER_BLOCK
2021-07-20 21:27:29 +03:00
Jens Steube
a026171fb3
Merge pull request #2879 from jtojanen/master
...
Implement gpu_bzero
2021-07-20 13:00:37 +02:00
Jens Steube
7c5fe5e233
Slow Kernels: Set some of the slowest kernels to OPTS_TYPE_MP_MULTI_DISABLE
2021-07-20 12:47:40 +02:00
Jukka Ojanen
8674e23d79
Add async HIP memcpy functions: hc_hipMemcpyDtoDAsync(), hc_hipMemcpyDtoHAsync() and hc_hipMemcpyHtoDAsync(). Implement partially async HIP memset and bzero kernels.
2021-07-20 12:47:10 +03:00
Jukka Ojanen
4263cafdcf
Add async CUDA memcpy functions: hc_cuMemcpyDtoDAsync(), hc_cuMemcpyDtoHAsync() and hc_cuMemcpyHtoDAsync(). Implement partially async CUDA memset and bzero kernels.
2021-07-20 12:23:39 +03:00
Jukka Ojanen
f07ff6f03d
Use kernel_bzero instead of kernel_memset when value is zero
2021-07-20 12:04:12 +03:00
Jukka Ojanen
4c0f6dd263
Merge branch 'hashcat:master' into master
2021-07-20 12:00:41 +03:00
Jukka Ojanen
ea5180ac46
Include missing bzero kernel parameters
2021-07-20 11:59:44 +03:00
Jens Steube
7d43b6e1eb
Clear tmps memory region after autotune use
2021-07-19 11:59:53 +02:00
Jens Steube
0504498e86
Merge pull request #2886 from pellekuiters/issue-2876
...
Auto-tuning outside kernel can cause unpredictable behavior for tmp_t structs on CUDA
2021-07-19 11:52:07 +02:00
pelle
ac3ccfcc51
Run init() and loop_prepare() during autotune to prevent possible corruptions during loop().
2021-07-19 11:07:24 +02:00
Jens Steube
257098a301
Get rid of hip/hip_runtime.h dependancy
2021-07-18 21:14:45 +02:00
Jukka Ojanen
72a418932c
HIP: Use kernel_bzero instead of kernel_memset when value is zero
2021-07-17 19:12:09 +03:00
Jukka Ojanen
a2a1d04bcf
Implement gpu_bzero
2021-07-17 19:00:10 +03:00
Jukka Ojanen
8066a47ac5
Merge branch 'hashcat:master' into master
2021-07-17 16:21:08 +03:00
Jens Steube
261e0f42ab
Unset PRODUCTION option
2021-07-16 23:03:11 +02:00
Jens Steube
84082a952e
hashcat 6.2.3
2021-07-16 22:41:52 +02:00
Jens Steube
141b59b80e
Merge pull request #2883 from hashcat/HIP
...
Merge HIP branch into master
2021-07-16 19:54:00 +02:00
Jens Steube
45e65dd05a
Backport more ROCm based optimizations to HIP
2021-07-15 23:34:27 +02:00
Jens Steube
2e929e692e
Backport update module_unstable_warning() for -m 21800 on HIP
2021-07-15 21:26:07 +02:00
Jens Steube
4730cf6e79
WinZip Kernel: Increase supported data length from 8k to 16mb
2021-07-15 16:42:29 +02:00
Jens Steube
d130cc66b3
Optimize ISA code on HIP for V_ALIGNBIT_B32 using a different template for inline assembly
2021-07-15 09:57:41 +02:00
Gabriele Gristina
80a671eab7
Fix bug on hm_SYSFS_CPU_get_syspath_hwmon()
2021-07-14 19:21:21 +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
11295e4679
Fix missing OPTI_TYPE_USES_BITS_64 in several modules
2021-07-14 17:01:46 +02:00
Jukka Ojanen
43b5a138d4
Use kernel_bzero instead of kernel_memset when value is zero
2021-07-14 17:13:39 +03:00
Jens Steube
a82fce2c8f
Fixed invalid ETA if --limit was specified, several salts are in a hash list and some of them were found in a potfile
2021-07-14 10:26:57 +02:00
Jens Steube
9c134833a6
Update module_unstable_warning() for -m 19600 and -m 19700
2021-07-14 08:26:12 +02:00
Jens Steube
f3bd936971
Add hardware monitor mapping for HIP devices
2021-07-14 08:23:39 +02:00
Jens Steube
84fcea9592
Fix unusued variable warning
2021-07-14 08:17:55 +02:00
Jens Steube
4b68a8fb4f
Merge pull request #2878 from hashcat/master
...
Backports
2021-07-13 22:26:28 +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
74fca7bac1
Merge pull request #2875 from jtojanen/master
...
Fix leaks in dictstat_destroy
2021-07-13 13:28:37 +02:00
Jens Steube
219bed457f
Fix use of --gpu-max-threads-per-block
2021-07-12 14:02:43 +02:00
Jukka Ojanen
6b5af5e849
Fix leaks in dictstat_destroy
2021-07-12 14:33:24 +03:00
Jens Steube
674ca7d88f
Add GPU threads to kernel cache checksum because it has an influence on HIP offline compile options
...
Add V_ALIGNBIT_B32 inline assembly wrapper because HIP does not provide amd_bitalign()
2021-07-12 11:27:05 +02:00
Jens Steube
23c3c178bf
Limit max threads per block to 64 to enable offline compiler to make better use if available registers
...
Fix double free() for hip_event1/hip_event2 and hip_stream causes segfault
Replace hc_cuCtxSetCurrent() with hc_cuCtxPushCurrent() in order to align changes with HIP
Add vector datatype operators (if we decide to use them - currently unused)
2021-07-12 09:28:26 +02:00
Jens Steube
20f7febd4c
Workaround too intensive optimization in -m 2000 using HIPRTC
2021-07-11 15:54:13 +02:00
Jens Steube
fde4770219
Add misssing HIP backend information in -I output
2021-07-11 14:47:48 +02:00
Jens Steube
ca3beacd93
Disable dynamic shared memory on HIP, because hipFuncSetAttribute() maps to cudaFuncSetAttribute() and not to cuFuncSetAttribute()
2021-07-11 14:30:49 +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
Jens Steube
9fc6c26f8c
Added temperature watchdog and utilization for CPU on linux using sysfs and procfs
2021-07-10 14:24:51 +02:00
Jens Steube
bfe83ec138
Added temperature watchdog for CPU on linux using sysfs
2021-07-10 08:43:15 +02:00
Jens Steube
72d7048b7e
Merge branch 'master' of https://github.com/hashcat/hashcat
2021-07-09 20:48:29 +02:00
Jens Steube
979f9e9868
Rename hardware monitor interface sysfs to sysfs_amdgpu
2021-07-09 20:48:10 +02:00
Jens Steube
d6042035fc
Merge pull request #2870 from jtojanen/fixes
...
Avoid deadlocks in brain server
2021-07-09 08:43:18 +02:00
reger-men
ea7b74389f
First draft HIP Version
2021-07-09 03:50:40 +00:00
Jukka Ojanen
d93d130b6d
Avoid deadlocks in brain server
2021-07-08 12:43:29 +03:00
Jens Steube
71c0ea44b7
Update hwmon to easier distinguish between CPU and GPU for OpenCL backend
2021-07-08 09:04:14 +02:00
Jens Steube
47848d2b64
Merge pull request #2867 from jtojanen/fixes
...
Improve affinity mask handling
2021-07-08 08:21:52 +02:00
Jens Steube
4cbf7900b6
Remove unused code in -m 26300
2021-07-08 06:25:35 +02:00
Jens Steube
4535695e0a
Update some module_unstable_warning() on macOS after the workaround that was added before
2021-07-07 10:36:52 +02:00
Jens Steube
f146a05878
Added option --markov-inverse to inverse markov statistics, with the idea of reversing the order of the password candidates
2021-07-07 10:32:08 +02:00
Jukka Ojanen
0edab2b20f
Allow CPU affinity mask up to 64 processors in Windows. Remove call to SetThreadAffinityMask as SetProcessAffinityMask limits all threads in process. Report error code. pthread_setaffinity_np returns non-zero if failure; works also with OSX as it reports KERN_SUCCESS which is zero.
2021-07-06 16:47:18 +03:00
Jens Steube
17a47e390c
Merge branch 'master' of https://github.com/hashcat/hashcat
2021-07-06 13:06:28 +02:00
Jens Steube
a3178391c2
Fixed missing option to automatically disable kernel cache in -m 25600 and -m 25800
2021-07-06 13:05:49 +02:00
Jens Steube
29c971187a
Merge pull request #2863 from matrix/metamask
...
Added hash-mode: MetaMask Wallet
2021-07-06 09:49:19 +02:00
Jens Steube
13e2aa2508
Merge pull request #2865 from jtojanen/fixes
...
Fix several backend leaks
2021-07-06 09:34:47 +02:00
Jukka Ojanen
a0eaefa0c2
Missing whitespaces
2021-07-05 20:20:51 +03:00
Jukka Ojanen
e133bd4ec4
Change rc_board_name_amd initial value to CL_INVALID_VALUE. If CHECK_BOARD_NAME_AMD is changed to 0, there is a problem with CL_SUCCESS which equals to 0, device will be skipped.
2021-07-05 19:05:10 +03:00
Jukka Ojanen
b3d18f86e2
Fix early return leaks in backend_session_begin
2021-07-05 19:03:56 +03:00
Jukka Ojanen
b976e52bc7
Fix early return leaks in load_kernel; nvrtc_options, nvrtc_options_string, build_log. Ensure build log NULL termination.
2021-07-05 19:00:35 +03:00
Jukka Ojanen
2f7eec2fd7
Fix early return leaks in backend_ctx_init and backend_ctx_devices_init
2021-07-05 15:52:48 +03:00
Jukka Ojanen
bcbb9b0d2c
Fix skipped device param leak in backend_ctx_devices_destroy
2021-07-05 15:38:07 +03:00
Gabriele Gristina
6ce1c78f6d
fix CT_MAX_LEN_BASE64 define
2021-07-05 13:21:10 +02:00
Jens Steube
a2f883396b
Java Object hashCode: Add OPTS_TYPE_SUGGEST_KG as a default option
2021-07-05 09:25:06 +02:00
Gabriele Gristina
7371cbebe7
add min/max data length limits to MetaMask Wallet
2021-07-04 20:14:25 +02:00
Gabriele Gristina
2d149164d2
cleanup module_26600.c
2021-07-04 15:50:52 +02:00
Gabriele Gristina
f571329c4f
Added hash-mode: MetaMask Wallet
2021-07-04 15:47:52 +02:00
Jens Steube
d99b4cf46e
Merge pull request #2861 from jtojanen/fixes
...
Fix iconv_ctx and iconv_tmp leaks in backend.c
2021-07-04 09:52:29 +02:00
Jens Steube
7366c47447
Added hash-mode: iPhone passcode (UID key + System Keybag)
2021-07-03 19:03:06 +02:00
Jens Steube
3c899ec528
Potfile: Disable potfile for hash-mode 99999
2021-07-03 14:22:14 +02:00
Jukka Ojanen
5f109b5862
Fix iconv_ctx and iconv_tmp leaks in backend.c
2021-07-03 12:51:37 +03:00
Jens Steube
c32242980f
Partially revert
...
6967e706c2
because it breaks --hash-info. See
https://github.com/hashcat/hashcat/issues/2859 for details.
2021-07-02 08:58:04 +02:00
Jens Steube
ce41316ac3
Status View: Include time and duration info when pausing and resuming
2021-07-01 21:33:14 +02:00
Jens Steube
701ad7c441
Fix segfault in --hash-info by avoid huge stack buffer allocation
2021-06-30 21:01:55 +02:00
Jens Steube
56c2243dfb
KeePass: Increase supported size for KeePass 1 databases from 300kB to 16MB
2021-06-30 13:16:03 +02:00
Jens Steube
2aff6cba51
Add ranges to charset lists
2021-06-29 22:27:50 +02:00
Jens Steube
254e33c473
File handling: Do not abort on seeing a BOM in input files, just warn and ignore the BOM
2021-06-29 20:42:22 +02:00
Jens Steube
2af45cd03f
Rename hc_fopen_nozip() to hc_fopen_raw()
2021-06-29 20:13:11 +02:00
Jens Steube
5bea13200f
Add missing newline in warning message when using --force
2021-06-28 23:09:41 +02:00
Gabriele Gristina
04b9050af9
fix tuning_db_search() bug
2021-06-28 13:08:11 +02:00
Royce Williams
e24b82cc4e
minor diffs from comparing with example hashes wiki
2021-06-27 14:38:36 -08:00
Jens Steube
77d9a8deb2
minor whitespace fixes and changes.txt update
2021-06-27 10:43:23 +02:00
Jens Steube
3b07f90524
Merge pull request #2854 from TheWorkingDeveloper/patch-1
...
Hexify plains with delimiter in debug rules
2021-06-27 10:37:38 +02:00
Jens Steube
3a3bd07514
Make it more visible we use a hardcoded character
2021-06-27 10:37:15 +02:00
Jens Steube
8a3eee3fe5
OpenCL Runtime: Workaround JiT crash (SC failed. No reason given.) on macOS by limiting local memory allocations to 32k
2021-06-27 10:18:38 +02:00
TheWorkingDeveloper
1bf3392b1a
Hexify plains with delimiter
...
Hexify plains if the plain contains the : separator
2021-06-26 23:53:04 +02:00
Jens Steube
dcaba1f473
Brain: Add brain_ctx_t to hashcat_ctx_t to enable runtime check if hashcat was compiled with brain support
2021-06-26 20:24:45 +02:00
Jens Steube
909d5e64a5
Added hash-mode: AES-128/192/256-ECB NOKDF
...
This mode is probably very rare in real-life scenarios,
but it is a nice template for kernels which do
not use a KDF,
or use AES,
or simple fast hashes with lookup tables
or simple optimized kernels in general
2021-06-26 17:12:10 +02:00
Jens Steube
b30d55c5d9
Removed LZMA "Windows.h" capitalized file name workaround used to cross compile Windows binaries from Linux hosts.
...
This breaks compiling windows binaries if the underlaying filesystem is not case sensitive. An example would be using /mnt/c folder under WSL.
Any update on the LZMA SDK needs to be post processed with this command: sed -i 's/<Windows.h>/<windows.h>/' deps/LZMA-SDK/C/*.c deps/LZMA-SDK/C/*.h
2021-06-25 20:30:26 +02:00
Jens Steube
5194beb6f2
Merge pull request #2847 from philsmd/patch-3
...
fixes #2839 : change example pass for FortiGate256
2021-06-25 07:57:23 +02:00