arno/hashcat
1
0
Fork 0
mirror of https://github.com/hashcat/hashcat.git synced 2025-07-07 23:28:11 +00:00
Code Issues Releases Wiki Activity
9,736 Commits 1 Branch 27 Tags 227 MiB
957f3b8ac9
Commit Graph

396 Commits

Author SHA1 Message Date
Gabriele Gristina
7650894e02 fixed bug in benchmark engine, updated negative status code 2022-01-11 04:58:30 +01:00
Jens Steube
dfd316c653
Merge pull request #3103 from matrix/backend_session_update_mp_rl 
Removed hc_clSetKernelArg() call from backend_session_update_mp_rl()
 2022-01-09 13:22:00 +01:00
Jens Steube
7a9a1b37d0
Merge pull request #3104 from matrix/backend_session_update_mp 
Removed hc_clSetKernelArg() call from backend_session_update_mp()
 2022-01-09 13:21:52 +01:00
Jens Steube
56ef2b4bde
Merge pull request #3102 from matrix/backend_cuda_restyle 
CUDA Backend: moved functions to ext_cuda.c/ext_nvrtc.c and includes to ext_cuda.h/ext_nvrtc.h
 2022-01-06 11:29:56 +01:00
Jens Steube
045ca5cb7a Fixed method how OPTS_TYPE_AUX* kernels are called in association mode, for instance WPA/WPA2 kernels 2022-01-05 10:46:00 +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
Gabriele Gristina
994083eaf5 Removed hc_clSetKernelArg() call from backend_session_update_mp() 2022-01-03 18:28:19 +01:00
Gabriele Gristina
0f0cf1fe08 Removed hc_clSetKernelArg() call from backend_session_update_mp_rl() 2022-01-03 17:16:27 +01:00
Gabriele Gristina
f8ceb8785e CUDA Backend: moved functions to ext_cuda.c/ext_nvrtc.c and includes to ext_cuda.h/ext_nvrtc.h 2022-01-03 16:29:15 +01:00
Gabriele Gristina
78c7ee2af6 HIP Backend: moved functions to ext_hip.c/ext_hiprtc.c and includes to ext_hip.h/ext_hiprtc.h 2022-01-02 19:12:41 +01:00
Gabriele Gristina
26b6054cab OpenCL Backend: moved functions to ext_OpenCL.c and includes to ext_OpenCL.h 2021-12-28 21:55:05 +01:00
Gabriele Gristina
861e644057 OpenCL Backend: added workaround to make optimized kernels work on Apple Silicon 2021-12-24 17:30:49 +01:00
Jens Steube
df6e5480ca Print module_extra_tuningdb_block undefined compute device warning only on GPU 2021-12-23 19:23:35 +01:00
Gabriele Gristina
3fd6dac523 Set default device-type to GPU with Apple M1 2021-12-14 00:37:17 +01:00
Gabriele Gristina
0fae3a4394 Added support for Apple Silicon compute devices 2021-12-13 22:42:42 +01:00
Jens Steube
d4a54287b1 Add missing backslash for RUN_INSTRUCTION_CHECKS() on AMD 2021-12-09 14:03:32 +01:00
Jens Steube
3d53188cc3 Tuning Database: Added a warning if a module implements module_extra_tuningdb_block but the installed computing device is not found 2021-12-04 16:56:16 +01:00
Jens Steube
21f91c5bb8 Module Optimizaters: Added OPTS_TYPE_MAXIMUM_THREADS to deactivate the else branch route in the section to find -T before compilation 
Set the new flag based on some testings with RX6900XT
 2021-11-13 17:27:06 +01:00
