Commit Graph

  • 46f737c5af Use real constant memory on CUDA Jens Steube 2019-05-10 13:22:26 +02:00
  • 371991e079 included speed improvements and feedback from atom applied to all 172xx kernel variants Sein Coray 2019-05-10 12:50:03 +02:00
  • 5d14a59304 Need 3.x nvrtc minimum Jens Steube 2019-05-10 10:11:12 +02:00
  • 556db9a9e9 Fix: buffer sizes (again) vlo 2019-05-10 08:43:25 +02:00
  • 54feb62e94 brute-force nvrtc .dll name Jens Steube 2019-05-09 22:17:13 +02:00
  • a2b5981303 Fix some library names Jens Steube 2019-05-09 21:20:50 +02:00
  • ee500bec02 Fixed: buffer overflows vlo 2019-05-09 19:27:11 +02:00
  • e4d8e4a7ad bring fork up-to-date Sein Coray 2019-05-09 17:23:59 +02:00
  • ce20a5ab6b Fix uint4 rotate in scrypt based kernels for CUDA Jens Steube 2019-05-09 16:55:48 +02:00
  • be8f29ca39 Only warn about broken NVIDIA driver Jens Steube 2019-05-09 16:30:08 +02:00
  • 39e150fc1e Use xxx_v2 CUDA symbols Jens Steube 2019-05-09 14:37:14 +02:00
  • 82927c13c8 Get rid of uchar4 in -m 9100 Jens Steube 2019-05-09 13:09:27 +02:00
  • ec4d4218c0 Add some missing operators for vector types Jens Steube 2019-05-09 12:59:36 +02:00
  • 6db4ab7e60 Fix scrypt based algorithms to work on CUDA Jens Steube 2019-05-09 11:11:52 +02:00
  • 23c3e723f7 Implemeted table for alternate base64 vlo 2019-05-09 08:51:21 +02:00
  • 33028314f0 Add hc_cuCtxSetCacheConfig() Jens Steube 2019-05-09 00:04:05 +02:00
  • fb82bfc169 Improve thread handling based on FIXED_LOCAL_SIZE Jens Steube 2019-05-08 23:30:07 +02:00
  • 3a3df091c7 Fix CUDA num_elements Jens Steube 2019-05-08 22:42:52 +02:00
  • 027af75a39 Fix rotate function names Jens Steube 2019-05-08 20:42:46 +02:00
  • 6b7d064118 Replace (u32x) (...) with make_u32x (...) Jens Steube 2019-05-08 15:21:22 +02:00
  • 14dd1aaaeb Added support for passlib pbkdf2-sha256 and sha1 vlo 2019-05-08 09:24:12 +02:00
  • a01cc28ec6 Merge pull request #2014 from hashcat/master Jens Steube 2019-05-07 21:48:47 +02:00
  • b6cc3c7d55 Merge pull request #2013 from philsmd/master Jens Steube 2019-05-07 21:48:16 +02:00
  • 363e789b89 Assume local nvrtc.dll and cuda.dll on windows Jens Steube 2019-05-07 16:52:08 +02:00
  • 54dd2ea300 Use same settings for vector datatypes in inc_types.h as seen in cuda SDK vector_types.h Jens Steube 2019-05-07 16:07:28 +02:00
  • f45a726376 fixes #2012: add Deflate support for 7-Zip using zlib philsmd 2019-05-07 15:31:11 +02:00
  • 0c8768c520 add the zlib 1.2.11 source code to the dependencies philsmd 2019-05-07 15:18:49 +02:00
  • a7d04adba3 Fix opencl_devices_active and backend_devices_active Jens Steube 2019-05-07 14:17:29 +02:00
  • 7e5356126c Fix more use of LOCAL_VK and LOCAL_AS Jens Steube 2019-05-07 12:22:37 +02:00
  • 03b2d3fb69 Fix use of LOCAL_VK and LOCAL_AS in -m 3200 Jens Steube 2019-05-07 12:08:54 +02:00
  • 8ff8c5d536 Add LOCAL_VK to make use of __shared__ Jens Steube 2019-05-07 09:01:32 +02:00
  • bbed0cd67a Fix test.sh and bitsliced algos Jens Steube 2019-05-06 15:06:02 +02:00
  • d0bd33c9d1 Rename CONSTANT_AS to CONSTANT_VK Jens Steube 2019-05-06 14:34:16 +02:00
  • 64c495dfa5 Use CUDA stream for all cuLaunchKernel() invocations Jens Steube 2019-05-06 11:23:34 +02:00
  • d94f582097 Replace CEILDIV() with round_up_multiple_64() Jens Steube 2019-05-06 09:36:07 +02:00
  • e9c04c2446 More CUDA implementation Jens Steube 2019-05-05 21:15:46 +02:00
  • 08dc1acc02 More CUDA rewrites Jens Steube 2019-05-05 11:57:54 +02:00
  • ec9925f3b1 Warnings self-check and autotune with CUDA Jens Steube 2019-05-04 21:52:00 +02:00
  • 4df00033d7 Prepare CUDA events Jens Steube 2019-05-04 10:44:03 +02:00
  • f2948460c9 Some first kernel invocations Jens Steube 2019-05-04 10:13:43 +02:00
  • 5ee033673c Disable name mangling in NVRTC's PTX output and more Jens Steube 2019-05-03 15:50:07 +02:00
  • 503304f36a Add some first CUDA device memory allocations and host buffer copies Jens Steube 2019-05-03 12:07:06 +02:00
  • 003e23bae8 Change hash mode of addition, add test module vlo 2019-05-02 22:36:23 +02:00
  • 50a6e720ca More OpenCL variables rename Jens Steube 2019-05-02 17:30:46 +02:00
  • af8e317cf4 Begin renaming some OpenCL only variables Jens Steube 2019-05-02 17:12:59 +02:00
  • a6fa7a2749 Add support for some first CUDA module loader Jens Steube 2019-05-02 14:58:52 +02:00
  • 456c57a6d0 Set vector width size for CUDA Jens Steube 2019-05-01 18:20:19 +02:00
  • 4510504257 New module for Python passlib pbkdf2-sha512 vlo 2019-05-01 18:02:38 +02:00
  • 3c4f4df771 Rename some more variables Jens Steube 2019-05-01 15:52:56 +02:00
  • 495d89f831 Find alias devices across different backend API's Jens Steube 2019-05-01 07:27:10 +02:00
  • 6fd936b43a Removed --opencl-platforms filter in order to force backend device numbers to stay constant Jens Steube 2019-04-30 16:24:13 +02:00
  • e3500ff4aa Add CUDA device attributes to -I Jens Steube 2019-04-30 13:38:44 +02:00
  • c80e516c3c Merge pull request #2007 from hashcat/master Jens Steube 2019-04-29 12:27:27 +02:00
  • 6caa78695f Backport -m 19500 fix to -m 19300 Jens Steube 2019-04-29 12:26:22 +02:00
  • 922fa2e351 Use md5 for generatic unique salt for salt sorter in -m 19500 Jens Steube 2019-04-29 12:01:53 +02:00
  • d862458ab5 Begin renaming API specific variables in backend section Jens Steube 2019-04-29 10:21:59 +02:00
  • d73c0ac8a9 More CUDA attribute queries Jens Steube 2019-04-28 18:54:26 +02:00
  • a415422123 Initialize CUDA devices and some first attribute queries Jens Steube 2019-04-28 14:45:50 +02:00
  • 222be0b0dc Merge pull request #2005 from hashcat/master Jens Steube 2019-04-27 16:35:38 +02:00
  • 200e72dba3 Limit -T maximum on -m 3200 to what's possible based on device specific shared memory available Jens Steube 2019-04-27 16:15:18 +02:00
  • d67de66453 Disable kernel cache on -m 3200 Jens Steube 2019-04-27 16:00:29 +02:00
  • 58213c81d6 Add vector datatypes operators Jens Steube 2019-04-26 22:07:56 +02:00
  • 052e42ccef Fix CUDA_ARCH value Jens Steube 2019-04-26 15:14:48 +02:00
  • 06171958ee Add --gpu-architecture to NVRTC build options Jens Steube 2019-04-26 15:10:02 +02:00
  • 6a32e8ef18 Fix ulong datatype on Windows x64 Jens Steube 2019-04-26 14:11:13 +02:00
  • d9cb5cf8df Fix recursion in inc_common.cl Jens Steube 2019-04-26 14:03:57 +02:00
  • 3b7304c9d8 Fix recursion in inc_platform.cl Jens Steube 2019-04-26 14:01:14 +02:00
  • 89119bf24a Add missing inc_platform.h include Jens Steube 2019-04-26 13:59:43 +02:00
  • 00e1e32492 Replace barrier() with SYNC_THREADS() Jens Steube 2019-04-26 13:34:07 +02:00
  • 9faba41848 Use nvrtc to compile PTX (resulting PTX not yet used) Jens Steube 2019-04-26 13:28:44 +02:00
  • 4045e60021 Add nvrtc wrapper for later use Jens Steube 2019-04-26 10:03:16 +02:00
  • 4b986de5fb Prepare native CUDA hybrid integration Jens Steube 2019-04-25 14:45:17 +02:00
  • c02083281f Fix undefined-internal warning message on ROCM Jens Steube 2019-04-24 14:17:34 +02:00
  • c5c79feaaa More cam_feistel() optimization Jens Steube 2019-04-23 21:56:40 +02:00
  • f49d3f92e9 Reduce cam_feistel() xor count Jens Steube 2019-04-23 16:51:16 +02:00
  • f10d27b2c9 Get rid of extract_byte() in inc_cipher_twofish.cl Jens Steube 2019-04-23 15:16:42 +02:00
  • bf4b1a8e02 Remove duplicate 'static' declaration specifier in -m 19800 and -m 19900 Jens Steube 2019-04-22 17:55:00 +02:00
  • b9aaaf7809 Move 198xx DiskCryptor to 200xx to not collide with Kerberos 5, etype 17, Pre-Auth jsteube 2019-04-20 19:41:37 +02:00
  • 17ab30b29f Merge pull request #1991 from brandoncasaba/master Jens Steube 2019-04-20 19:10:33 +02:00
  • 8b51843bb2 Mark -m 15300 and -m 1590 as unstable on AMDGPU driver jsteube 2019-04-20 18:58:01 +02:00
  • 926e99811c Add some more NO_UNROLL to avoid module_unstable_warnings jsteube 2019-04-20 16:36:43 +02:00
  • 08a74cfcb5 Add NO_UNROLL to -m 1750 for AMDGPU driver jsteube 2019-04-20 11:47:41 +02:00
  • 70fc36bf01 Reorganize inc_common.cl and make better use of HAS_* macros jsteube 2019-04-20 11:25:34 +02:00
  • ac4f8e688a Dependencies: Updated OpenCL-Headers to latest version from GitHub master repository Dependencies: Updated OpenCL function wrapper declarations accordingly to updated OpenCL-Headers jsteube 2019-04-20 09:34:13 +02:00
  • 5f3d9e08b9 Kernel Compile: Removed -cl-std= from all kernel build options since we're compatible to all OpenCL versions jsteube 2019-04-20 08:46:25 +02:00
  • f424dd1edb Fix tools/install_modules.sh ro uninstall pycryptodome jsteube 2019-04-19 13:17:15 +02:00
  • 4548d15dde Fix spin_damp datatype jsteube 2019-04-19 12:26:37 +02:00
  • 75b92c1ab1 Use both LE and BE modes for nonce error correction if none or both modes are set jsteube 2019-04-18 22:50:53 +02:00
  • 5b97fe7514 Workaround volatile for -m 2500 and -m 2501 for macosx jsteube 2019-04-18 16:52:14 +02:00
  • 6d21c1633b Fix loop_step value in selftest.c, this also workarounds Trap 6 errors jsteube 2019-04-18 14:29:39 +02:00
  • e7ae8e6863 Requirements: Update runtime check for minimum NVIDIA driver version from 367.x to 418.56 or later Jens Steube 2019-04-18 12:21:12 +02:00
  • 38c1029f2e Need volatile for IRIS GPU on Mac OSX for -m 2500 and -m 2501 Jens Steube 2019-04-17 13:21:35 +02:00
  • b2fecc5828 Fix typo in m11800_a0-pure.cl jsteube 2019-04-17 12:28:01 +02:00
  • 0a4ce19915 Merge branch 'master' of https://github.com/hashcat/hashcat Brandon Chalk 2019-04-16 10:03:05 -07:00
  • 5adb941f9a Test module variable scope fix Brandon Chalk 2019-04-16 09:58:28 -07:00
  • 66aada0f6a Merge branch 'master' of https://github.com/hashcat/hashcat jsteube 2019-04-15 18:11:20 +02:00
  • 74e3ede391 Test optimization for kernel include in pure mode jsteube 2019-04-15 18:11:15 +02:00
  • 56ae7fa2b8 Remove warnings no longer relevant in -m 15600 and -m 16300 jsteube 2019-04-15 16:05:26 +02:00
  • a3b6e6f1b0 Fix hc_rotl64_S() for AMD devices jsteube 2019-04-15 12:11:37 +02:00
  • e0d0721109 Workaround for Page not present or supervisor privilege JiT fail on ROCM for -m 13400 jsteube 2019-04-15 11:19:02 +02:00