mirror of
https://github.com/hashcat/hashcat.git
synced 2024-12-24 07:28:10 +00:00
parent
556e64ffe6
commit
f75f612026
109
OpenCL/m15500.cl
109
OpenCL/m15500.cl
@ -1,109 +0,0 @@
|
||||
/**
|
||||
* Author......: See docs/credits.txt
|
||||
* License.....: MIT
|
||||
*/
|
||||
|
||||
#define NEW_SIMD_CODE
|
||||
|
||||
#include "inc_vendor.cl"
|
||||
#include "inc_hash_constants.h"
|
||||
#include "inc_hash_functions.cl"
|
||||
#include "inc_types.cl"
|
||||
#include "inc_common.cl"
|
||||
#include "inc_simd.cl"
|
||||
#include "inc_rp.h"
|
||||
#include "inc_rp.cl"
|
||||
|
||||
#define COMPARE_S "inc_comp_single.cl"
|
||||
#define COMPARE_M "inc_comp_multi.cl"
|
||||
|
||||
|
||||
__kernel void m15500_init (__global pw_t *pws, __global const kernel_rule_t *rules_buf, __global const comb_t *combs_buf, __global const bf_t *bfs_buf, __global argon2_tmp_t *tmps, __global void *hooks, __global const u32 *bitmaps_buf_s1_a, __global const u32 *bitmaps_buf_s1_b, __global const u32 *bitmaps_buf_s1_c, __global const u32 *bitmaps_buf_s1_d, __global const u32 *bitmaps_buf_s2_a, __global const u32 *bitmaps_buf_s2_b, __global const u32 *bitmaps_buf_s2_c, __global const u32 *bitmaps_buf_s2_d, __global plain_t *plains_buf, __global const digest_t *digests_buf, __global u32 *hashes_shown, __global const salt_t *salt_bufs, __global argon2_t *esalt_bufs, __global u32 *d_return_buf, __global u32 *d_scryptV0_buf, __global u32 *d_scryptV1_buf, __global u32 *d_scryptV2_buf, __global u32 *d_scryptV3_buf, const u32 bitmap_mask, const u32 bitmap_shift1, const u32 bitmap_shift2, const u32 salt_pos, const u32 loop_pos, const u32 loop_cnt, const u32 rules_cnt, const u32 digests_cnt, const u32 digests_offset, const u32 combs_mode, const u32 gid_max)
|
||||
{
|
||||
/**
|
||||
* base
|
||||
*/
|
||||
|
||||
const u32 gid = get_global_id (0);
|
||||
|
||||
if (gid >= gid_max) return;
|
||||
|
||||
u32 w0[4];
|
||||
|
||||
w0[0] = pws[gid].i[0];
|
||||
w0[1] = pws[gid].i[1];
|
||||
w0[2] = pws[gid].i[2];
|
||||
w0[3] = pws[gid].i[3];
|
||||
|
||||
u32 w1[4];
|
||||
|
||||
w1[0] = pws[gid].i[4];
|
||||
w1[1] = pws[gid].i[5];
|
||||
w1[2] = pws[gid].i[6];
|
||||
w1[3] = pws[gid].i[7];
|
||||
|
||||
u32 w2[4];
|
||||
|
||||
w2[0] = 0;
|
||||
w2[1] = 0;
|
||||
w2[2] = 0;
|
||||
w2[3] = 0;
|
||||
|
||||
u32 w3[4];
|
||||
|
||||
w3[0] = 0;
|
||||
w3[1] = 0;
|
||||
w3[2] = 0;
|
||||
w3[3] = 0;
|
||||
|
||||
const u32 pw_len = pws[gid].pw_len;
|
||||
|
||||
/**
|
||||
* main
|
||||
*/
|
||||
|
||||
|
||||
|
||||
/**
|
||||
* Store tmps
|
||||
*/
|
||||
|
||||
printf("I was in m15500_init()\n");
|
||||
}
|
||||
|
||||
__kernel void m15500_loop (__global pw_t *pws, __global const kernel_rule_t *rules_buf, __global const comb_t *combs_buf, __global const bf_t *bfs_buf, __global argon2_tmp_t *tmps, __global void *hooks, __global const u32 *bitmaps_buf_s1_a, __global const u32 *bitmaps_buf_s1_b, __global const u32 *bitmaps_buf_s1_c, __global const u32 *bitmaps_buf_s1_d, __global const u32 *bitmaps_buf_s2_a, __global const u32 *bitmaps_buf_s2_b, __global const u32 *bitmaps_buf_s2_c, __global const u32 *bitmaps_buf_s2_d, __global plain_t *plains_buf, __global const digest_t *digests_buf, __global u32 *hashes_shown, __global const salt_t *salt_bufs, __global argon2_t *esalt_bufs, __global u32 *d_return_buf, __global u32 *d_scryptV0_buf, __global u32 *d_scryptV1_buf, __global u32 *d_scryptV2_buf, __global u32 *d_scryptV3_buf, const u32 bitmap_mask, const u32 bitmap_shift1, const u32 bitmap_shift2, const u32 salt_pos, const u32 loop_pos, const u32 loop_cnt, const u32 rules_cnt, const u32 digests_cnt, const u32 digests_offset, const u32 combs_mode, const u32 gid_max)
|
||||
{
|
||||
/**
|
||||
* base
|
||||
*/
|
||||
|
||||
const u32 gid = get_global_id (0);
|
||||
|
||||
if ((gid * VECT_SIZE) >= gid_max) return;
|
||||
|
||||
|
||||
/**
|
||||
* iter1
|
||||
*/
|
||||
|
||||
|
||||
/**
|
||||
* Update tmps
|
||||
*/
|
||||
|
||||
printf("I was in m15500_loop()\n");
|
||||
}
|
||||
}
|
||||
|
||||
__kernel void m15500_comp (__global pw_t *pws, __global const kernel_rule_t *rules_buf, __global const comb_t *combs_buf, __global const bf_t *bfs_buf, __global argon2_tmp_t *tmps, __global void *hooks, __global const u32 *bitmaps_buf_s1_a, __global const u32 *bitmaps_buf_s1_b, __global const u32 *bitmaps_buf_s1_c, __global const u32 *bitmaps_buf_s1_d, __global const u32 *bitmaps_buf_s2_a, __global const u32 *bitmaps_buf_s2_b, __global const u32 *bitmaps_buf_s2_c, __global const u32 *bitmaps_buf_s2_d, __global plain_t *plains_buf, __global const digest_t *digests_buf, __global u32 *hashes_shown, __global const salt_t *salt_bufs, __global argon2_t *esalt_bufs, __global u32 *d_return_buf, __global u32 *d_scryptV0_buf, __global u32 *d_scryptV1_buf, __global u32 *d_scryptV2_buf, __global u32 *d_scryptV3_buf, const u32 bitmap_mask, const u32 bitmap_shift1, const u32 bitmap_shift2, const u32 salt_pos, const u32 loop_pos, const u32 loop_cnt, const u32 rules_cnt, const u32 digests_cnt, const u32 digests_offset, const u32 combs_mode, const u32 gid_max)
|
||||
{
|
||||
const u32 gid = get_global_id (0);
|
||||
const u32 lid = get_local_id (0);
|
||||
const u32 lsz = get_local_size (0);
|
||||
|
||||
/**
|
||||
* Mark hashes
|
||||
*/
|
||||
|
||||
printf("I was in m15500_comp()\n");
|
||||
}
|
Loading…
Reference in New Issue
Block a user