Gabriele Gristina
9be7bc71a5 OpenCL Backend: added workaround to support Apple Silicon 2021-11-11 00:57:54 +01:00
Jens Steube
53f6693495 Temporary enable HIP 4.4/ROCM 4.5 on Linux and globally set native thread count 2021-11-10 19:32:54 +01:00
Jens Steube
f84aca82ca Backend types: The default filter for the device types is now set so that only the GPU is used, except for APPLE, where we set CPU 2021-11-01 10:49:20 +01:00
Jens Steube
49a68cd6c1 AMD Driver: Updated requirements for AMD Linux drivers to ROCm 4.5 or later due to new HIP interface 2021-10-21 20:03:01 +02:00
Jens Steube
576a71af5c Update minimum HIP version from 4.4 to upcoming 4.5 2021-10-19 19:26:10 +02:00
Jens Steube
756c29ec57 Add missing cleanup on windows if outdated HIP version is detected 2021-10-13 20:00:28 +02:00
Jens Steube
733f9c2d77 Add better detection future HIP 4.4 2021-10-12 20:01:14 +02:00
Jens Steube
07e58631a5 Backend devices: In non -S mode, limit the number of workitems so that no more than 4GB of host memory is required per backend device 2021-10-04 08:36:56 +02:00
Jens Steube
4b6654b503 Fix unstable plugin to driver warning 2021-09-24 13:35:18 +02:00
Jens Steube
c1fd42fe72 Reduce work item maximum in -S mode even further, tested with NTLM 2021-09-23 12:46:00 +02:00
Jens Steube
bd2cde31ae Back-end devices: In -S mode, limit the number of workitems so that no more than 2GB of host memory is required per backend device 2021-09-22 13:03:25 +02:00
Jens Steube
4ef1509bc7 Backend Devices: Reduce maximum workitems limited derived from available host memory down from to 8GB to 4GB per backend device 2021-09-21 13:01:15 +02:00
Jens Steube
721e1ea54d Fixed division by zero because backend_ctx->hardware_power_all was not re-inserted after refactoring device_param->hardware_power 2021-09-08 17:37:18 +02:00
Jens Steube
8c14fd85ea POCL: Added a workaround for an issue in POCL that uses a quote character as part of the path itself given to a path for the -I option 2021-09-05 10:55:59 +02:00
Jens Steube
b4b2195fa5 OpenCL Runtime: The use of clUnloadPlatformCompiler () was disabled after some users received unexpected return codes 2021-08-24 09:37:49 +02:00
Jens Steube
85854236d1
Merge pull request #2935 from matrix/apple_gpu_workaround 
workaround to 'clEnqueueWriteBuffer(): CL_INVALID_VALUE' with apple gpu
 2021-08-21 08:04:37 +02:00
Jens Steube
50e28ff306
Merge pull request #2926 from jtojanen/master 
Code cleanup and small fixes
 2021-08-19 14:59:16 +02:00
Jukka Ojanen
6b4786de84 Make blocking clEnqueueWriteBuffer() non-blocking 2021-08-18 20:19:51 +03:00
Jukka Ojanen
f3cd32bffd Avoid memory copy in rebuild_pws_compressed_append() and assign instead 2021-08-18 17:46:39 +03:00
Jens Steube
76e388ea47 Memory Management: Check in OpenCL that CL_DEVICE_HOST_UNIFIED_MEMORY is true and if so, then half of the available memory 
NEO driver: Activate the NEO driver (Intel GPU) after it passed several tests in hashcat using the latest NEO driver version
 2021-08-18 16:10:51 +02:00
Jens Steube
9254603960 Fixed out-of-boundary read in PKZIP masterkey kernel if the password candidate has length zero. 
Now that kernel threads are no longer fixed over the entire session, hardware_power and hardware_power_all
needs to be updated the same way as kernel_power and kernel_power_all for each call to inner2_loop().
 2021-08-17 17:38:45 +02:00
Gabriele Gristina
72735ed7c4 update workaround 2021-08-13 20:17:09 +02:00
Gabriele Gristina
fef18965c3 workaround to 'clEnqueueWriteBuffer(): CL_INVALID_VALUE' with apple gpu 2021-08-13 16:00:32 +02:00
Jukka Ojanen
5c0820b8d3 Assume hashconfig->tmp_size is the element size 2021-08-13 13:13:12 +03:00
Jukka Ojanen
bcb4a8f391 Add clFlush() to backend_session_update_mp() and backend_session_update_mp_rl() 2021-08-10 01:49:55 +03:00
Jukka Ojanen
3b11b51113 Remove unnecessary hc_cuEventSynchronize() call 2021-08-09 18:58:04 +03:00
Jukka Ojanen
1edb649c1f Remove duplicate conditions 2021-08-09 18:48:20 +03:00
Jens Steube
5e1d37c82e Make unsupported AMD HIP runtime version error message a bit more human readable 2021-08-05 18:56:33 +02:00
Jens Steube
cb69e2d413 Added some HIP version checks, fall back to OpenCL automatically 
Switched HIP version check from driverVersion to runtimeVersion
Fixed syntax check of HAS_VPERM macro in several kernel includes causing invalid error message for AMD GPUs on Windows
Updated AMD driver requirements
Updated docs/changes.txt with missing changes from previous commits
Fixed invalid vector data type in Murmur Hash in -a 3 mode
Fixed uninitialized variable warning in src/hashes.c
Fixed broken support for --generate-rules-func-min
 2021-08-04 20:49:22 +02: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
af5d346244
Merge pull request #2894 from jtojanen/master 
Allow async execution
 2021-08-02 10:05:20 +02:00
Jens Steube
20a7b9f992 Tuning-Database: Add new module function module_extra_tuningdb_block() to extend hashcat.hctune content from a plugin 
See src/modules/module_08900.c as an example
 2021-08-01 16:25:37 +02:00
Jukka Ojanen
3d7ce7162b Merge branch 'master' of https://github.com/hashcat/hashcat 2021-08-01 15:36:52 +03:00
Jukka Ojanen
62a06f735f kernel_memset32(): assume offset and size are bytes, not elements 2021-08-01 15:19:15 +03:00
Jukka Ojanen
fdbfae9a28 Modify OpenCL clEnqueueFillBuffer() workaround 2021-08-01 14:47:08 +03:00
Jukka Ojanen
dbe2bad098 Maintain code style 2021-08-01 14:26:33 +03:00
Jukka Ojanen
1ee222d43f Replace free() with hcfree() 2021-08-01 14:21:18 +03:00
Jukka Ojanen
e352a79a05 Extend context in gidd_to_pw_t() 2021-08-01 14:15:03 +03:00
Jens Steube
d4997d1255 Added support for auto-tuning --kernel-threads (-T) on startup 2021-07-31 16:49:39 +02:00
Jukka Ojanen
e154f9e781 Few cleanups 2021-07-30 12:24:21 +03:00
Jukka Ojanen
c3195d0603 Merge branch 'master' of https://github.com/hashcat/hashcat 2021-07-30 11:34:25 +03:00
Jukka Ojanen
1064cce08c Synchronize before hooks 2021-07-29 15:22:25 +03:00
Jukka Ojanen
81c2ec3caf Small cleanup in gidd_to_pw_t() 2021-07-29 14:48:36 +03:00
Jukka Ojanen
7a8065d090 Do not call clWaitForEvents() after spin damper when we know that event status is CL_COMPLETE 2021-07-29 14:39:11 +03:00
Jukka Ojanen
c48e6a25a8 Enqueue several commands before clFlush() 2021-07-29 14:28:01 +03:00
Jens Steube
03ed06849a
Merge branch 'HIP44' into master 2021-07-29 11:00:28 +02:00
Jens Steube
d38d40c8ba Unlock all GPU threads for AMD GPUs if WaveFront size is 32 (basically new models) 
Add new hash-modes to tools/benchmark_deep.pl
Fix MINGW issue on 64 bit constant in refactored kernel-accel limiting section
 2021-07-29 10:49:44 +02:00
Jukka Ojanen
9ed231c99c Add comment to blocking OpenCL calls 2021-07-29 00:07:10 +03:00
Jens Steube
a4299b74af Memory Management: Refactored the code responsible for limiting kernel accel in order to avoid out of -host- memory situations 2021-07-28 20:50:05 +02:00
Jukka Ojanen
d7cc8d7cd2 Revert due to module_03200.c, module_25600.c and module_25800.c using device_param->kernel_dynamic_local_mem_size_memset 2021-07-28 19:49:02 +03:00
Jukka Ojanen
8b590f651b Remove unused gpu_memset and its references 2021-07-28 19:26:30 +03:00
Jens Steube
9c0a37accf Update driver requirement 2021-07-28 07:56:34 +02:00
Jukka Ojanen
cdf27a1cb3 Implement async run_cuda_kernel_memset() and run_cuda_kernel_memset32() 2021-07-27 18:56:59 +03:00
Jukka Ojanen
d7de3550b1 Add fast event for CUDA and HIP, and use it to synchronize check_hash() 2021-07-27 12:57:26 +03:00
Jens Steube
f6de3e61e0
Merge pull request #2906 from hashcat/master 
Backports
 2021-07-27 09:48:56 +02:00
Jens Steube
fd2cb59d26 AMD GPUs: On Apple OpenCL platform, we ask for the preferred kernel thread size rather than hard-coding 32 
ECC secp256k1: Removed the inline assembly code for AMD GPUs because the latest JIT compilers optimize it with the same efficiency
 2021-07-27 09:37:31 +02:00
Jukka Ojanen
a642f7b233 Remove synchronous GPU memory copy functions 2021-07-26 15:36:42 +03:00
Jukka Ojanen
de5200cffc Allow async execution of backend 2021-07-26 15:17:25 +03:00
Jens Steube
72e307fbce
Merge pull request #2901 from hashcat/master 
Backport changes
 2021-07-25 10:57:39 +02:00
Jens Steube
959a232828
Merge pull request #2885 from neheb/charfixes 
const and char fixes
 2021-07-25 10:36:40 +02:00
Jens Steube
84a4058edf
Merge pull request #2900 from hashcat/master 
Backport changes
 2021-07-25 10:34:05 +02:00
Jens Steube
640d95a00f Vendor Detection: Add "Intel" as a valid vendor name for GPU on macOS 2021-07-24 16:01:30 +02:00
Jens Steube
5ffcaa980d HIP Backend: Added support to support HIP 4.4 and later, but added check to rule out older versions because they are incompatible 2021-07-23 16:04:34 +02:00
Rosen Penev
adaf3f293b make const char pointers actually const 
const char* is a non const pointer that points to const data. Add
missing const.

Signed-off-by: Rosen Penev <rosenp@gmail.com>
 2021-07-22 18:13:46 -07:00
Jukka Ojanen
9f9333f2ef Allow async execution of run_opencl_kernel_bzero(), run_hip_kernel_bzero() and run_opencl_kernel_bzero() 2021-07-22 15:18:10 +03:00
Jukka Ojanen
fafd24237d Define HC_ALIGN macro to control data alignment and use common constant zero buffer in run_cuda_kernel_bzero(), run_hip_kernel_bzero() and run_opencl_kernel_bzero(). 2021-07-22 14:24:03 +03:00
Jukka Ojanen
2c2988518d Remove all calls to clFlush() 2021-07-22 13:59:19 +03:00
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
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
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
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
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
257098a301 Get rid of hip/hip_runtime.h dependancy 2021-07-18 21:14:45 +02:00
Jukka Ojanen
a2a1d04bcf Implement gpu_bzero 2021-07-17 19:00:10 +03: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
f3bd936971 Add hardware monitor mapping for HIP devices 2021-07-14 08:23:39 +02:00
Jens Steube
219bed457f Fix use of --gpu-max-threads-per-block 2021-07-12 14:02:43 +02: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
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
bfe83ec138 Added temperature watchdog for CPU on linux using sysfs 2021-07-10 08:43:15 +02:00
Jens Steube
979f9e9868 Rename hardware monitor interface sysfs to sysfs_amdgpu 2021-07-09 20:48:10 +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
Jukka Ojanen
5f109b5862 Fix iconv_ctx and iconv_tmp leaks in backend.c 2021-07-03 12:51:37 +03: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
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
28c08de9ef Folders: Do not escape the variable cpath_real to prevent certain OpenCL runtimes from running into an error which do not support escape characters 2021-06-24 09:24:02 +02:00
Jens Steube
90f0e78b5b Add event about autodetection process start 
Remove "..." substring from final messages
Add patient message on "..." startup messages
Add missing docs/changes.txt entry
 2021-06-22 11:37:59 +02:00
Jens Steube
7e267b9b37
Merge pull request #2825 from matrix/hwmon_osx_v2 
Add OSX HW Monitor initial support (2021)
 2021-06-20 11:09:35 +02:00
Jens Steube
2d7f67fd23 Alias Devices: Show a warning in case the user specifically listed a device to use which in a later step is skipped because it is an alias of another active device 2021-06-15 14:06:24 +02:00
Gabriele Gristina
bc4ce4cbeb Add support for CPU/GPU device temperature and fanspeed using iokit (Apple) 2021-06-12 20:13:31 +02:00
Jens Steube
ee7fca82f5 Scrypt Kernels: Re-enable scrypt based kernels to use kernel cache 2021-06-12 11:42:19 +02:00
Jens Steube
7fc0ac4ef1 Do not initialize backend devices in case --id is used and some bug fixes 2021-06-12 10:47:48 +02:00
Jens Steube
65d81c0f7b Hashrate: Innerloop hashrate prediction requires update because of the new salt_repeats feature and also respect _loop2 kernel runtime 2021-06-10 22:43:06 +02:00
Jens Steube
fdccc8287d Fixed free memory size output for skipped GPU (both automatic and manual) of --backend-info information screen 2021-06-07 12:12:01 +02:00
Royce Williams
7bea7ca177 tighten output to be physical-terminal friendly 2021-06-06 10:47:18 -08:00
nycex
470e844e5d
use XDG_CACHE_HOME for kernels 2021-06-05 00:38:43 +02:00
Jens Steube
ce8c121b50 BCRYPT Kernels: Improved bcrypt performance by 6.5% for high-end NVIDIA GPU devices using CUDA backend 2021-06-01 22:52:07 +02:00
Jens Steube
9f5a22a3ab OpenCL Backend: Use CL_DEVICE_BOARD_NAME_AMD instead of CL_DEVICE_NAME for device name in case OpenCL runtime supports this query 2021-05-26 19:09:50 +02:00
Jens Steube
5c6501444a Kernels: Add standalone true UTF8 to UTF16 converter kernel that runs after amplifier. Use OPTS_TYPE_POST_AMP_UTF16LE from plugin 2021-05-20 14:34:24 +02:00
Jens Steube
fe91f6276d CUDA Backend: Do not warn about missing CUDA SDK installation if --stdout is used 2021-05-16 18:31:43 +02:00
Jens Steube
a2fcb03fe3 Update AMD GPU check on macOS warning message 2021-05-11 10:54:39 +02:00
Jens Steube
282eb75fe9 Update module_unstable_warning for benchmark long selection on macOS for CPU and GPU; Allow use of GPU without --force testwise 2021-05-10 19:57:56 +02:00
Jens Steube
bbd6e55968 Add missing null pointer to cuda_module_shared and opencl_program_shared in backend cleanup function 2021-05-09 14:55:52 +02:00
Jens Steube
0c2afde83b Add support for clUnloadPlatformCompiler() 2021-05-02 08:15:25 +00:00
Jens Steube
59459d0e5b Fixed memory leak causing problems in sessions with many iterations. for instance, --benchmark-all or large mask files 2021-05-01 17:27:33 +02:00
Jens Steube
9a87d5aa01 Fixed out-of-boundary reads in case user activates -S for fast but pure hashes in -a 1 or -a 3 mode 2021-04-27 19:55:30 +02:00
Jens Steube
0ba77fe761 Kernel Development: Kernel cache is disabled automatically in casehashcat is compiled with DEBUG=1 
See https://github.com/hashcat/hashcat/issues/2750
 2021-04-26 09:51:50 +02:00
Jens Steube
1dac869cb7 Removed unnecessary swaps in SCRYPT based algorithms 2021-04-23 20:55:13 +02:00
Jens Steube
15f35fa68c Scrypt Kernels: Reduced kernel wait times by making it a true split kernel where iteration count = N value 2021-04-21 15:59:14 +02:00
Jens Steube
8e47fdf8f5 Add 4 times single workitem extra buffer size to total extra buffer size to workaround single workitem buffer overflows 2021-04-19 10:27:51 +02:00
Jens Steube
57a8923b81 Update complete SCRYPT workload tuning logic. 
A detailed description will follow.
Set -m 8900 defaults to 16k:8:1 (default scrypt settings).
 2021-04-16 20:17:53 +02:00
Jens Steube
ff96015f53 Add OPTS_TYPE_NATIVE_THREADS for use by plugin developer to enforce native thread count (useful for scrypt) 2021-04-14 15:22:30 +02:00
Jens Steube
67d189e10a Update calculation of EXTRA_SPACE in backend.c and add upper and lower hard limit 2021-04-13 12:02:52 +02:00
Jens Steube
51e8661070 Update calculation of EXTRA_SPACE in backend.c to make it depending from kernel-accel 2021-04-13 11:47:37 +02:00
Jens Steube
9033975efd Allow plugins to disable the multiplication of the kernel-accel value with the multiprocessor count of the compute device. Will be used later. 2021-04-13 11:26:17 +02:00
Jens Steube
a0eae9050c OpenCL Runtime: Workaround JiT compiler deadlock on NVIDIA driver >= 465.89 2021-04-11 13:35:40 +02:00
Jens Steube
3c199bfa1b
Merge pull request #2693 from matrix/out_of_host_memory 
[backend.c] skipping devices instead of stop with error
 2021-04-05 12:58:47 +02:00
Jens Steube
d53913f444
Merge pull request #2672 from matrix/example2info 
Replaced --example-hashes with --hash-info
 2021-04-03 13:25:51 +02:00
Chick3nman
9b6235a5fc
Downgrade Kernel Exec Timeout Warning 
Kernel times of >450ms are very uncommon and this warning is not a blocking, downgrading it to advice to allow for it to be hidden along with other advice messages.
 2021-04-01 20:01:44 -05:00
Gabriele Gristina
fda0d668e5 use skip also with first checks of backend_session_begin() 2021-01-23 18:51:25 +01:00