diff --git a/Makefile b/Makefile index 69cf68b..90d5f0f 100644 --- a/Makefile +++ b/Makefile @@ -1,9 +1,16 @@ LIBS=-lpcre -lcrypto -lm -lpthread CFLAGS=-ggdb -O3 -Wall -OBJS=vanitygen.o +OBJS=vanitygen.o oclvanitygen.o pattern.o +PROGS=vanitygen +TESTS= -vanitygen: $(OBJS) - $(CC) $(OBJS) -o $@ $(CFLAGS) $(LIBS) +all: $(PROGS) + +vanitygen: vanitygen.o pattern.o + $(CC) $^ -o $@ $(CFLAGS) $(LIBS) + +oclvanitygen: oclvanitygen.o pattern.o + $(CC) $^ -o $@ $(CFLAGS) $(LIBS) -lOpenCL clean: - rm -f $(OBJS) vanitygen + rm -f $(OBJS) $(PROGS) $(TESTS) diff --git a/Makefile.Win32 b/Makefile.Win32 index c9107e0..a8c2cdb 100644 --- a/Makefile.Win32 +++ b/Makefile.Win32 @@ -4,7 +4,7 @@ PTHREADS_DIR = C:\pthreads-w32 PCRE_DIR = C:\pcre-7.9-src CFLAGS = /D_WIN32 /DPTW32_STATIC_LIB /DPCRE_STATIC /I$(OPENSSL_DIR)\include /I$(PTHREADS_DIR) /I$(PCRE_DIR) LIBS = $(OPENSSL_DIR)\lib\libeay32.lib $(PTHREADS_DIR)\pthread.lib $(PCRE_DIR)\pcre.lib ws2_32.lib -OBJS = vanitygen.obj +OBJS = vanitygen.obj pattern.obj winglue.obj all: vanitygen.exe @@ -15,5 +15,4 @@ vanitygen.exe: $(OBJS) $(CC) $(CFLAGS) /c /Tp$< /Fo$@ clean: - del vanitygen.exe - del vanitygen.obj + del vanitygen.exe $(OBJS) diff --git a/calc_addrs.cl b/calc_addrs.cl new file mode 100644 index 0000000..f6c2e0f --- /dev/null +++ b/calc_addrs.cl @@ -0,0 +1,860 @@ +/* + * Vanitygen, vanity bitcoin address generator + * Copyright (C) 2011 + * + * Vanitygen is free software: you can redistribute it and/or modify + * it under the terms of the GNU Affero General Public License as published by + * the Free Software Foundation, either version 3 of the License, or + * any later version. + * + * Vanitygen is distributed in the hope that it will be useful, + * but WITHOUT ANY WARRANTY; without even the implied warranty of + * MERCHANTABILITY or FITNESS FOR A PARTICULAR PURPOSE. See the + * GNU Affero General Public License for more details. + * + * You should have received a copy of the GNU Affero General Public License + * along with Vanitygen. If not, see . + */ + +/* + * This file contains an OpenCL kernel for performing certain parts of + * the bitcoin address calculation process. + * + * Kernel: calc_addrs + * + * Inputs: + * - Row of (sequential) EC points + * - Array of column increment EC points (= rowsize * Pgenerator) + * + * Steps: + * - For each row increment value C: + * - For each row point P: + * - Compute P + C + * - Normalize and hash with SHA256 and RIPEMD160 + * - Store hash value in output array + * + * Output: + * - Array of 20-byte address hash values + * + * Each instance of the kernel computes one full row. With a typical + * row size of 256 points, this makes each kernel instance very heavy. + * This tradeoff is chosen in favor of batched modular inversion, which + * substantially reduces the cost of performing modular inversion. + */ + +/* + * BIGNUM mini-library + * This module deals with fixed-size 256-bit bignums. + * Where modular arithmetic is performed, the SECP256k1 prime + * modulus (below) is assumed. + * + * Methods include: + * - bn_is_zero/bn_is_one/bn_is_odd/bn_is_even/bn_is_bit_set + * - bn_rshift[1]/bn_lshift[1] + * - bn_neg + * - bn_uadd/bn_uadd_p + * - bn_usub/bn_usub_p + */ + +typedef uint bn_word; +#define BN_NBITS 256 +#define BN_WSHIFT 5 +#define BN_WBITS (1 << BN_WSHIFT) +#define BN_NWORDS ((BN_NBITS/8) / sizeof(bn_word)) +#define BN_WORDMAX 0xffffffff + +#define MODULUS_BYTES \ + 0xfffffc2f, 0xfffffffe, 0xffffffff, 0xffffffff, \ + 0xffffffff, 0xffffffff, 0xffffffff, 0xffffffff + +typedef struct { + bn_word d[BN_NWORDS]; +} bignum; + +__constant bn_word modulus[] = { MODULUS_BYTES }; +__constant bn_word bn_one[BN_NWORDS] = { 1, 0, }; +__constant bignum bn_zero; + +__constant bn_word mont_rr[BN_NWORDS] = { 0xe90a1, 0x7a2, 0x1, 0, }; +__constant bn_word mont_n0[2] = { 0xd2253531, 0xd838091d }; + + +#define bn_is_odd(bn) (bn.d[0] & 1) +#define bn_is_even(bn) (!bn_is_odd(bn)) +#define bn_is_zero(bn) (!bn.d[0] && !bn.d[1] && !bn.d[2] && \ + !bn.d[3] && !bn.d[4] && !bn.d[5] && \ + !bn.d[6] && !bn.d[7]) +#define bn_is_one(bn) ((bn.d[0] == 1) && !bn.d[1] && !bn.d[2] && \ + !bn.d[3] && !bn.d[4] && !bn.d[5] && \ + !bn.d[6] && !bn.d[7]) +#define bn_is_bit_set(bn, n) \ + ((((bn_word*)&bn)[n >> BN_WSHIFT]) & (1 << (n & (BN_WBITS-1)))) + + +/* + * Bitwise shift + */ + +void +bn_lshift1(bignum *bn) +{ + int i; + for (i = (BN_NWORDS - 1); i > 0; i--) + bn->d[i] = (bn->d[i] << 1) | (bn->d[i-1] >> 31); + bn->d[i] <<= 1; +} + +void +bn_rshift(bignum *bn, int shift) +{ + int i, wd, iws; + bn_word *op, *ip, ihw, ilw; + iws = (shift & (BN_WBITS-1)); + wd = (shift >> BN_WSHIFT); + ip = ((bn_word*)bn); + op = ip + wd; + wd = BN_NWORDS - wd; + ihw = ip[0]; + for (i = 1; i < wd; i++) { + ilw = ihw; + ihw = ip[i]; + op[i-1] = ((ilw >> iws) | (ihw << (BN_WBITS - iws))); + } + op[i-1] = (ihw >> iws); + if (i < BN_NWORDS) { + while (i < BN_NWORDS) + op[i++] = 0; + } +} + +void +bn_rshift1(bignum *bn) +{ + int i; + for (i = 0; i < (BN_NWORDS - 1); i++) + bn->d[i] = (bn->d[i+1] << 31) | (bn->d[i] >> 1); + bn->d[i] >>= 1; +} + + +/* + * Unsigned comparison + */ + +int +bn_ucmp(bignum *a, bignum *b) +{ + int i; + for (i = (BN_NWORDS - 1); i >= 0; i--) { + if (a->d[i] < b->d[i]) return -1; + if (a->d[i] > b->d[i]) return 1; + } + return 0; +} + +int +bn_ucmp_c(bignum *a, __constant bn_word *b) +{ + int i; + for (i = (BN_NWORDS - 1); i >= 0; i--) { + if (a->d[i] < b[i]) return -1; + if (a->d[i] > b[i]) return 1; + } + return 0; +} + +/* + * Negate + */ + +void +bn_neg(bignum *n) +{ + int i, c; + for (i = 0, c = 1; i < BN_NWORDS; i++) + if ((n->d[i] = (~n->d[i]) + c) && c) + c = 0; +} + +/* + * Add/subtract + */ + +#define bn_add_word(r, a, b, t, c) do { \ + t = a + b; \ + c = (t < a) ? 1 : 0; \ + r = t; \ + } while (0) + +#define bn_addc_word(r, a, b, t, c) do { \ + t = a + b + c; \ + c = (t < a) ? 1 : ((c && (t == a)) ? 1 : 0); \ + r = t; \ + } while (0) + +bn_word +bn_uadd(bignum *r, bignum *a, bignum *b) +{ + bn_word t, c = 0; + int i; + bn_add_word(r->d[0], a->d[0], b->d[0], t, c); + for (i = 1; i < BN_NWORDS; i++) + bn_addc_word(r->d[i], a->d[i], b->d[i], t, c); + return c; +} + +bn_word +bn_uadd_c(bignum *r, bignum *a, __constant bn_word *b) +{ + bn_word t, c = 0; + int i; + bn_add_word(r->d[0], a->d[0], b[0], t, c); + for (i = 1; i < BN_NWORDS; i++) + bn_addc_word(r->d[i], a->d[i], b[i], t, c); + return c; +} + +#define bn_sub_word(r, a, b, t, c) do { \ + t = a - b; \ + c = (a < b) ? 1 : 0; \ + r = t; \ + } while (0) + +#define bn_subb_word(r, a, b, t, c) do { \ + t = a - (b + c); \ + c = ((a < b) || (!a && c)) ? 1 : 0; \ + r = t; \ + } while (0) + +bn_word +bn_usub(bignum *r, bignum *a, bignum *b) +{ + bn_word t, c = 0; + int i; + bn_sub_word(r->d[0], a->d[0], b->d[0], t, c); + for (i = 1; i < BN_NWORDS; i++) + bn_subb_word(r->d[i], a->d[i], b->d[i], t, c); + return c; +} + +bn_word +bn_usub_c(bignum *r, bignum *a, __constant bn_word *b) +{ + bn_word t, c = 0; + int i; + bn_sub_word(r->d[0], a->d[0], b[0], t, c); + for (i = 1; i < BN_NWORDS; i++) + bn_subb_word(r->d[i], a->d[i], b[i], t, c); + return c; +} + +/* + * Modular add/sub + */ + +void +bn_mod_add(bignum *r, bignum *a, bignum *b) +{ + if (bn_uadd(r, a, b) || + (bn_ucmp_c(r, modulus) >= 0)) + bn_usub_c(r, r, modulus); +} + +void +bn_mod_sub(bignum *r, bignum *a, bignum *b) +{ + if (bn_usub(r, a, b)) + bn_uadd_c(r, r, modulus); +} + +void +bn_mod_lshift1(bignum *bn) +{ + bn_word c = (bn->d[BN_NWORDS-1] & 0x80000000); + bn_lshift1(bn); + if (c || (bn_ucmp_c(bn, modulus) >= 0)) + bn_usub_c(bn, bn, modulus); +} + +/* + * Montgomery multiplication + * + * This includes normal multiplication of two "Montgomeryized" + * bignums, and bn_from_mont for de-Montgomeryizing a bignum. + */ + +#define bn_mul_word(r, a, w, c, p, s) do { \ + p = mul_hi(a, w); \ + r = (a * w) + c; \ + c = (r < c) ? p + 1 : p; \ + } while (0) + +#define bn_mul_add_word(r, a, w, c, p, s) do { \ + p = mul_hi(a, w); \ + s = r + c; \ + r = (a * w) + s; \ + c = (s < c) ? p + 1 : p; \ + if (r < s) c++; \ + } while (0) + +void +bn_mul_mont(bignum *r, bignum *a, bignum *b) +{ + bignum t; + bn_word tea, teb, c, p, s, m; + int i, j; + + c = 0; + for (j = 0; j < BN_NWORDS; j++) + bn_mul_word(t.d[j], a->d[j], b->d[0], c, p, s); + tea = c; + teb = 0; + + c = 0; + m = t.d[0] * mont_n0[0]; + bn_mul_add_word(t.d[0], modulus[0], m, c, p, s); + for (j = 1; j < BN_NWORDS; j++) { + bn_mul_add_word(t.d[j], modulus[j], m, c, p, s); + t.d[j-1] = t.d[j]; + } + t.d[BN_NWORDS-1] = tea + c; + tea = teb + ((t.d[BN_NWORDS-1] < c) ? 1 : 0); + + for (i = 1; i < BN_NWORDS; i++) { + c = 0; + for (j = 0; j < BN_NWORDS; j++) + bn_mul_add_word(t.d[j], a->d[j], b->d[i], c, p, s); + tea += c; + teb = ((tea < c) ? 1 : 0); + + c = 0; + m = t.d[0] * mont_n0[0]; + bn_mul_add_word(t.d[0], modulus[0], m, c, p, s); + for (j = 1; j < BN_NWORDS; j++) { + bn_mul_add_word(t.d[j], modulus[j], m, c, p, s); + t.d[j-1] = t.d[j]; + } + t.d[BN_NWORDS-1] = tea + c; + tea = teb + ((t.d[BN_NWORDS-1] < c) ? 1 : 0); + } + + if (tea || (t.d[BN_NWORDS-1] >= modulus[7])) { + c = bn_usub_c(r, &t, modulus); + if (tea || !c) + return; + } + *r = t; +} + +void +bn_from_mont(bignum *rb, bignum *b) +{ +#define WORKSIZE ((2*BN_NWORDS) + 1) + bn_word r[WORKSIZE]; + bn_word m, c, p, s; + int i, j, top, tl; + /* Copy the input to the working area */ + for (i = 0; i < BN_NWORDS; i++) + r[i] = b->d[i]; + /* Zero the upper words */ + for (i = BN_NWORDS; i < WORKSIZE; i++) + r[i] = 0; + /* Multiply (long) by modulus */ + for (i = 0; i < BN_NWORDS; i++) { + m = r[i] * mont_n0[0]; + c = 0; + for (j = 0; j < BN_NWORDS; j++) + bn_mul_add_word(r[i+j], modulus[j], m, c, p, s); + r[BN_NWORDS + i] += c; + if (r[BN_NWORDS + i] < c) { + if (++r[BN_NWORDS + i + 1] == 0) + ++r[BN_NWORDS + i + 2]; /* The end..? */ + } + } + for (top = WORKSIZE - 1; (top > BN_NWORDS) && (r[top] == 0); top--); + if (top <= BN_NWORDS) { + *rb = bn_zero; + return; + } + tl = top - BN_NWORDS; + c = 0; + for (j = 0; j < BN_NWORDS; j++) + bn_subb_word(rb->d[j], r[BN_NWORDS + j], modulus[j], p, c); + if (c) { + for (j = 0; j < BN_NWORDS; j++) + rb->d[j] = r[BN_NWORDS + j]; + } +} + + +/* Montgomery multiplication test kernel */ +__kernel void +test_mul_mont(__global bignum *products_out, __global bignum *nums_in, + int count) +{ + bignum x, y, tmp; + int i, o, p; + o = get_global_id(0) * count; + p = o * 2; + for (i = 0; i < count; i++) { + x = nums_in[p++]; + y = nums_in[p++]; + bn_mul_mont(&tmp, &x, &y); + bn_mul_mont(&tmp, &tmp, &x); + bn_mul_mont(&tmp, &tmp, &y); + bn_from_mont(&x, &tmp); + products_out[o++] = x; + } +} + +/* + * Modular inversion + */ + +void +bn_mod_inverse(bignum *r, bignum *n) +{ + bignum a, b, x, y; + int shift; + bn_word xc, yc; + for (shift = 0; shift < BN_NWORDS; shift++) { + a.d[shift] = modulus[shift]; + x.d[shift] = 0; + y.d[shift] = 0; + } + b = *n; + x.d[0] = 1; + xc = 0; + yc = 0; + while (!bn_is_zero(b)) { + shift = 0; + while (!bn_is_bit_set(b, shift)) { + shift++; + if (bn_is_odd(x)) + xc += bn_uadd_c(&x, &x, modulus); + bn_rshift1(&x); + x.d[7] |= (xc << 31); + xc >>= 1; + } + if (shift) + bn_rshift(&b, shift); + + shift = 0; + while (!bn_is_bit_set(a, shift)) { + shift++; + if (bn_is_odd(y)) + yc += bn_uadd_c(&y, &y, modulus); + bn_rshift1(&y); + y.d[7] |= (yc << 31); + yc >>= 1; + } + if (shift) + bn_rshift(&a, shift); + + if (bn_ucmp(&b, &a) >= 0) { + xc += yc + bn_uadd(&x, &x, &y); + bn_usub(&b, &b, &a); + } else { + yc += xc + bn_uadd(&y, &y, &x); + bn_usub(&a, &a, &b); + } + } + + if (!bn_is_one(a)) { + /* no modular inverse */ + *r = bn_zero; + return; + } + /* Compute y % m as cheaply as possible */ + while (yc < 0x80000000) + yc -= bn_usub_c(&y, &y, modulus); + bn_neg(&y); + *r = y; + return; +} + +/* modular inversion test kernel */ +__kernel void +test_mod_inverse(__global bignum *inv_out, __global bignum *nums_in, + int count) +{ + bignum x, xp; + int i, o; + o = get_global_id(0) * count; + for (i = 0; i < count; i++) { + x = nums_in[o]; + bn_mod_inverse(&xp, &x); + inv_out[o++] = xp; + } +} + + +/* + * HASH FUNCTIONS + * + * BYTE ORDER NOTE: None of the hash functions below deal with byte + * order. The caller is expected to be aware of this when it stuffs + * data into in the native integer. + * + * NOTE #2: Endianness of the OpenCL device makes no difference here. + */ + +/* + * SHA-2 256 + * + * CAUTION: Input buffer will be overwritten/mangled. + * Data expected in big-endian format. + * This implementation is designed for space efficiency more than + * raw speed. + */ + +__constant uint sha2_init[8] = { + 0x6a09e667, 0xbb67ae85, 0x3c6ef372, 0xa54ff53a, + 0x510e527f, 0x9b05688c, 0x1f83d9ab, 0x5be0cd19 +}; + +__constant uint sha2_k[64] = { + 0x428a2f98, 0x71374491, 0xb5c0fbcf, 0xe9b5dba5, + 0x3956c25b, 0x59f111f1, 0x923f82a4, 0xab1c5ed5, + 0xd807aa98, 0x12835b01, 0x243185be, 0x550c7dc3, + 0x72be5d74, 0x80deb1fe, 0x9bdc06a7, 0xc19bf174, + 0xe49b69c1, 0xefbe4786, 0x0fc19dc6, 0x240ca1cc, + 0x2de92c6f, 0x4a7484aa, 0x5cb0a9dc, 0x76f988da, + 0x983e5152, 0xa831c66d, 0xb00327c8, 0xbf597fc7, + 0xc6e00bf3, 0xd5a79147, 0x06ca6351, 0x14292967, + 0x27b70a85, 0x2e1b2138, 0x4d2c6dfc, 0x53380d13, + 0x650a7354, 0x766a0abb, 0x81c2c92e, 0x92722c85, + 0xa2bfe8a1, 0xa81a664b, 0xc24b8b70, 0xc76c51a3, + 0xd192e819, 0xd6990624, 0xf40e3585, 0x106aa070, + 0x19a4c116, 0x1e376c08, 0x2748774c, 0x34b0bcb5, + 0x391c0cb3, 0x4ed8aa4a, 0x5b9cca4f, 0x682e6ff3, + 0x748f82ee, 0x78a5636f, 0x84c87814, 0x8cc70208, + 0x90befffa, 0xa4506ceb, 0xbef9a3f7, 0xc67178f2 +}; + +void +sha2_256_init(uint *out) +{ + int i; + for (i = 0; i < 8; i++) + out[i] = sha2_init[i]; +} + +/* The state variable remapping is really contorted */ +#define sha2_stvar(vals, i, v) vals[(i+(7-v)) % 8] + +void +sha2_256_block(uint *out, uint *in) +{ + int i; + uint state[8], s0, s1, t1, t2; + for (i = 0; i < 8; i++) + state[7-i] = out[i]; + for (i = 0; i < 64; i++) { + if (i >= 16) { + /* Advance the input window */ + t1 = in[(i + 1) % 16]; + t2 = in[(i + 14) % 16]; + in[i % 16] += (in[(i + 9) % 16] + + (rotate(t1, 25U) ^ rotate(t1, 14U) ^ (t1 >> 3)) + + (rotate(t2, 15U) ^ rotate(t2, 13U) ^ (t2 >> 10))); + } + + /* Compute the t1, t2 augmentations */ + t1 = sha2_stvar(state, i, 4); + t2 = sha2_stvar(state, i, 0); + s0 = (rotate(t2, 30U) ^ rotate(t2, 19U) ^ rotate(t2, 10U)); + s1 = (rotate(t1, 26U) ^ rotate(t1, 21U) ^ rotate(t1, 7U)); + + t1 = (sha2_stvar(state, i, 7) + s1 + sha2_k[i] + in[i % 16] + + ((t1 & sha2_stvar(state, i, 5)) ^ + (~t1 & sha2_stvar(state, i, 6)))); + t2 = s0 + ((t2 & sha2_stvar(state, i, 1)) ^ + (t2 & sha2_stvar(state, i, 2)) ^ + (sha2_stvar(state, i, 1) & sha2_stvar(state, i, 2))); + + sha2_stvar(state, i, 3) += t1; + sha2_stvar(state, i, 7) = t1 + t2; + } + for (i = 0; i < 8; i++) + out[i] += state[7-i]; +} + + +/* + * RIPEMD160 + * + * Data expected in little-endian format. + */ + +__constant uint ripemd160_iv[] = { + 0x67452301, 0xEFCDAB89, 0x98BADCFE, 0x10325476, 0xC3D2E1F0 }; +__constant uint ripemd160_k[] = { + 0x00000000, 0x5A827999, 0x6ED9EBA1, 0x8F1BBCDC, 0xA953FD4E }; +__constant uint ripemd160_kp[] = { + 0x50A28BE6, 0x5C4DD124, 0x6D703EF3, 0x7A6D76E9, 0x00000000 }; +__constant uchar ripemd160_ws[] = { + 0, 1, 2, 3, 4, 5, 6, 7, 8, 9, 10, 11, 12, 13, 14, 15, + 7, 4, 13, 1, 10, 6, 15, 3, 12, 0, 9, 5, 2, 14, 11, 8, + 3, 10, 14, 4, 9, 15, 8, 1, 2, 7, 0, 6, 13, 11, 5, 12, + 1, 9, 11, 10, 0, 8, 12, 4, 13, 3, 7, 15, 14, 5, 6, 2, + 4, 0, 5, 9, 7, 12, 2, 10, 14, 1, 3, 8, 11, 6, 15, 13, +}; +__constant uchar ripemd160_wsp[] = { + 5, 14, 7, 0, 9, 2, 11, 4, 13, 6, 15, 8, 1, 10, 3, 12, + 6, 11, 3, 7, 0, 13, 5, 10, 14, 15, 8, 12, 4, 9, 1, 2, + 15, 5, 1, 3, 7, 14, 6, 9, 11, 8, 12, 2, 10, 0, 4, 13, + 8, 6, 4, 1, 3, 11, 15, 0, 5, 12, 2, 13, 9, 7, 10, 14, + 12, 15, 10, 4, 1, 5, 8, 7, 6, 2, 13, 14, 0, 3, 9, 11 +}; +__constant uchar ripemd160_rl[] = { + 11, 14, 15, 12, 5, 8, 7, 9, 11, 13, 14, 15, 6, 7, 9, 8, + 7, 6, 8, 13, 11, 9, 7, 15, 7, 12, 15, 9, 11, 7, 13, 12, + 11, 13, 6, 7, 14, 9, 13, 15, 14, 8, 13, 6, 5, 12, 7, 5, + 11, 12, 14, 15, 14, 15, 9, 8, 9, 14, 5, 6, 8, 6, 5, 12, + 9, 15, 5, 11, 6, 8, 13, 12, 5, 12, 13, 14, 11, 8, 5, 6, +}; +__constant uchar ripemd160_rlp[] = { + 8, 9, 9, 11, 13, 15, 15, 5, 7, 7, 8, 11, 14, 14, 12, 6, + 9, 13, 15, 7, 12, 8, 9, 11, 7, 7, 12, 7, 6, 15, 13, 11, + 9, 7, 15, 11, 8, 6, 6, 14, 12, 13, 5, 14, 13, 13, 7, 5, + 15, 5, 8, 11, 14, 14, 6, 14, 6, 9, 12, 9, 12, 5, 15, 8, + 8, 5, 12, 9, 12, 5, 14, 6, 8, 13, 6, 5, 15, 13, 11, 11 +}; + +#define ripemd160_f0(x, y, z) (x ^ y ^ z) +#define ripemd160_f1(x, y, z) ((x & y) | (~x & z)) +#define ripemd160_f2(x, y, z) ((x | ~y) ^ z) +#define ripemd160_f3(x, y, z) ((x & z) | (y & ~z)) +#define ripemd160_f4(x, y, z) (x ^ (y | ~z)) +#define ripemd160_round(i, in, vals, f, fp, t) do { \ + t = rotate(vals[0] + \ + f(vals[1], vals[2], vals[3]) + \ + in[ripemd160_ws[i]] + \ + ripemd160_k[i / 16], \ + (uint)ripemd160_rl[i]) + vals[4]; \ + vals[0] = vals[4]; vals[4] = vals[3]; \ + vals[3] = rotate(vals[2], 10U); vals[2] = vals[1]; \ + vals[1] = t; \ + t = rotate(vals[5] + \ + fp(vals[6], vals[7], vals[8]) + \ + in[ripemd160_wsp[i]] + \ + ripemd160_kp[i / 16], \ + (uint)ripemd160_rlp[i]) + vals[9]; \ + vals[5] = vals[9]; vals[9] = vals[8]; \ + vals[8] = rotate(vals[7], 10U); vals[7] = vals[6]; \ + vals[6] = t; \ + } while (0) + +void +ripemd160_init(uint *out) +{ + int i; + for(i = 0; i < 5; i++) + out[i] = ripemd160_iv[i]; +} + +void +ripemd160_block(uint *out, uint *in) +{ + uint vals[10], t; + int i; + for (i = 0; i < 5; i++) + vals[i] = vals[i + 5] = out[i]; + for (i = 0; i < 16; i++) + ripemd160_round(i, in, vals, + ripemd160_f0, ripemd160_f4, t); + for (i = 16; i < 32; i++) + ripemd160_round(i, in, vals, + ripemd160_f1, ripemd160_f3, t); + for (i = 32; i < 48; i++) + ripemd160_round(i, in, vals, + ripemd160_f2, ripemd160_f2, t); + for (i = 48; i < 64; i++) + ripemd160_round(i, in, vals, + ripemd160_f3, ripemd160_f1, t); + for (i = 64; i < 80; i++) + ripemd160_round(i, in, vals, + ripemd160_f4, ripemd160_f0, t); + t = out[1] + vals[2] + vals[8]; + out[1] = out[2] + vals[3] + vals[9]; + out[2] = out[3] + vals[4] + vals[5]; + out[3] = out[4] + vals[0] + vals[6]; + out[4] = out[0] + vals[1] + vals[7]; + out[0] = t; +} + + +#define bswap32(v) \ + (((v) >> 24) | (((v) >> 8) & 0xff00) | \ + (((v) << 8) & 0xff0000) | ((v) << 24)) + + +__kernel void +calc_addrs(__global uint *hashes_out, + __global bignum *z_heap, __global bignum *point_tmp, + __global bignum *row_in, __global bignum *col_in, int ncols) +{ + uint hash1[16]; + uint hash2[16]; + uint wl, wh; + bignum rx, ry; + bignum x1, y1, a, b, c, d, e, z; + bn_word cy; + int i, o; + + /* Load the row increment point */ + o = get_global_id(0); + rx = col_in[2*o]; + ry = col_in[(2*o) + 1]; + hashes_out += (o * 5 * ncols); + z_heap += (o * 2 * ncols); + point_tmp += (o * 2 * ncols); + + /* + * Perform the EC point add. + * Add the row increment to all row points. + * Save the X,Y in the point temporary space. + * Save the Z in the z_heap for modular inversion. + */ + for (i = 0; i < ncols; i++) { + x1 = row_in[(2*i)]; + y1 = row_in[(2*i) + 1]; + + bn_mod_sub(&z, &x1, &rx); + z_heap[(ncols - 1) + i] = z; + + bn_mod_sub(&b, &y1, &ry); + bn_mod_add(&c, &x1, &rx); + bn_mod_add(&d, &y1, &ry); + bn_mul_mont(&y1, &b, &b); + bn_mul_mont(&x1, &z, &z); + bn_mul_mont(&e, &c, &x1); + bn_mod_sub(&y1, &y1, &e); + point_tmp[2*i] = y1; + bn_mod_lshift1(&y1); + bn_mod_sub(&y1, &e, &y1); + bn_mul_mont(&y1, &y1, &b); + bn_mul_mont(&a, &x1, &z); + bn_mul_mont(&c, &d, &a); + bn_mod_sub(&y1, &y1, &c); + cy = 0; + if (bn_is_odd(y1)) + cy = bn_uadd_c(&y1, &y1, modulus); + bn_rshift1(&y1); + if (cy) + y1.d[BN_NWORDS-1] |= 0x80000000; + point_tmp[(2*i)+1] = y1; + } + + /* Compute the product hierarchy in z_heap */ + for (i = ncols - 1; i > 0; i--) { + a = z_heap[(i*2) - 1]; + b = z_heap[(i*2)]; + bn_mul_mont(&z, &a, &b); + z_heap[i-1] = z; + } + + /* Invert the root, fix up 1/ZR -> R/Z */ + z = z_heap[0]; + bn_mod_inverse(&z, &z); + + for (i = 0; i < BN_NWORDS; i++) + a.d[i] = mont_rr[i]; + bn_mul_mont(&z, &z, &a); + bn_mul_mont(&z, &z, &a); + z_heap[0] = z; + + for (i = 1; i < ncols; i++) { + a = z_heap[i - 1]; + b = z_heap[(i*2) - 1]; + c = z_heap[i*2]; + bn_mul_mont(&z, &a, &c); + z_heap[(i*2) - 1] = z; + bn_mul_mont(&z, &a, &b); + z_heap[i*2] = z; + } + + for (i = 0; i < ncols; i++) { + /* + * Multiply the coordinates by the inverted Z values. + * Stash the coordinates in the hash buffer. + * SHA-2 requires big endian, and our intended hash input + * is big-endian, so swapping is unnecessary, but + * inserting the format byte in front causes a headache. + */ + a = z_heap[(ncols - 1) + i]; + bn_mul_mont(&b, &a, &a); /* Z^2 */ + x1 = point_tmp[2*i]; + bn_mul_mont(&x1, &x1, &b); /* X / Z^2 */ + bn_from_mont(&x1, &x1); + + wh = 0x00000004; /* POINT_CONVERSION_UNCOMPRESSED */ + for (o = 0; o < BN_NWORDS; o++) { + wl = wh; + wh = x1.d[(BN_NWORDS - 1) - o]; + hash1[o] = (wl << 24) | (wh >> 8); + } + + bn_mul_mont(&a, &a, &b); /* Z^3 */ + y1 = point_tmp[(2*i)+1]; + bn_mul_mont(&y1, &y1, &a); /* Y / Z^3 */ + bn_from_mont(&y1, &y1); + + for (o = 0; o < BN_NWORDS; o++) { + wl = wh; + wh = y1.d[(BN_NWORDS - 1) - o]; + hash1[BN_NWORDS + o] = (wl << 24) | (wh >> 8); + } + + /* + * Hash the first 64 bytes of the buffer + */ + sha2_256_init(hash2); + sha2_256_block(hash2, hash1); + + /* + * Hash the last byte of the buffer + SHA-2 padding + */ + hash1[0] = wh << 24 | 0x800000; + hash1[1] = 0; + hash1[2] = 0; + hash1[3] = 0; + hash1[4] = 0; + hash1[5] = 0; + hash1[6] = 0; + hash1[7] = 0; + hash1[8] = 0; + hash1[9] = 0; + hash1[10] = 0; + hash1[11] = 0; + hash1[12] = 0; + hash1[13] = 0; + hash1[14] = 0; + hash1[15] = 65 * 8; + sha2_256_block(hash2, hash1); + + /* + * Hash the SHA-2 result with RIPEMD160 + * Unfortunately, SHA-2 outputs big-endian, but + * RIPEMD160 expects little-endian. Need to swap! + */ + for (o = 0; o < 8; o++) + hash2[o] = bswap32(hash2[o]); + hash2[8] = bswap32(0x80000000); + hash2[9] = 0; + hash2[10] = 0; + hash2[11] = 0; + hash2[12] = 0; + hash2[13] = 0; + hash2[14] = 32 * 8; + hash2[15] = 0; + ripemd160_init(hash1); + ripemd160_block(hash1, hash2); + + /* Copy the hash to the output buffer */ + for (o = 0; o < 5; o++) + *(hashes_out++) = hash1[o]; + } + +} diff --git a/oclvanitygen.c b/oclvanitygen.c new file mode 100644 index 0000000..a3b04aa --- /dev/null +++ b/oclvanitygen.c @@ -0,0 +1,1160 @@ +/* + * Vanitygen, vanity bitcoin address generator + * Copyright (C) 2011 + * + * Vanitygen is free software: you can redistribute it and/or modify + * it under the terms of the GNU Affero General Public License as published by + * the Free Software Foundation, either version 3 of the License, or + * any later version. + * + * Vanitygen is distributed in the hope that it will be useful, + * but WITHOUT ANY WARRANTY; without even the implied warranty of + * MERCHANTABILITY or FITNESS FOR A PARTICULAR PURPOSE. See the + * GNU Affero General Public License for more details. + * + * You should have received a copy of the GNU Affero General Public License + * along with Vanitygen. If not, see . + */ + +#include +#include +#include +#include + +#include + +#include +#include +#include + +#include + +#include "pattern.h" + + +const char *version = "0.13"; + +#define MAX_SLOT 2 + +typedef struct _vg_ocl_context_s { + vg_exec_context_t base; + cl_device_id voc_ocldid; + cl_context voc_oclctx; + cl_command_queue voc_oclcmdq; + cl_program voc_oclprog; + cl_kernel voc_oclkernel[MAX_SLOT]; + cl_event voc_oclkrnwait[MAX_SLOT]; + cl_mem voc_args[MAX_SLOT][6]; + size_t voc_arg_size[MAX_SLOT][6]; + + pthread_t voc_cpu_thread; + pthread_mutex_t voc_lock; + pthread_cond_t voc_wait; + int voc_cpu_slot; + int voc_cpu_worksize; + int voc_halt; + int voc_rekey; +} vg_ocl_context_t; + + +/* Thread synchronization stubs */ +void +vg_exec_downgrade_lock(vg_exec_context_t *vxcp) +{ +} + +int +vg_exec_upgrade_lock(vg_exec_context_t *vxcp) +{ + return 0; +} + + +/* + * OpenCL per-exec functions + */ + +int +vg_ocl_load_program(vg_context_t *vcp, vg_ocl_context_t *vocp, + const char *filename, const char *opts, const char *func) +{ + FILE *kfp; + char *buf; + int len; + size_t sz; + cl_program prog; + cl_kernel krn; + cl_int ret; + int i; + + buf = (char *) malloc(128 * 1024); + if (!buf) + return 0; + + kfp = fopen(filename, "r"); + if (!kfp) { + printf("Error loading CL kernel: %s\n", strerror(errno)); + free(buf); + return 0; + } + + len = fread(buf, 1, 128 * 1024, kfp); + fclose(kfp); + + sz = len; + prog = clCreateProgramWithSource(vocp->voc_oclctx, + 1, (const char **) &buf, &sz, + &ret); + free(buf); + if (!prog) { + printf("clCreateProgramWithSource: %d\n", ret); + return 0; + } + + if (vcp->vc_verbose > 0) { + printf("Compiling kernel..."); + fflush(stdout); + } + ret = clBuildProgram(prog, 1, &vocp->voc_ocldid, opts, NULL, NULL); + if (ret != CL_SUCCESS) { + if (vcp->vc_verbose > 0) + printf("failure.\n"); + printf("clBuildProgram: %d\n", ret); + } else if (vcp->vc_verbose > 0) { + printf("done!\n"); + } + if ((ret != CL_SUCCESS) || (vcp->vc_verbose > 1)) { + const size_t logbufsize = 1024 * 16; + char *log = (char*) malloc(logbufsize); + size_t logsize; + cl_int ret2; + + ret2 = clGetProgramBuildInfo(prog, + vocp->voc_ocldid, + CL_PROGRAM_BUILD_LOG, + logbufsize, + log, + &logsize); + if (ret2 != CL_SUCCESS) { + printf("clGetProgramBuildInfo: %d\n", ret2); + } else { + printf("Build log:%s\n", log); + } + free(log); + } + if (ret != CL_SUCCESS) { + clReleaseProgram(prog); + return 0; + } + + for (i = 0; i < 2; i++) { + krn = clCreateKernel(prog, func, &ret); + if (!krn) { + clReleaseProgram(prog); + printf("clCreateKernel(%d): %d\n", i, ret); + while (--i >= 0) { + clReleaseKernel(vocp->voc_oclkernel[i]); + vocp->voc_oclkernel[i] = NULL; + } + return 0; + } + vocp->voc_oclkernel[i] = krn; + vocp->voc_oclkrnwait[i] = NULL; + } + vocp->voc_oclprog = prog; + return 1; +} + +void +vg_ocl_context_callback(const char *errinfo, + const void *private_info, + size_t cb, + void *user_data) +{ + printf("vg_ocl_context_callback error: %s\n", errinfo); +} + +int +vg_ocl_init(vg_context_t *vcp, vg_ocl_context_t *vocp, cl_device_id did) +{ + cl_int ret; + + memset(vocp, 0, sizeof(*vocp)); + vg_exec_context_init(vcp, &vocp->base); + + pthread_mutex_init(&vocp->voc_lock, NULL); + pthread_cond_init(&vocp->voc_wait, NULL); + vocp->voc_cpu_slot = -1; + + vocp->voc_ocldid = did; + vocp->voc_oclctx = clCreateContext(NULL, + 1, &did, + vg_ocl_context_callback, + NULL, + &ret); + if (!vocp->voc_oclctx) { + printf("clCreateContext failed: %d\n", ret); + return 0; + } + + vocp->voc_oclcmdq = clCreateCommandQueue(vocp->voc_oclctx, + vocp->voc_ocldid, + 0, &ret); + if (!vocp->voc_oclcmdq) { + printf("clCreateCommandQueue failed: %d\n", ret); + return 0; + } + + if (!vg_ocl_load_program(vcp, vocp, + "calc_addrs.cl", + NULL, //"-cl-nv-verbose", + "calc_addrs")) { + printf("Could not load kernel\n"); + return 0; + } + return 1; +} + +int +vg_ocl_kernel_arg_alloc(vg_ocl_context_t *vocp, int slot, + int arg, size_t size) +{ + cl_mem clbuf; + cl_int ret; + int i; + + clbuf = clCreateBuffer(vocp->voc_oclctx, + CL_MEM_READ_WRITE | CL_MEM_ALLOC_HOST_PTR, + size, + NULL, + &ret); + if (!clbuf) { + printf("Could not create argument buffer: %d\n", ret); + return 0; + } + + for (i = 0; i < 2; i++) { + if ((i != slot) && (slot >= 0)) + continue; + ret = clSetKernelArg(vocp->voc_oclkernel[i], + arg, + sizeof(clbuf), + &clbuf); + if (ret) { + clReleaseMemObject(clbuf); + printf("Could not set kernel argument: %d\n", ret); + return 0; + } + vocp->voc_args[i][arg] = clbuf; + vocp->voc_arg_size[i][arg] = size; + } + return 1; +} + +void * +vg_ocl_map_arg_buffer(vg_ocl_context_t *vocp, int slot, + int arg, int rw) +{ + void *buf; + cl_int ret; + + assert(slot >= 0); + + buf = clEnqueueMapBuffer(vocp->voc_oclcmdq, + vocp->voc_args[slot][arg], + CL_TRUE, + rw ? CL_MAP_WRITE : CL_MAP_READ, + 0, vocp->voc_arg_size[slot][arg], + 0, NULL, + NULL, + &ret); + if (!buf) { + printf("Could not map argument buffer: %d\n", ret); + return NULL; + } + return buf; +} + +void +vg_ocl_unmap_arg_buffer(vg_ocl_context_t *vocp, int slot, + int arg, void *buf) +{ + cl_int ret; + cl_event ev; + + assert(slot >= 0); + + ret = clEnqueueUnmapMemObject(vocp->voc_oclcmdq, + vocp->voc_args[slot][arg], + buf, + 0, NULL, + &ev); + if (ret != CL_SUCCESS) { + printf("Could not unmap buffer: %d\n", ret); + return; + } + + ret = clWaitForEvents(1, &ev); + clReleaseEvent(ev); + if (ret != CL_SUCCESS) { + printf("Error waiting for event: %d\n", ret); + } +} + +int +vg_ocl_kernel_int_arg(vg_ocl_context_t *vocp, int slot, + int arg, int value) +{ + cl_int ret; + int i; + + for (i = 0; i < 2; i++) { + if ((i != slot) && (slot >= 0)) + continue; + ret = clSetKernelArg(vocp->voc_oclkernel[i], + arg, + sizeof(value), + &value); + if (ret) { + printf("Could not set kernel argument: %d\n", + ret); + return 0; + } + } + return 1; +} + +int +vg_ocl_kernel_dead(vg_ocl_context_t *vocp, int slot) +{ + return (vocp->voc_oclkrnwait[slot] == NULL); +} + +int +vg_ocl_kernel_start(vg_ocl_context_t *vocp, int slot, int worksize) +{ + cl_int ret; + cl_event ev; + size_t globalws = worksize; + + assert(!vocp->voc_oclkrnwait[slot]); + + ret = clEnqueueNDRangeKernel(vocp->voc_oclcmdq, + vocp->voc_oclkernel[slot], + 1, + NULL, &globalws, NULL, + 0, NULL, + &ev); + if (ret != CL_SUCCESS) { + printf("Could not queue kernel: %d\n", ret); + return 0; + } + + vocp->voc_oclkrnwait[slot] = ev; + return 1; +} + +int +vg_ocl_kernel_wait(vg_ocl_context_t *vocp, int slot) +{ + cl_event ev; + cl_int ret; + + ev = vocp->voc_oclkrnwait[slot]; + vocp->voc_oclkrnwait[slot] = NULL; + if (ev) { + ret = clWaitForEvents(1, &ev); + clReleaseEvent(ev); + if (ret != CL_SUCCESS) { + printf("Error waiting for event: %d\n", ret); + return 0; + } + } + return 1; +} + + +/* + * Absolutely disgusting. + * We want points in Montgomery form, and it's a lot easier to read the + * coordinates from the structure than to export and re-montgomeryize. + */ + +struct ec_point_st { + const EC_METHOD *meth; + BIGNUM X; + BIGNUM Y; + BIGNUM Z; + int Z_is_one; +}; + +INLINE void +vg_ocl_put_point(unsigned char *buf, EC_POINT *ppnt) +{ + assert(ppnt->Z_is_one); + memcpy(buf, ppnt->X.d, 32); + memcpy(buf + 32, ppnt->Y.d, 32); +} + + +void * +vg_opencl_cpu_thread(void *arg) +{ + vg_ocl_context_t *vocp = (vg_ocl_context_t *) arg; + vg_exec_context_t *vxcp = &vocp->base; + vg_context_t *vcp = vxcp->vxc_vc; + unsigned char *ocl_hashes_out; + vg_test_func_t test_func = vcp->vc_test; + int i, c = 0, output_interval = 1000; + int rekey = 0; + int halt = 0; + int slot = -1; + int round; + struct timeval tvstart; + + gettimeofday(&tvstart, NULL); + + while (1) { + pthread_mutex_lock(&vocp->voc_lock); + if (rekey) { + rekey = 0; + vocp->voc_rekey = 1; + } + if (halt) { + halt = 0; + vocp->voc_halt = 1; + } + if (slot != -1) { + assert(vocp->voc_cpu_slot == slot); + vocp->voc_cpu_slot = -1; + slot = -1; + pthread_cond_signal(&vocp->voc_wait); + } + if (vocp->voc_halt) + break; + while (vocp->voc_cpu_slot == -1) { + pthread_cond_wait(&vocp->voc_wait, &vocp->voc_lock); + if (vocp->voc_halt) + break; + } + assert(!vocp->voc_rekey); + assert(!vocp->voc_halt); + slot = vocp->voc_cpu_slot; + round = vocp->voc_cpu_worksize; + pthread_mutex_unlock(&vocp->voc_lock); + + + ocl_hashes_out = (unsigned char *) + vg_ocl_map_arg_buffer(vocp, slot, 0, 0); + + for (i = 0; i < round; i++, vxcp->vxc_delta++) { + + memcpy(&vxcp->vxc_binres[1], + ocl_hashes_out + (20*i), + 20); + + switch (test_func(vxcp)) { + case 1: + rekey = 1; + i = round; + break; + case 2: + halt = 1; + i = round; + break; + default: + break; + } + } + + vg_ocl_unmap_arg_buffer(vocp, slot, 0, ocl_hashes_out); + + c += (i + 1); + if (c >= output_interval) { + output_interval = vg_output_timing(vcp, c, &tvstart); + c = 0; + } + } + pthread_mutex_unlock(&vocp->voc_lock); + return NULL; +} + +/* + * Address search thread main loop + */ + +void * +vg_opencl_loop(vg_context_t *vcp, cl_device_id did, int worksize) +{ + int i; + int batchsize, round; + + const BN_ULONG rekey_max = 20000000; + BN_ULONG npoints, rekey_at; + + EC_KEY *pkey = NULL; + const EC_GROUP *pgroup; + const EC_POINT *pgen; + EC_POINT **ppbase = NULL, **pprow, *pbatchinc = NULL, *poffset = NULL; + EC_POINT *pseek = NULL; + + unsigned char *ocl_points_in, *ocl_strides_in; + + vg_ocl_context_t ctx; + vg_ocl_context_t *vocp = &ctx; + vg_exec_context_t *vxcp = &vocp->base; + + int slot, nslots; + + if (!vg_ocl_init(vcp, &ctx, did)) + return NULL; + + pkey = vxcp->vxc_key; + pgroup = EC_KEY_get0_group(pkey); + pgen = EC_GROUP_get0_generator(pgroup); + + /* + * batchsize: number of points to process in each thread + * worksize: number of threads per kernel + * nslots: number of kernels + */ + + batchsize = 256; + if (!worksize) + worksize = 512; + nslots = 1; + slot = 0; + + ppbase = (EC_POINT **) malloc((batchsize + worksize) * + sizeof(EC_POINT*)); + if (!ppbase) + goto enomem; + + for (i = 0; i < (batchsize + worksize); i++) { + ppbase[i] = EC_POINT_new(pgroup); + if (!ppbase[i]) + goto enomem; + } + + pprow = ppbase + batchsize; + pbatchinc = EC_POINT_new(pgroup); + poffset = EC_POINT_new(pgroup); + pseek = EC_POINT_new(pgroup); + if (!pbatchinc || !poffset || !pseek) + goto enomem; + + BN_set_word(&vxcp->vxc_bntmp, batchsize); + EC_POINT_mul(pgroup, pbatchinc, &vxcp->vxc_bntmp, NULL, NULL, + vxcp->vxc_bnctx); + EC_POINT_make_affine(pgroup, pbatchinc, vxcp->vxc_bnctx); + + BN_set_word(&vxcp->vxc_bntmp, worksize * batchsize); + EC_POINT_mul(pgroup, poffset, &vxcp->vxc_bntmp, NULL, NULL, + vxcp->vxc_bnctx); + EC_POINT_make_affine(pgroup, poffset, vxcp->vxc_bnctx); + + round = batchsize * worksize; + + for (i = 0; i < nslots; i++) { + /* + * Each work group gets its own: + * - Hash output array + * - Point and z_heap scratch spaces + * - Column point array + */ + if (!vg_ocl_kernel_arg_alloc(vocp, i, 0, 20 * round) || + !vg_ocl_kernel_arg_alloc(vocp, i, 1, 32 * 2 * round) || + !vg_ocl_kernel_arg_alloc(vocp, i, 2, 32 * 2 * round) || + !vg_ocl_kernel_arg_alloc(vocp, i, 4, 32 * 2 * worksize)) + goto enomem; + } + + /* Same row point array for all instances */ + if (!vg_ocl_kernel_arg_alloc(vocp, -1, 3, 32 * 2 * batchsize)) + goto enomem; + + vg_ocl_kernel_int_arg(vocp, -1, 5, batchsize); + + npoints = 0; + rekey_at = 0; + vxcp->vxc_binres[0] = vcp->vc_addrtype; + + if (pthread_create(&vocp->voc_cpu_thread, NULL, + vg_opencl_cpu_thread, vocp)) + goto enomem; + +rekey: + /* Generate a new random private key */ + EC_KEY_generate_key(pkey); + npoints = 0; + + /* Determine rekey interval */ + EC_GROUP_get_order(pgroup, &vxcp->vxc_bntmp, vxcp->vxc_bnctx); + BN_sub(&vxcp->vxc_bntmp2, + &vxcp->vxc_bntmp, + EC_KEY_get0_private_key(pkey)); + rekey_at = BN_get_word(&vxcp->vxc_bntmp2); + if ((rekey_at == BN_MASK2) || (rekey_at > rekey_max)) + rekey_at = rekey_max; + assert(rekey_at > 0); + + EC_POINT_copy(ppbase[0], EC_KEY_get0_public_key(pkey)); + + /* Build the base array of sequential points */ + for (i = 1; i < batchsize; i++) { + EC_POINT_add(pgroup, + ppbase[i], + ppbase[i-1], + pgen, vxcp->vxc_bnctx); + } + + EC_POINTs_make_affine(pgroup, batchsize, ppbase, + vxcp->vxc_bnctx); + + /* Fill the sequential point array */ + ocl_points_in = (unsigned char *) + vg_ocl_map_arg_buffer(vocp, 0, 3, 1); + if (!ocl_points_in) + goto enomem; + for (i = 0; i < batchsize; i++) + vg_ocl_put_point(ocl_points_in + (64*i), ppbase[i]); + vg_ocl_unmap_arg_buffer(vocp, 0, 3, ocl_points_in); + + /* + * Set up the initial row increment table. + * Set the first element to pgen -- effectively + * skipping the exact key generated above. + */ + EC_POINT_copy(pprow[0], pgen); + for (i = 1; i < worksize; i++) { + EC_POINT_add(pgroup, + pprow[i], + pprow[i-1], + pbatchinc, vxcp->vxc_bnctx); + } + EC_POINTs_make_affine(pgroup, worksize, pprow, vxcp->vxc_bnctx); + + vxcp->vxc_delta = 1; + npoints = 1; + slot = 0; + + while (1) { + if (((npoints + round) < rekey_at) && + vg_ocl_kernel_dead(vocp, slot)) { + + if (npoints > 1) { + /* Move the row increments forward */ + for (i = 0; i < worksize; i++) { + EC_POINT_add(pgroup, + pprow[i], + pprow[i], + poffset, + vxcp->vxc_bnctx); + } + + EC_POINTs_make_affine(pgroup, worksize, pprow, + vxcp->vxc_bnctx); + } + + /* Copy the row stride array to the device */ + ocl_strides_in = (unsigned char *) + vg_ocl_map_arg_buffer(vocp, slot, 4, 1); + if (!ocl_strides_in) + goto enomem; + memset(ocl_strides_in, 0, 64*worksize); + for (i = 0; i < worksize; i++) + vg_ocl_put_point(ocl_strides_in + (64*i), + pprow[i]); + vg_ocl_unmap_arg_buffer(vocp, slot, 4, ocl_strides_in); + + /* Kick off the kernel */ + if (!vg_ocl_kernel_start(vocp, slot, worksize)) + exit(1); + + slot = (slot + 1) % nslots; + npoints += round; + continue; + } + + else if (vg_ocl_kernel_dead(vocp, slot)) { + slot = (slot + 1) % nslots; + if (vg_ocl_kernel_dead(vocp, slot)) + goto rekey; + } + + vg_ocl_kernel_wait(vocp, slot); + + if (npoints >= rekey_at) + continue; + + pthread_mutex_lock(&vocp->voc_lock); + recheck: + if (vocp->voc_halt) { + pthread_mutex_unlock(&vocp->voc_lock); + pthread_join(vocp->voc_cpu_thread, NULL); + goto out; + } + if (vocp->voc_rekey) { + vocp->voc_rekey = 0; + rekey_at = 0; + pthread_mutex_unlock(&vocp->voc_lock); + continue; + } + if (vocp->voc_cpu_slot != -1) { + pthread_cond_wait(&vocp->voc_wait, &vocp->voc_lock); + goto recheck; + } + + vocp->voc_cpu_slot = slot; + vocp->voc_cpu_worksize = round; + pthread_cond_signal(&vocp->voc_wait); + pthread_mutex_unlock(&vocp->voc_lock); + } + + if (0) { + enomem: + printf("ERROR: allocation failure?\n"); + } + +out: + if (ppbase) { + for (i = 0; i < (batchsize + worksize); i++) + if (ppbase[i]) + EC_POINT_free(ppbase[i]); + free(ppbase); + } + if (pbatchinc) + EC_POINT_free(pbatchinc); + return NULL; +} + + + + +/* + * OpenCL platform/device selection junk + */ + +int +get_device_list(cl_platform_id pid, cl_device_id **list_out) +{ + cl_uint nd; + cl_int res; + cl_device_id *ids; + res = clGetDeviceIDs(pid, CL_DEVICE_TYPE_ALL, 0, NULL, &nd); + if (res != CL_SUCCESS) { + printf("clGetDeviceIDs(0) failed: %d\n", res); + *list_out = NULL; + return -1; + } + if (nd) { + ids = (cl_device_id *) malloc(nd * sizeof(*ids)); + if (ids == NULL) { + printf("Could not allocate device ID list\n"); + *list_out = NULL; + return -1; + } + res = clGetDeviceIDs(pid, CL_DEVICE_TYPE_ALL, nd, ids, NULL); + if (res != CL_SUCCESS) { + printf("clGetDeviceIDs(n) failed: %d\n", res); + free(ids); + *list_out = NULL; + return -1; + } + *list_out = ids; + } + return nd; +} + +void +show_devices(cl_platform_id pid, cl_device_id *ids, int nd, int base) +{ + int i; + char nbuf[128]; + char vbuf[128]; + size_t len; + cl_int res; + + for (i = 0; i < nd; i++) { + res = clGetDeviceInfo(ids[i], CL_DEVICE_NAME, + sizeof(nbuf), nbuf, &len); + if (res != CL_SUCCESS) + continue; + if (len >= sizeof(nbuf)) + len = sizeof(nbuf) - 1; + nbuf[len] = '\0'; + res = clGetDeviceInfo(ids[i], CL_DEVICE_VENDOR, + sizeof(vbuf), vbuf, &len); + if (res != CL_SUCCESS) + continue; + if (len >= sizeof(vbuf)) + len = sizeof(vbuf) - 1; + vbuf[len] = '\0'; + printf(" %d: [%s] %s\n", i + base, vbuf, nbuf); + } +} + +cl_device_id +get_device(cl_platform_id pid, int num) +{ + int nd; + cl_device_id id, *ids; + + nd = get_device_list(pid, &ids); + if (nd < 0) + return NULL; + if (!nd) { + printf("No OpenCL devices found\n"); + return NULL; + } + if (num < 0) { + if (nd == 1) + num = 0; + else + num = nd; + } + if (num < nd) { + id = ids[num]; + free(ids); + return id; + } + free(ids); + return NULL; +} + +int +get_platform_list(cl_platform_id **list_out) +{ + cl_uint np; + cl_int res; + cl_platform_id *ids; + res = clGetPlatformIDs(0, NULL, &np); + if (res != CL_SUCCESS) { + printf("clGetPlatformIDs(0) failed: %d\n", res); + *list_out = NULL; + return -1; + } + if (np) { + ids = (cl_platform_id *) malloc(np * sizeof(*ids)); + if (ids == NULL) { + printf("Could not allocate platform ID list\n"); + *list_out = NULL; + return -1; + } + res = clGetPlatformIDs(np, ids, NULL); + if (res != CL_SUCCESS) { + printf("clGetPlatformIDs(n) failed: %d\n", res); + free(ids); + *list_out = NULL; + return -1; + } + *list_out = ids; + } + return np; +} + +void +show_platforms(cl_platform_id *ids, int np, int base) +{ + int i; + char nbuf[128]; + char vbuf[128]; + size_t len; + cl_int res; + + for (i = 0; i < np; i++) { + res = clGetPlatformInfo(ids[i], CL_PLATFORM_NAME, + sizeof(nbuf), nbuf, &len); + if (res != CL_SUCCESS) { + printf("Failed to enumerate platform ID: %d\n", res); + continue; + } + if (len >= sizeof(nbuf)) + len = sizeof(nbuf) - 1; + nbuf[len] = '\0'; + res = clGetPlatformInfo(ids[i], CL_PLATFORM_VENDOR, + sizeof(vbuf), vbuf, &len); + if (res != CL_SUCCESS) { + printf("Failed to enumerate platform ID: %d\n", res); + continue; + } + if (len >= sizeof(vbuf)) + len = sizeof(vbuf) - 1; + vbuf[len] = '\0'; + printf("%d: [%s] %s\n", i + base, vbuf, nbuf); + } +} + +cl_platform_id +get_platform(int num) +{ + int np; + cl_platform_id id, *ids; + + np = get_platform_list(&ids); + if (np < 0) + return NULL; + if (!np) { + printf("No OpenCL platforms available\n"); + return NULL; + } + if (num < 0) { + if (np == 1) + num = 0; + else + num = np; + } + if (num < np) { + id = ids[num]; + free(ids); + return id; + } + free(ids); + return NULL; +} + +void +enumerate_opencl(void) +{ + cl_platform_id *pids; + cl_device_id *dids; + int np, nd, i; + + np = get_platform_list(&pids); + if (!np) { + printf("No OpenCL platforms available\n"); + return; + } + printf("Available OpenCL platforms:\n"); + for (i = 0; i < np; i++) { + show_platforms(&pids[i], 1, i); + nd = get_device_list(pids[i], &dids); + if (!nd) { + printf(" -- No devices\n"); + } else { + show_devices(pids[i], dids, nd, 0); + } + } +} + +cl_device_id +get_opencl_device(int platformidx, int deviceidx) +{ + cl_platform_id pid; + cl_device_id did = NULL; + + pid = get_platform(platformidx); + if (pid) { + did = get_device(pid, deviceidx); + if (did) + return did; + } + enumerate_opencl(); + return NULL; +} + + + +void +usage(const char *name) +{ + printf( +"oclVanitygen %s (" OPENSSL_VERSION_TEXT ")\n" +"Usage: %s [-vqrikNT] [-t ] [-f |-] [...]\n" +"Generates a bitcoin receiving address matching , and outputs the\n" +"address and associated private key. The private key may be stored in a safe\n" +"location or imported into a bitcoin client to spend any balance received on\n" +"the address.\n" +"By default, is interpreted as an exact prefix.\n" +"\n" +"Options:\n" +"-v Verbose output\n" +"-q Quiet output\n" +"-r Use regular expression match instead of prefix\n" +" (Feasibility of expression is not checked)\n" +"-i Case-insensitive prefix search\n" +"-k Keep pattern and continue search after finding a match\n" +"-N Generate namecoin address\n" +"-T Generate bitcoin testnet address\n" +"-p Select OpenCL platform\n" +"-d Select OpenCL device\n" +"-w Set OpenCL work size (Default: number of CPUs)\n" +"-f File containing list of patterns, one per line\n" +" (Use \"-\" as the file name for stdin)\n" +"-o Write pattern matches to \n" +"-s Seed random number generator from \n", +version, name); +} + +int +main(int argc, char **argv) +{ + int addrtype = 0; + int privtype = 128; + int regex = 0; + int caseinsensitive = 0; + int opt; + int platformidx = -1, deviceidx = -1; + char *seedfile = NULL; + FILE *fp = NULL; + char **patterns; + int verbose = 1; + int npatterns = 0; + int worksize = 0; + int remove_on_match = 1; + vg_context_t *vcp = NULL; + cl_device_id did; + const char *result_file = NULL; + + while ((opt = getopt(argc, argv, "vqrikNTp:d:w:h?f:o:s:")) != -1) { + switch (opt) { + case 'v': + verbose = 2; + break; + case 'q': + verbose = 0; + break; + case 'r': + regex = 1; + break; + case 'i': + caseinsensitive = 1; + break; + case 'k': + remove_on_match = 0; + break; + case 'N': + addrtype = 52; + break; + case 'T': + addrtype = 111; + privtype = 239; + break; + case 'p': + platformidx = atoi(optarg); + break; + case 'd': + deviceidx = atoi(optarg); + break; + case 'w': + worksize = atoi(optarg); + if (worksize == 0) { + printf("Invalid thread count '%s'\n", optarg); + return 1; + } + break; + case 'f': + if (fp) { + printf("Multiple files specified\n"); + return 1; + } + if (!strcmp(optarg, "-")) { + fp = stdin; + } else { + fp = fopen(optarg, "r"); + if (!fp) { + printf("Could not open %s: %s\n", + optarg, strerror(errno)); + return 1; + } + } + break; + case 'o': + if (result_file) { + printf("Multiple output files specified\n"); + return 1; + } + result_file = optarg; + break; + case 's': + if (seedfile != NULL) { + printf("Multiple RNG seeds specified\n"); + return 1; + } + seedfile = optarg; + break; + default: + usage(argv[0]); + return 1; + } + } + +#if OPENSSL_VERSION_NUMBER < 0x10000000L + /* Complain about older versions of OpenSSL */ + if (verbose > 0) { + printf("WARNING: Built with " OPENSSL_VERSION_TEXT "\n" + "WARNING: Use OpenSSL 1.0.0d+ for best performance\n"); + } +#endif + + if (caseinsensitive && regex) + printf("WARNING: case insensitive mode incompatible with " + "regular expressions\n"); + + if (seedfile) { + opt = -1; +#if !defined(_WIN32) + { struct stat st; + if (!stat(seedfile, &st) && + (st.st_mode & (S_IFBLK|S_IFCHR))) { + opt = 32; + } } +#endif + opt = RAND_load_file(seedfile, opt); + if (!opt) { + printf("Could not load RNG seed %s\n", optarg); + return 1; + } + if (verbose > 0) { + printf("Read %d bytes from RNG seed file\n", opt); + } + } + + if (fp) { + if (!vg_read_file(fp, &patterns, &npatterns)) { + printf("Failed to load pattern file\n"); + return 1; + } + if (fp != stdin) + fclose(fp); + + } else { + if (optind >= argc) { + usage(argv[0]); + return 1; + } + patterns = &argv[optind]; + npatterns = argc - optind; + } + + if (regex) { + vcp = vg_regex_context_new(addrtype, privtype); + + } else { + vcp = vg_prefix_context_new(addrtype, privtype, + caseinsensitive); + } + + vcp->vc_verbose = verbose; + vcp->vc_result_file = result_file; + vcp->vc_remove_on_match = remove_on_match; + + if (!vg_context_add_patterns(vcp, patterns, npatterns)) + return 1; + + if (!vcp->vc_npatterns) { + printf("No patterns to search\n"); + return 1; + } + + if ((verbose > 0) && regex && (vcp->vc_npatterns > 1)) + printf("Regular expressions: %ld\n", vcp->vc_npatterns); + + did = get_opencl_device(platformidx, deviceidx); + if (!did) { + return 1; + } + + vg_opencl_loop(vcp, did, worksize); + return 0; +} diff --git a/pattern.c b/pattern.c new file mode 100644 index 0000000..184dc02 --- /dev/null +++ b/pattern.c @@ -0,0 +1,1892 @@ +/* + * Vanitygen, vanity bitcoin address generator + * Copyright (C) 2011 + * + * Vanitygen is free software: you can redistribute it and/or modify + * it under the terms of the GNU Affero General Public License as published by + * the Free Software Foundation, either version 3 of the License, or + * any later version. + * + * Vanitygen is distributed in the hope that it will be useful, + * but WITHOUT ANY WARRANTY; without even the implied warranty of + * MERCHANTABILITY or FITNESS FOR A PARTICULAR PURPOSE. See the + * GNU Affero General Public License for more details. + * + * You should have received a copy of the GNU Affero General Public License + * along with Vanitygen. If not, see . + */ + +#include +#include +#include +#include + +#include + +#include +#include +#include +#include +#include + +#include + +#include "pattern.h" + +static const char *b58_alphabet = "123456789ABCDEFGHJKLMNPQRSTUVWXYZabcdefghijkmnopqrstuvwxyz"; + +static void +encode_b58_check(void *buf, size_t len, char *result) +{ + unsigned char hash1[32]; + unsigned char hash2[32]; + + int d, p; + + BN_CTX *bnctx; + BIGNUM *bn, *bndiv, *bntmp; + BIGNUM bna, bnb, bnbase, bnrem; + unsigned char *binres; + int brlen, zpfx; + + bnctx = BN_CTX_new(); + BN_init(&bna); + BN_init(&bnb); + BN_init(&bnbase); + BN_init(&bnrem); + BN_set_word(&bnbase, 58); + + bn = &bna; + bndiv = &bnb; + + brlen = (2 * len) + 4; + binres = (unsigned char*) malloc(brlen); + memcpy(binres, buf, len); + + SHA256(binres, len, hash1); + SHA256(hash1, sizeof(hash1), hash2); + memcpy(&binres[len], hash2, 4); + + BN_bin2bn(binres, len + 4, bn); + + for (zpfx = 0; zpfx < (len + 4) && binres[zpfx] == 0; zpfx++); + + p = brlen; + while (!BN_is_zero(bn)) { + BN_div(bndiv, &bnrem, bn, &bnbase, bnctx); + bntmp = bn; + bn = bndiv; + bndiv = bntmp; + d = BN_get_word(&bnrem); + binres[--p] = b58_alphabet[d]; + } + + while (zpfx--) { + binres[--p] = b58_alphabet[0]; + } + + memcpy(result, &binres[p], brlen - p); + result[brlen - p] = '\0'; + + free(binres); + BN_clear_free(&bna); + BN_clear_free(&bnb); + BN_clear_free(&bnbase); + BN_clear_free(&bnrem); + BN_CTX_free(bnctx); +} + +void +vg_encode_address(EC_KEY *pkey, int addrtype, char *result) +{ + unsigned char eckey_buf[128], *pend; + unsigned char binres[21] = {0,}; + unsigned char hash1[32]; + + pend = eckey_buf; + + i2o_ECPublicKey(pkey, &pend); + + binres[0] = addrtype; + SHA256(eckey_buf, pend - eckey_buf, hash1); + RIPEMD160(hash1, sizeof(hash1), &binres[1]); + + encode_b58_check(binres, sizeof(binres), result); +} + +void +vg_encode_privkey(EC_KEY *pkey, int addrtype, char *result) +{ + unsigned char eckey_buf[128]; + const BIGNUM *bn; + int nbytes; + + bn = EC_KEY_get0_private_key(pkey); + + eckey_buf[0] = addrtype; + nbytes = BN_bn2bin(bn, &eckey_buf[1]); + + encode_b58_check(eckey_buf, nbytes + 1, result); +} + + +void +dumphex(const unsigned char *src, size_t len) +{ + size_t i; + for (i = 0; i < len; i++) { + printf("%02x", src[i]); + } + printf("\n"); +} + +void +dumpbn(const BIGNUM *bn) +{ + char *buf; + buf = BN_bn2hex(bn); + printf("%s\n", buf ? buf : "0"); + if (buf) + OPENSSL_free(buf); +} + +/* + * Common code for execution helper + */ + +int +vg_exec_context_init(vg_context_t *vcp, vg_exec_context_t *vxcp) +{ + memset(vxcp, 0, sizeof(*vxcp)); + + vxcp->vxc_vc = vcp; + + BN_init(&vxcp->vxc_bntarg); + BN_init(&vxcp->vxc_bnbase); + BN_init(&vxcp->vxc_bntmp); + BN_init(&vxcp->vxc_bntmp2); + + BN_set_word(&vxcp->vxc_bnbase, 58); + + vxcp->vxc_bnctx = BN_CTX_new(); + assert(vxcp->vxc_bnctx); + vxcp->vxc_key = EC_KEY_new_by_curve_name(NID_secp256k1); + assert(vxcp->vxc_key); + vxcp->vxc_point = EC_POINT_new(EC_KEY_get0_group(vxcp->vxc_key)); + assert(vxcp->vxc_point); + EC_KEY_precompute_mult(vxcp->vxc_key, vxcp->vxc_bnctx); + return 1; +} + +void +vg_exec_context_del(vg_exec_context_t *vxcp) +{ + BN_clear_free(&vxcp->vxc_bntarg); + BN_clear_free(&vxcp->vxc_bnbase); + BN_clear_free(&vxcp->vxc_bntmp); + BN_clear_free(&vxcp->vxc_bntmp2); + BN_CTX_free(vxcp->vxc_bnctx); + vxcp->vxc_bnctx = NULL; +} + +void +vg_exec_context_consolidate_key(vg_exec_context_t *vxcp) +{ + if (vxcp->vxc_delta) { + BN_clear(&vxcp->vxc_bntmp); + BN_set_word(&vxcp->vxc_bntmp, vxcp->vxc_delta); + BN_add(&vxcp->vxc_bntmp2, + EC_KEY_get0_private_key(vxcp->vxc_key), + &vxcp->vxc_bntmp); + EC_KEY_set_private_key(vxcp->vxc_key, &vxcp->vxc_bntmp2); + EC_POINT_mul(EC_KEY_get0_group(vxcp->vxc_key), + vxcp->vxc_point, &vxcp->vxc_bntmp2, + NULL, NULL, vxcp->vxc_bnctx); + EC_KEY_set_public_key(vxcp->vxc_key, vxcp->vxc_point); + vxcp->vxc_delta = 0; + } +} + + +typedef struct _timing_info_s { + struct _timing_info_s *ti_next; + pthread_t ti_thread; + unsigned long ti_last_rate; +} timing_info_t; + +int +vg_output_timing(vg_context_t *vcp, int cycle, struct timeval *last) +{ + static unsigned long long total = 0, prevfound = 0, sincelast = 0; + static pthread_mutex_t mutex = PTHREAD_MUTEX_INITIALIZER; + static timing_info_t *timing_head = NULL; + + pthread_t me; + struct timeval tvnow, tv; + timing_info_t *tip, *mytip; + unsigned long long rate, myrate; + double count, prob, time, targ; + char linebuf[80]; + char *unit; + int rem, p, i; + + const double targs[] = { 0.5, 0.75, 0.8, 0.9, 0.95, 1.0 }; + + /* Compute the rate */ + gettimeofday(&tvnow, NULL); + timersub(&tvnow, last, &tv); + memcpy(last, &tvnow, sizeof(*last)); + myrate = tv.tv_usec + (1000000 * tv.tv_sec); + myrate = (1000000ULL * cycle) / myrate; + + pthread_mutex_lock(&mutex); + me = pthread_self(); + rate = myrate; + for (tip = timing_head, mytip = NULL; tip != NULL; tip = tip->ti_next) { + if (pthread_equal(tip->ti_thread, me)) { + mytip = tip; + tip->ti_last_rate = myrate; + } else + rate += tip->ti_last_rate; + } + if (!mytip) { + mytip = (timing_info_t *) malloc(sizeof(*tip)); + mytip->ti_next = timing_head; + mytip->ti_thread = me; + timing_head = mytip; + mytip->ti_last_rate = myrate; + } + + total += cycle; + if (prevfound != vcp->vc_found) { + prevfound = vcp->vc_found; + sincelast = 0; + } + sincelast += cycle; + count = sincelast; + + if (mytip != timing_head) { + pthread_mutex_unlock(&mutex); + return myrate; + } + pthread_mutex_unlock(&mutex); + + targ = rate; + unit = "key/s"; + if (targ > 1000) { + unit = "Kkey/s"; + targ /= 1000.0; + if (targ > 1000) { + unit = "Mkey/s"; + targ /= 1000.0; + } + } + + rem = sizeof(linebuf); + p = snprintf(linebuf, rem, "[%.2f %s][total %lld]", + targ, unit, total); + assert(p > 0); + rem -= p; + if (rem < 0) + rem = 0; + + if (vcp->vc_chance >= 1.0) { + prob = 1.0f - exp(-count/vcp->vc_chance); + + if (prob <= 0.999) { + p = snprintf(&linebuf[p], rem, "[Prob %.1f%%]", + prob * 100); + assert(p > 0); + rem -= p; + if (rem < 0) + rem = 0; + p = sizeof(linebuf) - rem; + } + + for (i = 0; i < sizeof(targs)/sizeof(targs[0]); i++) { + targ = targs[i]; + if ((targ < 1.0) && (prob <= targ)) + break; + } + + if (targ < 1.0) { + time = ((-vcp->vc_chance * log(1.0 - targ)) - count) / + rate; + unit = "s"; + if (time > 60) { + time /= 60; + unit = "min"; + if (time > 60) { + time /= 60; + unit = "h"; + if (time > 24) { + time /= 24; + unit = "d"; + if (time > 365) { + time /= 365; + unit = "y"; + } + } + } + } + + if (time > 1000000) { + p = snprintf(&linebuf[p], rem, + "[%d%% in %e%s]", + (int) (100 * targ), time, unit); + } else { + p = snprintf(&linebuf[p], rem, + "[%d%% in %.1f%s]", + (int) (100 * targ), time, unit); + } + assert(p > 0); + rem -= p; + if (rem < 0) + rem = 0; + p = sizeof(linebuf) - rem; + } + } + + if (vcp->vc_found) { + if (vcp->vc_remove_on_match) + p = snprintf(&linebuf[p], rem, "[Found %lld/%ld]", + vcp->vc_found, vcp->vc_npatterns_start); + else + p = snprintf(&linebuf[p], rem, "[Found %lld]", + vcp->vc_found); + assert(p > 0); + rem -= p; + if (rem < 0) + rem = 0; + } + + if (rem) { + memset(&linebuf[sizeof(linebuf)-rem], 0x20, rem); + linebuf[sizeof(linebuf)-1] = '\0'; + } + printf("\r%s", linebuf); + fflush(stdout); + return myrate; +} + +void +vg_output_match(vg_context_t *vcp, EC_KEY *pkey, const char *pattern) +{ + unsigned char key_buf[512], *pend; + char addr_buf[64]; + char privkey_buf[128]; + int len; + + assert(EC_KEY_check_key(pkey)); + vg_encode_address(pkey, vcp->vc_addrtype, addr_buf); + vg_encode_privkey(pkey, vcp->vc_privtype, privkey_buf); + + if (!vcp->vc_result_file || (vcp->vc_verbose > 0)) { + printf("\r%79s\rPattern: %s\n", "", pattern); + } + + if (vcp->vc_verbose > 0) { + if (vcp->vc_verbose > 1) { + /* Hexadecimal OpenSSL notation */ + pend = key_buf; + len = i2o_ECPublicKey(pkey, &pend); + printf("Pubkey (ASN1): "); + dumphex(key_buf, len); + pend = key_buf; + len = i2d_ECPrivateKey(pkey, &pend); + printf("Privkey (ASN1): "); + dumphex(key_buf, len); + } + + } + + if (!vcp->vc_result_file || (vcp->vc_verbose > 0)) { + printf("Address: %s\n" + "Privkey: %s\n", + addr_buf, privkey_buf); + } + + if (vcp->vc_result_file) { + FILE *fp = fopen(vcp->vc_result_file, "a"); + if (!fp) { + printf("ERROR: could not open result file: %s\n", + strerror(errno)); + } else { + fprintf(fp, + "Pattern: %s\n" + "Address: %s\n" + "Privkey: %s\n", + pattern, addr_buf, privkey_buf); + fclose(fp); + } + } +} + + + +void +vg_context_free(vg_context_t *vcp) +{ + vcp->vc_free(vcp); +} + +int +vg_context_add_patterns(vg_context_t *vcp, + char ** const patterns, int npatterns) +{ + return vcp->vc_add_patterns(vcp, patterns, npatterns); +} + + +static const signed char b58_reverse_map[256] = { + -1, -1, -1, -1, -1, -1, -1, -1, -1, -1, -1, -1, -1, -1, -1, -1, + -1, -1, -1, -1, -1, -1, -1, -1, -1, -1, -1, -1, -1, -1, -1, -1, + -1, -1, -1, -1, -1, -1, -1, -1, -1, -1, -1, -1, -1, -1, -1, -1, + -1, 0, 1, 2, 3, 4, 5, 6, 7, 8, -1, -1, -1, -1, -1, -1, + -1, 9, 10, 11, 12, 13, 14, 15, 16, -1, 17, 18, 19, 20, 21, -1, + 22, 23, 24, 25, 26, 27, 28, 29, 30, 31, 32, -1, -1, -1, -1, -1, + -1, 33, 34, 35, 36, 37, 38, 39, 40, 41, 42, 43, -1, 44, 45, 46, + 47, 48, 49, 50, 51, 52, 53, 54, 55, 56, 57, -1, -1, -1, -1, -1, + -1, -1, -1, -1, -1, -1, -1, -1, -1, -1, -1, -1, -1, -1, -1, -1, + -1, -1, -1, -1, -1, -1, -1, -1, -1, -1, -1, -1, -1, -1, -1, -1, + -1, -1, -1, -1, -1, -1, -1, -1, -1, -1, -1, -1, -1, -1, -1, -1, + -1, -1, -1, -1, -1, -1, -1, -1, -1, -1, -1, -1, -1, -1, -1, -1, + -1, -1, -1, -1, -1, -1, -1, -1, -1, -1, -1, -1, -1, -1, -1, -1, + -1, -1, -1, -1, -1, -1, -1, -1, -1, -1, -1, -1, -1, -1, -1, -1, + -1, -1, -1, -1, -1, -1, -1, -1, -1, -1, -1, -1, -1, -1, -1, -1 +}; + +/* + * Find the bignum ranges that produce a given prefix. + */ +static int +get_prefix_ranges(int addrtype, const char *pfx, BIGNUM **result, + BN_CTX *bnctx) +{ + int i, p, c; + int zero_prefix = 0; + int check_upper = 0; + int b58pow, b58ceil, b58top = 0; + int ret = -1; + + BIGNUM bntarg, bnceil, bnfloor; + BIGNUM bnbase; + BIGNUM *bnap, *bnbp, *bntp; + BIGNUM *bnhigh = NULL, *bnlow = NULL, *bnhigh2 = NULL, *bnlow2 = NULL; + BIGNUM bntmp, bntmp2; + + BN_init(&bntarg); + BN_init(&bnceil); + BN_init(&bnfloor); + BN_init(&bnbase); + BN_init(&bntmp); + BN_init(&bntmp2); + + BN_set_word(&bnbase, 58); + + p = strlen(pfx); + + for (i = 0; i < p; i++) { + c = b58_reverse_map[(int)pfx[i]]; + if (c == -1) { + printf("Invalid character '%c' in prefix '%s'\n", + pfx[i], pfx); + goto out; + } + if (i == zero_prefix) { + if (c == 0) { + /* Add another zero prefix */ + zero_prefix++; + if (zero_prefix > 19) { + printf("Prefix '%s' is too long\n", + pfx); + goto out; + } + continue; + } + + /* First non-zero character */ + b58top = c; + BN_set_word(&bntarg, c); + + } else { + BN_set_word(&bntmp2, c); + BN_mul(&bntmp, &bntarg, &bnbase, bnctx); + BN_add(&bntarg, &bntmp, &bntmp2); + } + } + + /* Power-of-two ceiling and floor values based on leading 1s */ + BN_clear(&bntmp); + BN_set_bit(&bntmp, 200 - (zero_prefix * 8)); + BN_sub(&bnceil, &bntmp, BN_value_one()); + BN_set_bit(&bnfloor, 192 - (zero_prefix * 8)); + + bnlow = BN_new(); + bnhigh = BN_new(); + + if (b58top) { + /* + * If a non-zero was given in the prefix, find the + * numeric boundaries of the prefix. + */ + + BN_copy(&bntmp, &bnceil); + bnap = &bntmp; + bnbp = &bntmp2; + b58pow = 0; + while (BN_cmp(bnap, &bnbase) > 0) { + b58pow++; + BN_div(bnbp, NULL, bnap, &bnbase, bnctx); + bntp = bnap; + bnap = bnbp; + bnbp = bntp; + } + b58ceil = BN_get_word(bnap); + + if ((b58pow - (p - zero_prefix)) < 6) { + /* + * Do not allow the prefix to constrain the + * check value, this is ridiculous. + */ + printf("Prefix '%s' is too long\n", pfx); + goto out; + } + + BN_set_word(&bntmp2, b58pow - (p - zero_prefix)); + BN_exp(&bntmp, &bnbase, &bntmp2, bnctx); + BN_mul(bnlow, &bntmp, &bntarg, bnctx); + BN_sub(&bntmp2, &bntmp, BN_value_one()); + BN_add(bnhigh, bnlow, &bntmp2); + + if (b58top <= b58ceil) { + /* Fill out the upper range too */ + check_upper = 1; + bnlow2 = BN_new(); + bnhigh2 = BN_new(); + + BN_mul(bnlow2, bnlow, &bnbase, bnctx); + BN_mul(&bntmp2, bnhigh, &bnbase, bnctx); + BN_set_word(&bntmp, 57); + BN_add(bnhigh2, &bntmp2, &bntmp); + + /* + * Addresses above the ceiling will have one + * fewer "1" prefix in front than we require. + */ + if (BN_cmp(&bnceil, bnlow2) < 0) { + /* High prefix is above the ceiling */ + check_upper = 0; + BN_free(bnhigh2); + bnhigh2 = NULL; + BN_free(bnlow2); + bnlow2 = NULL; + } + else if (BN_cmp(&bnceil, bnhigh2) < 0) + /* High prefix is partly above the ceiling */ + BN_copy(bnhigh2, &bnceil); + + /* + * Addresses below the floor will have another + * "1" prefix in front instead of our target. + */ + if (BN_cmp(&bnfloor, bnhigh) >= 0) { + /* Low prefix is completely below the floor */ + assert(check_upper); + check_upper = 0; + BN_free(bnhigh); + bnhigh = bnhigh2; + bnhigh2 = NULL; + BN_free(bnlow); + bnlow = bnlow2; + bnlow2 = NULL; + } + else if (BN_cmp(&bnfloor, bnlow) > 0) { + /* Low prefix is partly below the floor */ + BN_copy(bnlow, &bnfloor); + } + } + + } else { + BN_copy(bnhigh, &bnceil); + BN_clear(bnlow); + } + + /* Limit the prefix to the address type */ + BN_clear(&bntmp); + BN_set_word(&bntmp, addrtype); + BN_lshift(&bntmp2, &bntmp, 192); + + if (check_upper) { + if (BN_cmp(&bntmp2, bnhigh2) > 0) { + check_upper = 0; + BN_free(bnhigh2); + bnhigh2 = NULL; + BN_free(bnlow2); + bnlow2 = NULL; + } + else if (BN_cmp(&bntmp2, bnlow2) > 0) + BN_copy(bnlow2, &bntmp2); + } + + if (BN_cmp(&bntmp2, bnhigh) > 0) { + if (!check_upper) + goto not_possible; + check_upper = 0; + BN_free(bnhigh); + bnhigh = bnhigh2; + bnhigh2 = NULL; + BN_free(bnlow); + bnlow = bnlow2; + bnlow2 = NULL; + } + else if (BN_cmp(&bntmp2, bnlow) > 0) { + BN_copy(bnlow, &bntmp2); + } + + BN_set_word(&bntmp, addrtype + 1); + BN_lshift(&bntmp2, &bntmp, 192); + + if (check_upper) { + if (BN_cmp(&bntmp2, bnlow2) < 0) { + check_upper = 0; + BN_free(bnhigh2); + bnhigh2 = NULL; + BN_free(bnlow2); + bnlow2 = NULL; + } + else if (BN_cmp(&bntmp2, bnhigh2) < 0) + BN_copy(bnlow2, &bntmp2); + } + + if (BN_cmp(&bntmp2, bnlow) < 0) { + if (!check_upper) + goto not_possible; + check_upper = 0; + BN_free(bnhigh); + bnhigh = bnhigh2; + bnhigh2 = NULL; + BN_free(bnlow); + bnlow = bnlow2; + bnlow2 = NULL; + } + else if (BN_cmp(&bntmp2, bnhigh) < 0) { + BN_copy(bnhigh, &bntmp2); + } + + /* Address ranges are complete */ + assert(check_upper || ((bnlow2 == NULL) && (bnhigh2 == NULL))); + result[0] = bnlow; + result[1] = bnhigh; + result[2] = bnlow2; + result[3] = bnhigh2; + bnlow = NULL; + bnhigh = NULL; + bnlow2 = NULL; + bnhigh2 = NULL; + ret = 0; + + if (0) { + not_possible: + printf("Prefix '%s' not possible\n", pfx); + ret = -2; + } + +out: + BN_clear_free(&bntarg); + BN_clear_free(&bnceil); + BN_clear_free(&bnfloor); + BN_clear_free(&bnbase); + BN_clear_free(&bntmp); + BN_clear_free(&bntmp2); + if (bnhigh) + BN_free(bnhigh); + if (bnlow) + BN_free(bnlow); + if (bnhigh2) + BN_free(bnhigh2); + if (bnlow2) + BN_free(bnlow2); + + return ret; +} + +/* + * AVL tree implementation + */ + +typedef enum { CENT = 1, LEFT = 0, RIGHT = 2 } avl_balance_t; + +typedef struct _avl_item_s { + struct _avl_item_s *ai_left, *ai_right, *ai_up; + avl_balance_t ai_balance; +#ifndef NDEBUG + int ai_indexed; +#endif +} avl_item_t; + +typedef struct _avl_root_s { + avl_item_t *ar_root; +} avl_root_t; + +static INLINE void +avl_root_init(avl_root_t *rootp) +{ + rootp->ar_root = NULL; +} + +static INLINE int +avl_root_empty(avl_root_t *rootp) +{ + return (rootp->ar_root == NULL) ? 1 : 0; +} + +static INLINE void +avl_item_init(avl_item_t *itemp) +{ + memset(itemp, 0, sizeof(*itemp)); + itemp->ai_balance = CENT; +} + +#define container_of(ptr, type, member) \ + (((type*) (((unsigned char *)ptr) - \ + (size_t)&(((type *)((unsigned char *)0))->member)))) + +#define avl_item_entry(ptr, type, member) \ + container_of(ptr, type, member) + + + +static INLINE void +_avl_rotate_ll(avl_root_t *rootp, avl_item_t *itemp) +{ + avl_item_t *tmp; + tmp = itemp->ai_left; + itemp->ai_left = tmp->ai_right; + if (itemp->ai_left) + itemp->ai_left->ai_up = itemp; + tmp->ai_right = itemp; + + if (itemp->ai_up) { + if (itemp->ai_up->ai_left == itemp) { + itemp->ai_up->ai_left = tmp; + } else { + assert(itemp->ai_up->ai_right == itemp); + itemp->ai_up->ai_right = tmp; + } + } else { + rootp->ar_root = tmp; + } + tmp->ai_up = itemp->ai_up; + itemp->ai_up = tmp; +} + +static INLINE void +_avl_rotate_lr(avl_root_t *rootp, avl_item_t *itemp) +{ + avl_item_t *rcp, *rlcp; + rcp = itemp->ai_left; + rlcp = rcp->ai_right; + if (itemp->ai_up) { + if (itemp == itemp->ai_up->ai_left) { + itemp->ai_up->ai_left = rlcp; + } else { + assert(itemp == itemp->ai_up->ai_right); + itemp->ai_up->ai_right = rlcp; + } + } else { + rootp->ar_root = rlcp; + } + rlcp->ai_up = itemp->ai_up; + rcp->ai_right = rlcp->ai_left; + if (rcp->ai_right) + rcp->ai_right->ai_up = rcp; + itemp->ai_left = rlcp->ai_right; + if (itemp->ai_left) + itemp->ai_left->ai_up = itemp; + rlcp->ai_left = rcp; + rlcp->ai_right = itemp; + rcp->ai_up = rlcp; + itemp->ai_up = rlcp; +} + +static INLINE void +_avl_rotate_rr(avl_root_t *rootp, avl_item_t *itemp) +{ + avl_item_t *tmp; + tmp = itemp->ai_right; + itemp->ai_right = tmp->ai_left; + if (itemp->ai_right) + itemp->ai_right->ai_up = itemp; + tmp->ai_left = itemp; + + if (itemp->ai_up) { + if (itemp->ai_up->ai_right == itemp) { + itemp->ai_up->ai_right = tmp; + } else { + assert(itemp->ai_up->ai_left == itemp); + itemp->ai_up->ai_left = tmp; + } + } else { + rootp->ar_root = tmp; + } + tmp->ai_up = itemp->ai_up; + itemp->ai_up = tmp; +} + +static INLINE void +_avl_rotate_rl(avl_root_t *rootp, avl_item_t *itemp) +{ + avl_item_t *rcp, *rlcp; + rcp = itemp->ai_right; + rlcp = rcp->ai_left; + if (itemp->ai_up) { + if (itemp == itemp->ai_up->ai_right) { + itemp->ai_up->ai_right = rlcp; + } else { + assert(itemp == itemp->ai_up->ai_left); + itemp->ai_up->ai_left = rlcp; + } + } else { + rootp->ar_root = rlcp; + } + rlcp->ai_up = itemp->ai_up; + rcp->ai_left = rlcp->ai_right; + if (rcp->ai_left) + rcp->ai_left->ai_up = rcp; + itemp->ai_right = rlcp->ai_left; + if (itemp->ai_right) + itemp->ai_right->ai_up = itemp; + rlcp->ai_right = rcp; + rlcp->ai_left = itemp; + rcp->ai_up = rlcp; + itemp->ai_up = rlcp; +} + +static void +avl_delete_fix(avl_root_t *rootp, avl_item_t *itemp, avl_item_t *parentp) +{ + avl_item_t *childp; + + if ((parentp->ai_left == NULL) && + (parentp->ai_right == NULL)) { + assert(itemp == NULL); + parentp->ai_balance = CENT; + itemp = parentp; + parentp = itemp->ai_up; + } + + while (parentp) { + if (itemp == parentp->ai_right) { + itemp = parentp->ai_left; + if (parentp->ai_balance == LEFT) { + /* Parent was left-heavy, now worse */ + if (itemp->ai_balance == LEFT) { + /* If left child is also + * left-heavy, LL fixes it. */ + _avl_rotate_ll(rootp, parentp); + itemp->ai_balance = CENT; + parentp->ai_balance = CENT; + parentp = itemp; + } else if (itemp->ai_balance == CENT) { + _avl_rotate_ll(rootp, parentp); + itemp->ai_balance = RIGHT; + parentp->ai_balance = LEFT; + break; + } else { + childp = itemp->ai_right; + _avl_rotate_lr(rootp, parentp); + itemp->ai_balance = CENT; + parentp->ai_balance = CENT; + if (childp->ai_balance == RIGHT) + itemp->ai_balance = LEFT; + if (childp->ai_balance == LEFT) + parentp->ai_balance = RIGHT; + childp->ai_balance = CENT; + parentp = childp; + } + } else if (parentp->ai_balance == CENT) { + parentp->ai_balance = LEFT; + break; + } else { + parentp->ai_balance = CENT; + } + + } else { + itemp = parentp->ai_right; + if (parentp->ai_balance == RIGHT) { + if (itemp->ai_balance == RIGHT) { + _avl_rotate_rr(rootp, parentp); + itemp->ai_balance = CENT; + parentp->ai_balance = CENT; + parentp = itemp; + } else if (itemp->ai_balance == CENT) { + _avl_rotate_rr(rootp, parentp); + itemp->ai_balance = LEFT; + parentp->ai_balance = RIGHT; + break; + } else { + childp = itemp->ai_left; + _avl_rotate_rl(rootp, parentp); + + itemp->ai_balance = CENT; + parentp->ai_balance = CENT; + if (childp->ai_balance == RIGHT) + parentp->ai_balance = LEFT; + if (childp->ai_balance == LEFT) + itemp->ai_balance = RIGHT; + childp->ai_balance = CENT; + parentp = childp; + } + } else if (parentp->ai_balance == CENT) { + parentp->ai_balance = RIGHT; + break; + } else { + parentp->ai_balance = CENT; + } + } + + itemp = parentp; + parentp = itemp->ai_up; + } +} + +static void +avl_insert_fix(avl_root_t *rootp, avl_item_t *itemp) +{ + avl_item_t *childp, *parentp = itemp->ai_up; + itemp->ai_left = itemp->ai_right = NULL; +#ifndef NDEBUG + assert(!itemp->ai_indexed); + itemp->ai_indexed = 1; +#endif + while (parentp) { + if (itemp == parentp->ai_left) { + if (parentp->ai_balance == LEFT) { + /* Parent was left-heavy, now worse */ + if (itemp->ai_balance == LEFT) { + /* If left child is also + * left-heavy, LL fixes it. */ + _avl_rotate_ll(rootp, parentp); + itemp->ai_balance = CENT; + parentp->ai_balance = CENT; + break; + } else { + assert(itemp->ai_balance != CENT); + childp = itemp->ai_right; + _avl_rotate_lr(rootp, parentp); + itemp->ai_balance = CENT; + parentp->ai_balance = CENT; + if (childp->ai_balance == RIGHT) + itemp->ai_balance = LEFT; + if (childp->ai_balance == LEFT) + parentp->ai_balance = RIGHT; + childp->ai_balance = CENT; + break; + } + } else if (parentp->ai_balance == CENT) { + parentp->ai_balance = LEFT; + } else { + parentp->ai_balance = CENT; + return; + } + } else { + if (parentp->ai_balance == RIGHT) { + if (itemp->ai_balance == RIGHT) { + _avl_rotate_rr(rootp, parentp); + itemp->ai_balance = CENT; + parentp->ai_balance = CENT; + break; + } else { + assert(itemp->ai_balance != CENT); + childp = itemp->ai_left; + _avl_rotate_rl(rootp, parentp); + itemp->ai_balance = CENT; + parentp->ai_balance = CENT; + if (childp->ai_balance == RIGHT) + parentp->ai_balance = LEFT; + if (childp->ai_balance == LEFT) + itemp->ai_balance = RIGHT; + childp->ai_balance = CENT; + break; + } + } else if (parentp->ai_balance == CENT) { + parentp->ai_balance = RIGHT; + } else { + parentp->ai_balance = CENT; + break; + } + } + + itemp = parentp; + parentp = itemp->ai_up; + } +} + +static INLINE avl_item_t * +avl_next(avl_item_t *itemp) +{ + if (itemp->ai_right) { + itemp = itemp->ai_right; + while (itemp->ai_left) + itemp = itemp->ai_left; + return itemp; + } + + while (itemp->ai_up && (itemp == itemp->ai_up->ai_right)) + itemp = itemp->ai_up; + + if (!itemp->ai_up) + return NULL; + + return itemp->ai_up; +} + +static void +avl_remove(avl_root_t *rootp, avl_item_t *itemp) +{ + avl_item_t *relocp, *replacep, *parentp = NULL; +#ifndef NDEBUG + assert(itemp->ai_indexed); + itemp->ai_indexed = 0; +#endif + /* If the item is directly replaceable, do it. */ + if ((itemp->ai_left == NULL) || (itemp->ai_right == NULL)) { + parentp = itemp->ai_up; + replacep = itemp->ai_left; + if (replacep == NULL) + replacep = itemp->ai_right; + if (replacep != NULL) + replacep->ai_up = parentp; + if (parentp == NULL) { + rootp->ar_root = replacep; + } else { + if (itemp == parentp->ai_left) + parentp->ai_left = replacep; + else + parentp->ai_right = replacep; + + avl_delete_fix(rootp, replacep, parentp); + } + return; + } + + /* + * Otherwise we do an indirect replacement with + * the item's leftmost right descendant. + */ + relocp = avl_next(itemp); + assert(relocp); + assert(relocp->ai_up != NULL); + assert(relocp->ai_left == NULL); + replacep = relocp->ai_right; + relocp->ai_left = itemp->ai_left; + if (relocp->ai_left != NULL) + relocp->ai_left->ai_up = relocp; + if (itemp->ai_up == NULL) + rootp->ar_root = relocp; + else { + if (itemp == itemp->ai_up->ai_left) + itemp->ai_up->ai_left = relocp; + else + itemp->ai_up->ai_right = relocp; + } + if (relocp == relocp->ai_up->ai_left) { + assert(relocp->ai_up != itemp); + relocp->ai_up->ai_left = replacep; + parentp = relocp->ai_up; + if (replacep != NULL) + replacep->ai_up = relocp->ai_up; + relocp->ai_right = itemp->ai_right; + } else { + assert(relocp->ai_up == itemp); + relocp->ai_right = replacep; + parentp = relocp; + } + if (relocp->ai_right != NULL) + relocp->ai_right->ai_up = relocp; + relocp->ai_up = itemp->ai_up; + relocp->ai_balance = itemp->ai_balance; + avl_delete_fix(rootp, replacep, parentp); +} + + + +/* + * Address prefix AVL tree node + */ + +const int vpk_nwords = (25 + sizeof(BN_ULONG) - 1) / sizeof(BN_ULONG); + +typedef struct _vg_prefix_s { + avl_item_t vp_item; + struct _vg_prefix_s *vp_sibling; + const char *vp_pattern; + BIGNUM *vp_low; + BIGNUM *vp_high; +} vg_prefix_t; + +static void +vg_prefix_free(vg_prefix_t *vp) +{ + if (vp->vp_low) + BN_free(vp->vp_low); + if (vp->vp_high) + BN_free(vp->vp_high); + free(vp); +} + +static vg_prefix_t * +vg_prefix_avl_search(avl_root_t *rootp, BIGNUM *targ) +{ + vg_prefix_t *vp; + avl_item_t *itemp = rootp->ar_root; + + while (itemp) { + vp = avl_item_entry(itemp, vg_prefix_t, vp_item); + if (BN_cmp(vp->vp_low, targ) > 0) { + itemp = itemp->ai_left; + } else { + if (BN_cmp(vp->vp_high, targ) < 0) { + itemp = itemp->ai_right; + } else + return vp; + } + } + return NULL; +} + +static vg_prefix_t * +vg_prefix_avl_insert(avl_root_t *rootp, vg_prefix_t *vpnew) +{ + vg_prefix_t *vp; + avl_item_t *itemp = NULL; + avl_item_t **ptrp = &rootp->ar_root; + while (*ptrp) { + itemp = *ptrp; + vp = avl_item_entry(itemp, vg_prefix_t, vp_item); + if (BN_cmp(vp->vp_low, vpnew->vp_high) > 0) { + ptrp = &itemp->ai_left; + } else { + if (BN_cmp(vp->vp_high, vpnew->vp_low) < 0) { + ptrp = &itemp->ai_right; + } else + return vp; + } + } + vpnew->vp_item.ai_up = itemp; + itemp = &vpnew->vp_item; + *ptrp = itemp; + avl_insert_fix(rootp, itemp); + return NULL; +} + +static vg_prefix_t * +vg_prefix_add(avl_root_t *rootp, const char *pattern, BIGNUM *low, BIGNUM *high) +{ + vg_prefix_t *vp, *vp2; + assert(BN_cmp(low, high) < 0); + vp = (vg_prefix_t *) malloc(sizeof(*vp)); + if (vp) { + avl_item_init(&vp->vp_item); + vp->vp_sibling = NULL; + vp->vp_pattern = pattern; + vp->vp_low = low; + vp->vp_high = high; + vp2 = vg_prefix_avl_insert(rootp, vp); + if (vp2 != NULL) { + printf("Prefix '%s' ignored, overlaps '%s'\n", + pattern, vp2->vp_pattern); + vg_prefix_free(vp); + vp = NULL; + } + } + return vp; +} + +static void +vg_prefix_delete(avl_root_t *rootp, vg_prefix_t *vp) +{ + vg_prefix_t *sibp, *delp; + + avl_remove(rootp, &vp->vp_item); + sibp = vp->vp_sibling; + while (sibp && sibp != vp) { + avl_remove(rootp, &sibp->vp_item); + delp = sibp; + sibp = sibp->vp_sibling; + vg_prefix_free(delp); + } + vg_prefix_free(vp); +} + +static vg_prefix_t * +vg_prefix_add_ranges(avl_root_t *rootp, const char *pattern, BIGNUM **ranges, + vg_prefix_t *master) +{ + vg_prefix_t *vp, *vp2 = NULL; + + assert(ranges[0]); + vp = vg_prefix_add(rootp, pattern, ranges[0], ranges[1]); + if (!vp) + return NULL; + + if (ranges[2]) { + vp2 = vg_prefix_add(rootp, pattern, ranges[2], ranges[3]); + if (!vp2) { + vg_prefix_delete(rootp, vp); + return NULL; + } + } + + if (!master) { + vp->vp_sibling = vp2; + if (vp2) + vp2->vp_sibling = vp; + } else if (vp2) { + vp->vp_sibling = vp2; + vp2->vp_sibling = (master->vp_sibling ? + master->vp_sibling : + master); + master->vp_sibling = vp; + } else { + vp->vp_sibling = (master->vp_sibling ? + master->vp_sibling : + master); + master->vp_sibling = vp; + } + return vp; +} + +static void +vg_prefix_range_sum(vg_prefix_t *vp, BIGNUM *result, BIGNUM *tmp1) +{ + vg_prefix_t *startp; + + startp = vp; + BN_clear(result); + do { + BN_sub(tmp1, vp->vp_high, vp->vp_low); + BN_add(result, result, tmp1); + vp = vp->vp_sibling; + } while (vp && (vp != startp)); +} + + +typedef struct _prefix_case_iter_s { + char ci_prefix[32]; + char ci_case_map[32]; + char ci_nbits; + int ci_value; +} prefix_case_iter_t; + +static const unsigned char b58_case_map[256] = { + 0, 0, 0, 0, 0, 0, 0, 0, 0, 0, 0, 0, 0, 0, 0, 0, + 0, 0, 0, 0, 0, 0, 0, 0, 0, 0, 0, 0, 0, 0, 0, 0, + 0, 0, 0, 0, 0, 0, 0, 0, 0, 0, 0, 0, 0, 0, 0, 0, + 0, 0, 0, 0, 0, 0, 0, 0, 0, 0, 0, 0, 0, 0, 0, 0, + 0, 1, 1, 1, 1, 1, 1, 1, 1, 0, 1, 1, 0, 1, 1, 0, + 1, 1, 1, 1, 1, 1, 1, 1, 1, 1, 1, 0, 0, 0, 0, 0, + 0, 1, 1, 1, 1, 1, 1, 1, 1, 0, 1, 1, 0, 1, 1, 0, + 1, 1, 1, 1, 1, 1, 1, 1, 1, 1, 1, 0, 0, 0, 0, 0, +}; + +static int +prefix_case_iter_init(prefix_case_iter_t *cip, const char *pfx) +{ + int i; + + cip->ci_nbits = 0; + cip->ci_value = 0; + for (i = 0; pfx[i]; i++) { + if (i > sizeof(cip->ci_prefix)) + return 0; + if (!b58_case_map[(int)pfx[i]]) { + cip->ci_prefix[i] = pfx[i]; + continue; + } + cip->ci_prefix[i] = pfx[i] | 0x20; + cip->ci_case_map[(int)cip->ci_nbits] = i; + cip->ci_nbits++; + } + cip->ci_prefix[i] = '\0'; + return 1; +} + +static int +prefix_case_iter_next(prefix_case_iter_t *cip) +{ + unsigned long val, max, mask; + int i, nbits; + + nbits = cip->ci_nbits; + max = (1UL << nbits) - 1; + val = cip->ci_value + 1; + if (val >= max) + return 0; + + for (i = 0, mask = 1; i < nbits; i++, mask <<= 1) { + if (val & mask) + cip->ci_prefix[(int)cip->ci_case_map[i]] &= 0xdf; + else + cip->ci_prefix[(int)cip->ci_case_map[i]] |= 0x20; + } + cip->ci_value = val; + return 1; +} + + +typedef struct _vg_prefix_context_s { + vg_context_t base; + avl_root_t vcp_avlroot; + BIGNUM vcp_difficulty; + int vcp_caseinsensitive; +} vg_prefix_context_t; + +static void +vg_prefix_context_free(vg_context_t *vcp) +{ + vg_prefix_context_t *vcpp = (vg_prefix_context_t *) vcp; + vg_prefix_t *vp; + unsigned long npfx_left = 0; + + while (!avl_root_empty(&vcpp->vcp_avlroot)) { + vp = avl_item_entry(vcpp->vcp_avlroot.ar_root, + vg_prefix_t, vp_item); + vg_prefix_delete(&vcpp->vcp_avlroot, vp); + npfx_left++; + } + + assert(npfx_left == vcpp->base.vc_npatterns); + BN_clear_free(&vcpp->vcp_difficulty); + free(vcpp); +} + +static void +vg_prefix_context_next_difficulty(vg_prefix_context_t *vcpp, + BIGNUM *bntmp, BIGNUM *bntmp2, BN_CTX *bnctx) +{ + char *dbuf; + + BN_clear(bntmp); + BN_set_bit(bntmp, 192); + BN_div(bntmp2, NULL, bntmp, &vcpp->vcp_difficulty, bnctx); + + dbuf = BN_bn2dec(bntmp2); + if (vcpp->base.vc_verbose > 0) { + if (vcpp->base.vc_npatterns > 1) + printf("Next match difficulty: %s (%ld prefixes)\n", + dbuf, vcpp->base.vc_npatterns); + else + printf("Difficulty: %s\n", dbuf); + } + vcpp->base.vc_chance = atof(dbuf); + OPENSSL_free(dbuf); +} + +static int +vg_prefix_context_add_patterns(vg_context_t *vcp, + char ** const patterns, int npatterns) +{ + vg_prefix_context_t *vcpp = (vg_prefix_context_t *) vcp; + prefix_case_iter_t caseiter; + vg_prefix_t *vp, *vp2; + BN_CTX *bnctx; + BIGNUM bntmp, bntmp2, bntmp3; + BIGNUM *ranges[4]; + int ret = 0; + int i, impossible = 0; + unsigned long npfx; + char *dbuf; + + bnctx = BN_CTX_new(); + BN_init(&bntmp); + BN_init(&bntmp2); + BN_init(&bntmp3); + + npfx = 0; + for (i = 0; i < npatterns; i++) { + if (!vcpp->vcp_caseinsensitive) { + vp = NULL; + ret = get_prefix_ranges(vcpp->base.vc_addrtype, + patterns[i], + ranges, bnctx); + if (!ret) { + vp = vg_prefix_add_ranges(&vcpp->vcp_avlroot, + patterns[i], + ranges, NULL); + } + + } else { + /* Case-enumerate the prefix */ + if (!prefix_case_iter_init(&caseiter, patterns[i])) { + printf("Prefix '%s' is too long\n", + patterns[i]); + continue; + } + + if (caseiter.ci_nbits > 16) { + printf("WARNING: Prefix '%s' has " + "2^%d case-varied derivitives\n", + patterns[i], caseiter.ci_nbits); + } + + vp = NULL; + do { + ret = get_prefix_ranges(vcpp->base.vc_addrtype, + caseiter.ci_prefix, + ranges, bnctx); + if (ret) + break; + vp2 = vg_prefix_add_ranges(&vcpp->vcp_avlroot, + patterns[i], + ranges, + vp); + if (!vp2) { + ret = -1; + break; + } + if (!vp) + vp = vp2; + + } while (prefix_case_iter_next(&caseiter)); + + if (ret && vp) { + vg_prefix_delete(&vcpp->vcp_avlroot, vp); + vp = NULL; + } + } + + if (ret == -2) + impossible++; + + if (!vp) + continue; + + npfx++; + + /* Determine the probability of finding a match */ + vg_prefix_range_sum(vp, &bntmp, &bntmp2); + BN_add(&bntmp2, &vcpp->vcp_difficulty, &bntmp); + BN_copy(&vcpp->vcp_difficulty, &bntmp2); + + if (vcp->vc_verbose > 1) { + BN_clear(&bntmp2); + BN_set_bit(&bntmp2, 192); + BN_div(&bntmp3, NULL, &bntmp2, &bntmp, bnctx); + + dbuf = BN_bn2dec(&bntmp3); + printf("Prefix difficulty: %20s %s\n", + dbuf, patterns[i]); + OPENSSL_free(dbuf); + } + } + + vcpp->base.vc_npatterns += npfx; + vcpp->base.vc_npatterns_start += npfx; + + if (!npfx && impossible) { + const char *ats = "bitcoin", *bw = "\"1\""; + switch (vcpp->base.vc_addrtype) { + case 111: + ats = "testnet"; + bw = "\"m\" or \"n\""; + break; + case 52: + ats = "namecoin"; + bw = "\"M\" or \"N\""; + break; + default: + break; + } + printf("Hint: valid %s addresses begin with %s\n", ats, bw); + } + + if (npfx) + vg_prefix_context_next_difficulty(vcpp, &bntmp, &bntmp2, bnctx); + + ret = (npfx != 0); + + BN_clear_free(&bntmp); + BN_clear_free(&bntmp2); + BN_clear_free(&bntmp3); + BN_CTX_free(bnctx); + return ret; +} + + +static int +vg_prefix_test(vg_exec_context_t *vxcp) +{ + vg_prefix_context_t *vcpp = (vg_prefix_context_t *) vxcp->vxc_vc; + vg_prefix_t *vp; + int res = 0; + + /* + * We constrain the prefix so that we can check for + * a match without generating the lower four byte + * check code. + */ + + BN_bin2bn(vxcp->vxc_binres, 25, &vxcp->vxc_bntarg); + +research: + vp = vg_prefix_avl_search(&vcpp->vcp_avlroot, &vxcp->vxc_bntarg); + if (vp) { + if (vg_exec_upgrade_lock(vxcp)) + goto research; + + vg_exec_context_consolidate_key(vxcp); + vg_output_match(&vcpp->base, vxcp->vxc_key, vp->vp_pattern); + + vcpp->base.vc_found++; + + if (vcpp->base.vc_remove_on_match) { + /* Subtract the range from the difficulty */ + vg_prefix_range_sum(vp, + &vxcp->vxc_bntarg, + &vxcp->vxc_bntmp); + BN_sub(&vxcp->vxc_bntmp, + &vcpp->vcp_difficulty, + &vxcp->vxc_bntarg); + BN_copy(&vcpp->vcp_difficulty, &vxcp->vxc_bntmp); + + vg_prefix_delete(&vcpp->vcp_avlroot,vp); + vcpp->base.vc_npatterns--; + + if (!avl_root_empty(&vcpp->vcp_avlroot)) + vg_prefix_context_next_difficulty( + vcpp, &vxcp->vxc_bntmp, + &vxcp->vxc_bntmp2, + vxcp->vxc_bnctx); + } + res = 1; + } + if (avl_root_empty(&vcpp->vcp_avlroot)) { + return 2; + } + return res; +} + +vg_context_t * +vg_prefix_context_new(int addrtype, int privtype, int caseinsensitive) +{ + vg_prefix_context_t *vcpp; + + vcpp = (vg_prefix_context_t *) malloc(sizeof(*vcpp)); + if (vcpp) { + vcpp->base.vc_addrtype = addrtype; + vcpp->base.vc_privtype = privtype; + vcpp->base.vc_npatterns = 0; + vcpp->base.vc_npatterns_start = 0; + vcpp->base.vc_found = 0; + vcpp->base.vc_chance = 0.0; + vcpp->base.vc_free = vg_prefix_context_free; + vcpp->base.vc_add_patterns = vg_prefix_context_add_patterns; + vcpp->base.vc_test = vg_prefix_test; + avl_root_init(&vcpp->vcp_avlroot); + BN_init(&vcpp->vcp_difficulty); + vcpp->vcp_caseinsensitive = caseinsensitive; + } + return &vcpp->base; +} + + + + +typedef struct _vg_regex_context_s { + vg_context_t base; + pcre **vcr_regex; + pcre_extra **vcr_regex_extra; + const char **vcr_regex_pat; + unsigned long vcr_nalloc; +} vg_regex_context_t; + +static int +vg_regex_context_add_patterns(vg_context_t *vcp, + char ** const patterns, int npatterns) +{ + vg_regex_context_t *vcrp = (vg_regex_context_t *) vcp; + const char *pcre_errptr; + int pcre_erroffset; + unsigned long i, nres, count; + void **mem; + + if (!npatterns) + return 1; + + if (npatterns > (vcrp->vcr_nalloc - vcrp->base.vc_npatterns)) { + count = npatterns + vcrp->base.vc_npatterns; + if (count < (2 * vcrp->vcr_nalloc)) { + count = (2 * vcrp->vcr_nalloc); + } + if (count < 16) { + count = 16; + } + mem = (void **) malloc(3 * count * sizeof(void*)); + if (!mem) + return 0; + + for (i = 0; i < vcrp->base.vc_npatterns; i++) { + mem[i] = vcrp->vcr_regex[i]; + mem[count + i] = vcrp->vcr_regex_extra[i]; + mem[(2 * count) + i] = (void *) vcrp->vcr_regex_pat[i]; + } + + if (vcrp->vcr_nalloc) + free(vcrp->vcr_regex); + vcrp->vcr_regex = (pcre **) mem; + vcrp->vcr_regex_extra = (pcre_extra **) &mem[count]; + vcrp->vcr_regex_pat = (const char **) &mem[2 * count]; + vcrp->vcr_nalloc = count; + } + + nres = vcrp->base.vc_npatterns; + for (i = 0; i < npatterns; i++) { + vcrp->vcr_regex[nres] = + pcre_compile(patterns[i], 0, + &pcre_errptr, &pcre_erroffset, NULL); + if (!vcrp->vcr_regex[nres]) { + const char *spaces = " "; + printf("%s\n", patterns[i]); + while (pcre_erroffset > 16) { + printf("%s", spaces); + pcre_erroffset -= 16; + } + if (pcre_erroffset > 0) + printf("%s", &spaces[16 - pcre_erroffset]); + printf("^\nRegex error: %s\n", pcre_errptr); + continue; + } + vcrp->vcr_regex_extra[nres] = + pcre_study(vcrp->vcr_regex[nres], 0, &pcre_errptr); + if (pcre_errptr) { + printf("Regex error: %s\n", pcre_errptr); + pcre_free(vcrp->vcr_regex[nres]); + continue; + } + vcrp->vcr_regex_pat[nres] = patterns[i]; + nres += 1; + } + + if (nres == vcrp->base.vc_npatterns) + return 0; + + vcrp->base.vc_npatterns_start += (nres - vcrp->base.vc_npatterns); + vcrp->base.vc_npatterns = nres; + return 1; +} + +static void +vg_regex_context_free(vg_context_t *vcp) +{ + vg_regex_context_t *vcrp = (vg_regex_context_t *) vcp; + int i; + for (i = 0; i < vcrp->base.vc_npatterns; i++) { + if (vcrp->vcr_regex_extra[i]) + pcre_free(vcrp->vcr_regex_extra[i]); + pcre_free(vcrp->vcr_regex[i]); + } + if (vcrp->vcr_nalloc) + free(vcrp->vcr_regex); + free(vcrp); +} + +static int +vg_regex_test(vg_exec_context_t *vxcp) +{ + vg_regex_context_t *vcrp = (vg_regex_context_t *) vxcp->vxc_vc; + + unsigned char hash1[32], hash2[32]; + int i, zpfx, p, d, nres, re_vec[9]; + char b58[40]; + BIGNUM bnrem; + BIGNUM *bn, *bndiv, *bnptmp; + int res = 0; + + pcre *re; + + BN_init(&bnrem); + + /* Hash the hash and write the four byte check code */ + SHA256(vxcp->vxc_binres, 21, hash1); + SHA256(hash1, sizeof(hash1), hash2); + memcpy(&vxcp->vxc_binres[21], hash2, 4); + + bn = &vxcp->vxc_bntmp; + bndiv = &vxcp->vxc_bntmp2; + + BN_bin2bn(vxcp->vxc_binres, 25, bn); + + /* Compute the complete encoded address */ + for (zpfx = 0; zpfx < 25 && vxcp->vxc_binres[zpfx] == 0; zpfx++); + p = sizeof(b58) - 1; + b58[p] = '\0'; + while (!BN_is_zero(bn)) { + BN_div(bndiv, &bnrem, bn, &vxcp->vxc_bnbase, vxcp->vxc_bnctx); + bnptmp = bn; + bn = bndiv; + bndiv = bnptmp; + d = BN_get_word(&bnrem); + b58[--p] = b58_alphabet[d]; + } + while (zpfx--) { + b58[--p] = b58_alphabet[0]; + } + + /* + * Run the regular expressions on it + * SLOW, runs in linear time with the number of REs + */ +restart_loop: + nres = vcrp->base.vc_npatterns; + if (!nres) { + res = 2; + goto out; + } + for (i = 0; i < nres; i++) { + d = pcre_exec(vcrp->vcr_regex[i], + vcrp->vcr_regex_extra[i], + &b58[p], (sizeof(b58) - 1) - p, 0, + 0, + re_vec, sizeof(re_vec)/sizeof(re_vec[0])); + + if (d <= 0) { + if (d != PCRE_ERROR_NOMATCH) { + printf("PCRE error: %d\n", d); + res = 2; + goto out; + } + continue; + } + + re = vcrp->vcr_regex[i]; + + if (vg_exec_upgrade_lock(vxcp) && + ((i >= vcrp->base.vc_npatterns) || + (vcrp->vcr_regex[i] != re))) + goto restart_loop; + + vg_exec_context_consolidate_key(vxcp); + vg_output_match(&vcrp->base, vxcp->vxc_key, + vcrp->vcr_regex_pat[i]); + vcrp->base.vc_found++; + + if (vcrp->base.vc_remove_on_match) { + pcre_free(vcrp->vcr_regex[i]); + if (vcrp->vcr_regex_extra[i]) + pcre_free(vcrp->vcr_regex_extra[i]); + nres -= 1; + vcrp->base.vc_npatterns = nres; + if (!nres) { + res = 2; + goto out; + } + vcrp->vcr_regex[i] = vcrp->vcr_regex[nres]; + vcrp->vcr_regex_extra[i] = + vcrp->vcr_regex_extra[nres]; + vcrp->vcr_regex_pat[i] = vcrp->vcr_regex_pat[nres]; + vcrp->base.vc_npatterns = nres; + } + res = 1; + } +out: + BN_clear_free(&bnrem); + return res; +} + +vg_context_t * +vg_regex_context_new(int addrtype, int privtype) +{ + vg_regex_context_t *vcrp; + + vcrp = (vg_regex_context_t *) malloc(sizeof(*vcrp)); + if (vcrp) { + vcrp->base.vc_addrtype = addrtype; + vcrp->base.vc_privtype = privtype; + vcrp->base.vc_npatterns = 0; + vcrp->base.vc_npatterns_start = 0; + vcrp->base.vc_found = 0; + vcrp->base.vc_chance = 0.0; + vcrp->base.vc_free = vg_regex_context_free; + vcrp->base.vc_add_patterns = vg_regex_context_add_patterns; + vcrp->base.vc_test = vg_regex_test; + vcrp->vcr_regex = NULL; + vcrp->vcr_nalloc = 0; + } + return &vcrp->base; +} + + +/* + * Pattern file reader + * Absolutely disgusting, unable to free the pattern list when it's done + */ + +int +vg_read_file(FILE *fp, char ***result, int *rescount) +{ + int ret = 1; + + char **patterns; + char *buf = NULL, *obuf, *pat; + const int blksize = 16*1024; + int nalloc = 16; + int npatterns = 0; + int count, pos; + + patterns = (char**) malloc(sizeof(char*) * nalloc); + count = 0; + pos = 0; + + while (1) { + obuf = buf; + buf = (char *) malloc(blksize); + if (!buf) { + ret = 0; + break; + } + if (pos < count) { + memcpy(buf, &obuf[pos], count - pos); + } + pos = count - pos; + count = fread(&buf[pos], 1, blksize - pos, fp); + if (count < 0) { + printf("Error reading file: %s\n", strerror(errno)); + ret = 0; + } + if (count <= 0) + break; + count += pos; + pat = buf; + + while (pos < count) { + if ((buf[pos] == '\r') || (buf[pos] == '\n')) { + buf[pos] = '\0'; + if (pat) { + if (npatterns == nalloc) { + nalloc *= 2; + patterns = (char**) + realloc(patterns, + sizeof(char*) * + nalloc); + } + patterns[npatterns] = pat; + npatterns++; + pat = NULL; + } + } + else if (!pat) { + pat = &buf[pos]; + } + pos++; + } + + pos = pat ? (pat - buf) : count; + } + + *result = patterns; + *rescount = npatterns; + + return ret; +} diff --git a/pattern.h b/pattern.h new file mode 100644 index 0000000..aa05033 --- /dev/null +++ b/pattern.h @@ -0,0 +1,110 @@ +/* + * Vanitygen, vanity bitcoin address generator + * Copyright (C) 2011 + * + * Vanitygen is free software: you can redistribute it and/or modify + * it under the terms of the GNU Affero General Public License as published by + * the Free Software Foundation, either version 3 of the License, or + * any later version. + * + * Vanitygen is distributed in the hope that it will be useful, + * but WITHOUT ANY WARRANTY; without even the implied warranty of + * MERCHANTABILITY or FITNESS FOR A PARTICULAR PURPOSE. See the + * GNU Affero General Public License for more details. + * + * You should have received a copy of the GNU Affero General Public License + * along with Vanitygen. If not, see . + */ + +#if !defined (__VG_PATTERN_H__) +#define __VG_PATTERN_H__ + +#include + +#include +#include + +#ifdef _WIN32 +#include "winglue.h" +#else +#define INLINE inline +#include +#include +#include +#include +#endif + + +extern void dumphex(const unsigned char *src, size_t len); +extern void dumpbn(const BIGNUM *bn); +extern void dumpbin(const uint8_t *data, size_t nbytes); + +typedef struct _vg_context_s vg_context_t; + +/* Context of one pattern-matching unit within the process */ +typedef struct _vg_exec_context_s { + vg_context_t *vxc_vc; + BN_CTX *vxc_bnctx; + EC_KEY *vxc_key; + EC_POINT *vxc_point; + int vxc_delta; + unsigned char vxc_binres[28]; + BIGNUM vxc_bntarg; + BIGNUM vxc_bnbase; + BIGNUM vxc_bntmp; + BIGNUM vxc_bntmp2; +} vg_exec_context_t; + +/* Init/cleanup for common execution context */ +extern int vg_exec_context_init(vg_context_t *vcp, vg_exec_context_t *vxcp); +extern void vg_exec_context_del(vg_exec_context_t *vxcp); +extern void vg_exec_context_consolidate_key(vg_exec_context_t *vxcp); + +/* Implementation-specific lock/unlock/consolidate */ +extern void vg_exec_downgrade_lock(vg_exec_context_t *vxcp); +extern int vg_exec_upgrade_lock(vg_exec_context_t *vxcp); + + +typedef void (*vg_free_func_t)(vg_context_t *); +typedef int (*vg_add_pattern_func_t)(vg_context_t *, + char ** const patterns, int npatterns); +typedef int (*vg_test_func_t)(vg_exec_context_t *); + +/* Application-level context, incl. parameters and global pattern store */ +struct _vg_context_s { + int vc_addrtype; + int vc_privtype; + unsigned long vc_npatterns; + unsigned long vc_npatterns_start; + unsigned long long vc_found; + double vc_chance; + const char *vc_result_file; + int vc_remove_on_match; + int vc_verbose; + vg_free_func_t vc_free; + vg_add_pattern_func_t vc_add_patterns; + vg_test_func_t vc_test; +}; + + +extern void vg_context_free(vg_context_t *vcp); +extern int vg_context_add_patterns(vg_context_t *vcp, + char ** const patterns, int npatterns); + + +extern vg_context_t *vg_prefix_context_new(int addrtype, int privtype, + int caseinsensitive); +extern vg_context_t *vg_regex_context_new(int addrtype, int privtype); + +extern int vg_output_timing(vg_context_t *vcp, int cycle, struct timeval *last); +extern void vg_output_match(vg_context_t *vcp, EC_KEY *pkey, + const char *pattern); +extern void vg_encode_address(EC_KEY *pkey, int addrtype, char *result); +extern void vg_encode_privkey(EC_KEY *pkey, int addrtype, char *result); + + + +extern int vg_read_file(FILE *fp, char ***result, int *rescount); + + +#endif /* !defined (__VG_PATTERN_H__) */ diff --git a/vanitygen.c b/vanitygen.c index d3cfc30..5fa5fc1 100644 --- a/vanitygen.c +++ b/vanitygen.c @@ -26,395 +26,21 @@ #include #include #include -#include #include #include -#include - -#ifndef _WIN32 -#define INLINE inline -#include -#include -#include -#include -#else -#include "winglue.c" -#endif - -const char *version = "0.12"; -const int debug = 0; -int verbose = 1; -int remove_on_match = 1; -const char *result_file = NULL; - -const char *b58_alphabet = "123456789ABCDEFGHJKLMNPQRSTUVWXYZabcdefghijkmnopqrstuvwxyz"; - -void -encode_b58_check(void *buf, size_t len, char *result) -{ - unsigned char hash1[32]; - unsigned char hash2[32]; - - int d, p; - - BN_CTX *bnctx; - BIGNUM *bn, *bndiv, *bntmp; - BIGNUM bna, bnb, bnbase, bnrem; - unsigned char *binres; - int brlen, zpfx; - - bnctx = BN_CTX_new(); - BN_init(&bna); - BN_init(&bnb); - BN_init(&bnbase); - BN_init(&bnrem); - BN_set_word(&bnbase, 58); - - bn = &bna; - bndiv = &bnb; - - brlen = (2 * len) + 4; - binres = (unsigned char*) malloc(brlen); - memcpy(binres, buf, len); - - SHA256(binres, len, hash1); - SHA256(hash1, sizeof(hash1), hash2); - memcpy(&binres[len], hash2, 4); - - BN_bin2bn(binres, len + 4, bn); - - for (zpfx = 0; zpfx < (len + 4) && binres[zpfx] == 0; zpfx++); - - p = brlen; - while (!BN_is_zero(bn)) { - BN_div(bndiv, &bnrem, bn, &bnbase, bnctx); - bntmp = bn; - bn = bndiv; - bndiv = bntmp; - d = BN_get_word(&bnrem); - binres[--p] = b58_alphabet[d]; - } - - while (zpfx--) { - binres[--p] = b58_alphabet[0]; - } - - memcpy(result, &binres[p], brlen - p); - result[brlen - p] = '\0'; - - free(binres); - BN_clear_free(&bna); - BN_clear_free(&bnb); - BN_clear_free(&bnbase); - BN_clear_free(&bnrem); - BN_CTX_free(bnctx); -} - -void -encode_address(EC_KEY *pkey, int addrtype, char *result) -{ - unsigned char eckey_buf[128], *pend; - unsigned char binres[21] = {0,}; - unsigned char hash1[32]; - - pend = eckey_buf; - - i2o_ECPublicKey(pkey, &pend); - - binres[0] = addrtype; - SHA256(eckey_buf, pend - eckey_buf, hash1); - RIPEMD160(hash1, sizeof(hash1), &binres[1]); - - encode_b58_check(binres, sizeof(binres), result); -} - -void -encode_privkey(EC_KEY *pkey, int addrtype, char *result) -{ - unsigned char eckey_buf[128]; - const BIGNUM *bn; - int nbytes; - - bn = EC_KEY_get0_private_key(pkey); - - eckey_buf[0] = addrtype; - nbytes = BN_bn2bin(bn, &eckey_buf[1]); - - encode_b58_check(eckey_buf, nbytes + 1, result); -} - - -void -dumphex(const unsigned char *src, size_t len) -{ - size_t i; - for (i = 0; i < len; i++) { - printf("%02x", src[i]); - } - printf("\n"); -} - -void -dumpbn(const BIGNUM *bn) -{ - char *buf; - buf = BN_bn2hex(bn); - printf("%s\n", buf ? buf : "0"); - if (buf) - OPENSSL_free(buf); -} +#include "pattern.h" +const char *version = "0.13"; typedef struct _vg_thread_context_s { - BN_CTX *vct_bnctx; - EC_KEY *vct_key; - EC_POINT *vct_point; - int vct_delta; - int vct_flags; - unsigned char vct_binres[25]; - BIGNUM vct_bntarg; - BIGNUM vct_bnbase; - BIGNUM vct_bntmp; - BIGNUM vct_bntmp2; - - struct _vg_thread_context_s *vct_next; - int vct_mode; - int vct_stop; + vg_exec_context_t base; + struct _vg_thread_context_s *vt_next; + int vt_mode; + int vt_stop; } vg_thread_context_t; -struct _vg_context_s; - -typedef int (*vg_test_func_t)(struct _vg_context_s *, - vg_thread_context_t *); - -typedef struct _vg_context_s { - int vc_addrtype; - int vc_privtype; - unsigned long vc_npatterns; - unsigned long vc_npatterns_start; - unsigned long long vc_found; - double vc_chance; - vg_test_func_t vc_test; -} vg_context_t; - - -typedef struct _timing_info_s { - struct _timing_info_s *ti_next; - pthread_t ti_thread; - unsigned long ti_last_rate; -} timing_info_t; - -int -output_timing(int cycle, struct timeval *last, vg_context_t *vcp) -{ - static unsigned long long total = 0, prevfound = 0, sincelast = 0; - static pthread_mutex_t mutex = PTHREAD_MUTEX_INITIALIZER; - static timing_info_t *timing_head = NULL; - - pthread_t me; - struct timeval tvnow, tv; - timing_info_t *tip, *mytip; - unsigned long long rate, myrate; - double count, prob, time, targ; - char linebuf[80]; - char *unit; - int rem, p, i; - - const double targs[] = { 0.5, 0.75, 0.8, 0.9, 0.95, 1.0 }; - - /* Compute the rate */ - gettimeofday(&tvnow, NULL); - timersub(&tvnow, last, &tv); - memcpy(last, &tvnow, sizeof(*last)); - myrate = tv.tv_usec + (1000000 * tv.tv_sec); - myrate = (1000000ULL * cycle) / myrate; - - pthread_mutex_lock(&mutex); - me = pthread_self(); - rate = myrate; - for (tip = timing_head, mytip = NULL; tip != NULL; tip = tip->ti_next) { - if (pthread_equal(tip->ti_thread, me)) { - mytip = tip; - tip->ti_last_rate = myrate; - } else - rate += tip->ti_last_rate; - } - if (!mytip) { - mytip = (timing_info_t *) malloc(sizeof(*tip)); - mytip->ti_next = timing_head; - mytip->ti_thread = me; - timing_head = mytip; - mytip->ti_last_rate = myrate; - } - - total += cycle; - if (prevfound != vcp->vc_found) { - prevfound = vcp->vc_found; - sincelast = 0; - } - sincelast += cycle; - count = sincelast; - - if (mytip != timing_head) { - pthread_mutex_unlock(&mutex); - return myrate; - } - pthread_mutex_unlock(&mutex); - - targ = rate; - unit = "key/s"; - if (targ > 1000) { - unit = "Kkey/s"; - targ /= 1000.0; - if (targ > 1000) { - unit = "Mkey/s"; - targ /= 1000.0; - } - } - - rem = sizeof(linebuf); - p = snprintf(linebuf, rem, "[%.2f %s][total %lld]", - targ, unit, total); - assert(p > 0); - rem -= p; - if (rem < 0) - rem = 0; - - if (vcp->vc_chance >= 1.0) { - prob = 1.0f - exp(-count/vcp->vc_chance); - - if (prob <= 0.999) { - p = snprintf(&linebuf[p], rem, "[Prob %.1f%%]", - prob * 100); - assert(p > 0); - rem -= p; - if (rem < 0) - rem = 0; - p = sizeof(linebuf) - rem; - } - - for (i = 0; i < sizeof(targs)/sizeof(targs[0]); i++) { - targ = targs[i]; - if ((targ < 1.0) && (prob <= targ)) - break; - } - - if (targ < 1.0) { - time = ((-vcp->vc_chance * log(1.0 - targ)) - count) / - rate; - unit = "s"; - if (time > 60) { - time /= 60; - unit = "min"; - if (time > 60) { - time /= 60; - unit = "h"; - if (time > 24) { - time /= 24; - unit = "d"; - if (time > 365) { - time /= 365; - unit = "y"; - } - } - } - } - - if (time > 1000000) { - p = snprintf(&linebuf[p], rem, - "[%d%% in %e%s]", - (int) (100 * targ), time, unit); - } else { - p = snprintf(&linebuf[p], rem, - "[%d%% in %.1f%s]", - (int) (100 * targ), time, unit); - } - assert(p > 0); - rem -= p; - if (rem < 0) - rem = 0; - p = sizeof(linebuf) - rem; - } - } - - if (vcp->vc_found) { - if (remove_on_match) - p = snprintf(&linebuf[p], rem, "[Found %lld/%ld]", - vcp->vc_found, vcp->vc_npatterns_start); - else - p = snprintf(&linebuf[p], rem, "[Found %lld]", - vcp->vc_found); - assert(p > 0); - rem -= p; - if (rem < 0) - rem = 0; - } - - if (rem) { - memset(&linebuf[sizeof(linebuf)-rem], 0x20, rem); - linebuf[sizeof(linebuf)-1] = '\0'; - } - printf("\r%s", linebuf); - fflush(stdout); - return myrate; -} - -void -output_match(EC_KEY *pkey, const char *pattern, vg_context_t *vcp) -{ - unsigned char key_buf[512], *pend; - char addr_buf[64]; - char privkey_buf[128]; - int len; - - assert(EC_KEY_check_key(pkey)); - encode_address(pkey, vcp->vc_addrtype, addr_buf); - encode_privkey(pkey, vcp->vc_privtype, privkey_buf); - - if (!result_file || (verbose > 0)) { - printf("\r%79s\rPattern: %s\n", "", pattern); - } - - if (verbose > 0) { - if (verbose > 1) { - /* Hexadecimal OpenSSL notation */ - pend = key_buf; - len = i2o_ECPublicKey(pkey, &pend); - printf("Pubkey (ASN1): "); - dumphex(key_buf, len); - pend = key_buf; - len = i2d_ECPrivateKey(pkey, &pend); - printf("Privkey (ASN1): "); - dumphex(key_buf, len); - } - - } - - if (!result_file || (verbose > 0)) { - printf("Address: %s\n" - "Privkey: %s\n", - addr_buf, privkey_buf); - } - - if (result_file) { - FILE *fp = fopen(result_file, "a"); - if (!fp) { - printf("ERROR: could not open result file: %s\n", - strerror(errno)); - } else { - fprintf(fp, - "Pattern: %s\n" - "Address: %s\n" - "Privkey: %s\n", - pattern, addr_buf, privkey_buf); - fclose(fp); - } - } -} - - /* * To synchronize pattern lists, we use a special shared-exclusive lock * geared toward being held in shared mode 99.9% of the time. @@ -428,1799 +54,329 @@ static vg_thread_context_t *vg_threads = NULL; static int vg_thread_excl = 0; void -__vg_thread_yield(vg_thread_context_t *vctp) +__vg_thread_yield(vg_thread_context_t *vtcp) { - vctp->vct_mode = 0; + vtcp->vt_mode = 0; while (vg_thread_excl) { - if (vctp->vct_stop) { + if (vtcp->vt_stop) { assert(vg_thread_excl); - vctp->vct_stop = 0; + vtcp->vt_stop = 0; pthread_cond_signal(&vg_thread_upcond); } pthread_cond_wait(&vg_thread_rdcond, &vg_thread_lock); } - assert(!vctp->vct_stop); - assert(!vctp->vct_mode); - vctp->vct_mode = 1; -} - -void -vg_thread_downgrade_lock(vg_thread_context_t *vctp) -{ - pthread_mutex_lock(&vg_thread_lock); - - assert(vctp->vct_mode == 2); - assert(!vctp->vct_stop); - if (!--vg_thread_excl) { - vctp->vct_mode = 1; - pthread_cond_broadcast(&vg_thread_rdcond); - pthread_mutex_unlock(&vg_thread_lock); - return; - } - pthread_cond_signal(&vg_thread_wrcond); - __vg_thread_yield(vctp); - pthread_mutex_unlock(&vg_thread_lock); -} - -void -vg_thread_init_lock(vg_thread_context_t *vctp) -{ - vctp->vct_mode = 0; - vctp->vct_stop = 0; - - pthread_mutex_lock(&vg_thread_lock); - vctp->vct_next = vg_threads; - vg_threads = vctp; - __vg_thread_yield(vctp); - pthread_mutex_unlock(&vg_thread_lock); -} - -void -vg_thread_del_lock(vg_thread_context_t *vctp) -{ - vg_thread_context_t *tp, **pprev; - - if (vctp->vct_mode == 2) - vg_thread_downgrade_lock(vctp); - - pthread_mutex_lock(&vg_thread_lock); - assert(vctp->vct_mode == 1); - vctp->vct_mode = 0; - - for (pprev = &vg_threads, tp = *pprev; - (tp != vctp) && (tp != NULL); - pprev = &tp->vct_next, tp = *pprev); - - assert(tp == vctp); - *pprev = tp->vct_next; - - if (tp->vct_stop) - pthread_cond_signal(&vg_thread_upcond); - - pthread_mutex_unlock(&vg_thread_lock); -} - -int -vg_thread_upgrade_lock(vg_thread_context_t *vctp) -{ - vg_thread_context_t *tp; - - if (vctp->vct_mode == 2) - return 0; - - pthread_mutex_lock(&vg_thread_lock); - - assert(vctp->vct_mode == 1); - vctp->vct_mode = 0; - - if (vg_thread_excl++) { - assert(vctp->vct_stop); - vctp->vct_stop = 0; - pthread_cond_signal(&vg_thread_upcond); - pthread_cond_wait(&vg_thread_wrcond, &vg_thread_lock); - - for (tp = vg_threads; tp != NULL; tp = tp->vct_next) { - assert(!tp->vct_mode); - assert(!tp->vct_stop); - } - - } else { - for (tp = vg_threads; tp != NULL; tp = tp->vct_next) { - if (tp->vct_mode) { - assert(tp->vct_mode != 2); - tp->vct_stop = 1; - } - } - - do { - for (tp = vg_threads; tp != NULL; tp = tp->vct_next) { - if (tp->vct_mode) { - assert(tp->vct_mode != 2); - pthread_cond_wait(&vg_thread_upcond, - &vg_thread_lock); - break; - } - } - } while (tp); - } - - vctp->vct_mode = 2; - pthread_mutex_unlock(&vg_thread_lock); - return 1; -} - -void -vg_thread_yield(vg_thread_context_t *vctp) -{ - if (vctp->vct_mode == 2) - vg_thread_downgrade_lock(vctp); - - else if (vctp->vct_stop) { - assert(vctp->vct_mode == 1); - pthread_mutex_lock(&vg_thread_lock); - __vg_thread_yield(vctp); - pthread_mutex_unlock(&vg_thread_lock); - } - - assert(vctp->vct_mode == 1); -} - - -void -vg_thread_consolidate_key(vg_thread_context_t *vctp) -{ - if (vctp->vct_delta) { - BN_clear(&vctp->vct_bntmp); - BN_set_word(&vctp->vct_bntmp, vctp->vct_delta); - BN_add(&vctp->vct_bntmp2, - EC_KEY_get0_private_key(vctp->vct_key), - &vctp->vct_bntmp); - EC_KEY_set_private_key(vctp->vct_key, &vctp->vct_bntmp2); - EC_KEY_set_public_key(vctp->vct_key, vctp->vct_point); - vctp->vct_delta = 0; - } -} - -/* - * Address search thread main loop - */ - -void * -vg_thread_loop(vg_context_t *vcp) -{ - unsigned char eckey_buf[128]; - unsigned char hash1[32]; - - int i, c, len, output_interval; - - const BN_ULONG rekey_max = 10000000; - BN_ULONG npoints, rekey_at, nbatch; - - EC_KEY *pkey = NULL; - const EC_GROUP *pgroup; - const EC_POINT *pgen; - const int ptarraysize = 256; - EC_POINT *ppnt[ptarraysize]; - EC_POINT *pbatchinc; - - vg_test_func_t test_func = vcp->vc_test; - vg_thread_context_t ctx; - - struct timeval tvstart; - - - memset(ctx.vct_binres, 0, sizeof(ctx.vct_binres)); - - ctx.vct_bnctx = BN_CTX_new(); - BN_init(&ctx.vct_bntarg); - BN_init(&ctx.vct_bnbase); - BN_init(&ctx.vct_bntmp); - BN_init(&ctx.vct_bntmp2); - - BN_set_word(&ctx.vct_bnbase, 58); - - pthread_mutex_lock(&vg_thread_lock); - pkey = EC_KEY_new_by_curve_name(NID_secp256k1); - assert(pkey); - pgroup = EC_KEY_get0_group(pkey); - assert(pgroup); - pgen = EC_GROUP_get0_generator(pgroup); - assert(pgen); - - EC_KEY_precompute_mult(pkey, ctx.vct_bnctx); - - pthread_mutex_unlock(&vg_thread_lock); - - for (i = 0; i < ptarraysize; i++) { - ppnt[i] = EC_POINT_new(pgroup); - if (!ppnt[i]) { - printf("ERROR: out of memory?\n"); - exit(1); - } - } - pbatchinc = EC_POINT_new(pgroup); - if (!pbatchinc) { - printf("ERROR: out of memory?\n"); - exit(1); - } - - BN_set_word(&ctx.vct_bntmp, ptarraysize); - EC_POINT_mul(pgroup, pbatchinc, &ctx.vct_bntmp, NULL, NULL, - ctx.vct_bnctx); - EC_POINT_make_affine(pgroup, pbatchinc, ctx.vct_bnctx); - - vg_thread_init_lock(&ctx); - - npoints = 0; - rekey_at = 0; - nbatch = 0; - ctx.vct_flags = 0; - ctx.vct_key = pkey; - ctx.vct_binres[0] = vcp->vc_addrtype; - c = 0; - output_interval = 1000; - gettimeofday(&tvstart, NULL); - - while (1) { - if (++npoints >= rekey_at) { - pthread_mutex_lock(&vg_thread_lock); - /* Generate a new random private key */ - EC_KEY_generate_key(pkey); - npoints = 0; - - /* Determine rekey interval */ - EC_GROUP_get_order(pgroup, &ctx.vct_bntmp, - ctx.vct_bnctx); - BN_sub(&ctx.vct_bntmp2, - &ctx.vct_bntmp, - EC_KEY_get0_private_key(pkey)); - rekey_at = BN_get_word(&ctx.vct_bntmp2); - if ((rekey_at == BN_MASK2) || (rekey_at > rekey_max)) - rekey_at = rekey_max; - assert(rekey_at > 0); - - EC_POINT_copy(ppnt[0], EC_KEY_get0_public_key(pkey)); - pthread_mutex_unlock(&vg_thread_lock); - - npoints++; - ctx.vct_delta = 0; - - for (nbatch = 1; - (nbatch < ptarraysize) && (npoints < rekey_at); - nbatch++, npoints++) { - EC_POINT_add(pgroup, - ppnt[nbatch], - ppnt[nbatch-1], - pgen, ctx.vct_bnctx); - } - - } else { - /* - * Common case - * - * EC_POINT_add() can skip a few multiplies if - * one or both inputs are affine (Z_is_one). - * This is the case for every point in ppnt, as - * well as pbatchinc. - */ - assert(nbatch == ptarraysize); - for (nbatch = 0; - (nbatch < ptarraysize) && (npoints < rekey_at); - nbatch++, npoints++) { - EC_POINT_add(pgroup, - ppnt[nbatch], - ppnt[nbatch], - pbatchinc, - ctx.vct_bnctx); - } - } - - /* - * The single most expensive operation performed in this - * loop is modular inversion of ppnt->Z. There is an - * algorithm implemented in OpenSSL to do batched inversion - * that only does one actual BN_mod_inverse(), and saves - * a _lot_ of time. - * - * To take advantage of this, we batch up a few points, - * and feed them to EC_POINTs_make_affine() below. - */ - - EC_POINTs_make_affine(pgroup, nbatch, ppnt, ctx.vct_bnctx); - - for (i = 0; i < nbatch; i++, ctx.vct_delta++) { - /* Hash the public key */ - len = EC_POINT_point2oct(pgroup, ppnt[i], - POINT_CONVERSION_UNCOMPRESSED, - eckey_buf, - sizeof(eckey_buf), - ctx.vct_bnctx); - - SHA256(eckey_buf, len, hash1); - RIPEMD160(hash1, sizeof(hash1), &ctx.vct_binres[1]); - - ctx.vct_point = ppnt[i]; - - switch (test_func(vcp, &ctx)) { - case 1: - npoints = 0; - rekey_at = 0; - i = nbatch; - break; - case 2: - goto out; - default: - break; - } - } - - c += (i + 1); - if (c >= output_interval) { - output_interval = output_timing(c, &tvstart, vcp); - c = 0; - } - - vg_thread_yield(&ctx); - } - -out: - vg_thread_del_lock(&ctx); - - BN_clear_free(&ctx.vct_bntarg); - BN_clear_free(&ctx.vct_bnbase); - BN_clear_free(&ctx.vct_bntmp); - BN_clear_free(&ctx.vct_bntmp2); - BN_CTX_free(ctx.vct_bnctx); - EC_KEY_free(pkey); - for (i = 0; i < ptarraysize; i++) - if (ppnt[i]) - EC_POINT_free(ppnt[i]); - if (pbatchinc) - EC_POINT_free(pbatchinc); - return NULL; -} - - - -const signed char b58_reverse_map[256] = { - -1, -1, -1, -1, -1, -1, -1, -1, -1, -1, -1, -1, -1, -1, -1, -1, - -1, -1, -1, -1, -1, -1, -1, -1, -1, -1, -1, -1, -1, -1, -1, -1, - -1, -1, -1, -1, -1, -1, -1, -1, -1, -1, -1, -1, -1, -1, -1, -1, - -1, 0, 1, 2, 3, 4, 5, 6, 7, 8, -1, -1, -1, -1, -1, -1, - -1, 9, 10, 11, 12, 13, 14, 15, 16, -1, 17, 18, 19, 20, 21, -1, - 22, 23, 24, 25, 26, 27, 28, 29, 30, 31, 32, -1, -1, -1, -1, -1, - -1, 33, 34, 35, 36, 37, 38, 39, 40, 41, 42, 43, -1, 44, 45, 46, - 47, 48, 49, 50, 51, 52, 53, 54, 55, 56, 57, -1, -1, -1, -1, -1, - -1, -1, -1, -1, -1, -1, -1, -1, -1, -1, -1, -1, -1, -1, -1, -1, - -1, -1, -1, -1, -1, -1, -1, -1, -1, -1, -1, -1, -1, -1, -1, -1, - -1, -1, -1, -1, -1, -1, -1, -1, -1, -1, -1, -1, -1, -1, -1, -1, - -1, -1, -1, -1, -1, -1, -1, -1, -1, -1, -1, -1, -1, -1, -1, -1, - -1, -1, -1, -1, -1, -1, -1, -1, -1, -1, -1, -1, -1, -1, -1, -1, - -1, -1, -1, -1, -1, -1, -1, -1, -1, -1, -1, -1, -1, -1, -1, -1, - -1, -1, -1, -1, -1, -1, -1, -1, -1, -1, -1, -1, -1, -1, -1, -1 -}; - -/* - * Find the bignum ranges that produce a given prefix. - */ -int -get_prefix_ranges(int addrtype, const char *pfx, BIGNUM **result, - BN_CTX *bnctx) -{ - int i, p, c; - int zero_prefix = 0; - int check_upper = 0; - int b58pow, b58ceil, b58top = 0; - int ret = 0; - - BIGNUM bntarg, bnceil, bnfloor; - BIGNUM bnbase; - BIGNUM *bnap, *bnbp, *bntp; - BIGNUM *bnhigh = NULL, *bnlow = NULL, *bnhigh2 = NULL, *bnlow2 = NULL; - BIGNUM bntmp, bntmp2; - - BN_init(&bntarg); - BN_init(&bnceil); - BN_init(&bnfloor); - BN_init(&bnbase); - BN_init(&bntmp); - BN_init(&bntmp2); - - BN_set_word(&bnbase, 58); - - p = strlen(pfx); - - for (i = 0; i < p; i++) { - c = b58_reverse_map[(int)pfx[i]]; - if (c == -1) { - printf("Invalid character '%c' in prefix '%s'\n", - pfx[i], pfx); - goto out; - } - if (i == zero_prefix) { - if (c == 0) { - /* Add another zero prefix */ - zero_prefix++; - if (zero_prefix > 19) { - printf("Prefix '%s' is too long\n", - pfx); - goto out; - } - continue; - } - - /* First non-zero character */ - b58top = c; - BN_set_word(&bntarg, c); - - } else { - BN_set_word(&bntmp2, c); - BN_mul(&bntmp, &bntarg, &bnbase, bnctx); - BN_add(&bntarg, &bntmp, &bntmp2); - } - } - - /* Power-of-two ceiling and floor values based on leading 1s */ - BN_clear(&bntmp); - BN_set_bit(&bntmp, 200 - (zero_prefix * 8)); - BN_sub(&bnceil, &bntmp, BN_value_one()); - BN_set_bit(&bnfloor, 192 - (zero_prefix * 8)); - - bnlow = BN_new(); - bnhigh = BN_new(); - - if (b58top) { - /* - * If a non-zero was given in the prefix, find the - * numeric boundaries of the prefix. - */ - - BN_copy(&bntmp, &bnceil); - bnap = &bntmp; - bnbp = &bntmp2; - b58pow = 0; - while (BN_cmp(bnap, &bnbase) > 0) { - b58pow++; - BN_div(bnbp, NULL, bnap, &bnbase, bnctx); - bntp = bnap; - bnap = bnbp; - bnbp = bntp; - } - b58ceil = BN_get_word(bnap); - - if ((b58pow - (p - zero_prefix)) < 6) { - /* - * Do not allow the prefix to constrain the - * check value, this is ridiculous. - */ - printf("Prefix '%s' is too long\n", pfx); - goto out; - } - - BN_set_word(&bntmp2, b58pow - (p - zero_prefix)); - BN_exp(&bntmp, &bnbase, &bntmp2, bnctx); - BN_mul(bnlow, &bntmp, &bntarg, bnctx); - BN_sub(&bntmp2, &bntmp, BN_value_one()); - BN_add(bnhigh, bnlow, &bntmp2); - - if (b58top <= b58ceil) { - /* Fill out the upper range too */ - check_upper = 1; - bnlow2 = BN_new(); - bnhigh2 = BN_new(); - - BN_mul(bnlow2, bnlow, &bnbase, bnctx); - BN_mul(&bntmp2, bnhigh, &bnbase, bnctx); - BN_set_word(&bntmp, 57); - BN_add(bnhigh2, &bntmp2, &bntmp); - - /* - * Addresses above the ceiling will have one - * fewer "1" prefix in front than we require. - */ - if (BN_cmp(&bnceil, bnlow2) < 0) { - /* High prefix is above the ceiling */ - check_upper = 0; - BN_free(bnhigh2); - bnhigh2 = NULL; - BN_free(bnlow2); - bnlow2 = NULL; - } - else if (BN_cmp(&bnceil, bnhigh2) < 0) - /* High prefix is partly above the ceiling */ - BN_copy(bnhigh2, &bnceil); - - /* - * Addresses below the floor will have another - * "1" prefix in front instead of our target. - */ - if (BN_cmp(&bnfloor, bnhigh) >= 0) { - /* Low prefix is completely below the floor */ - assert(check_upper); - check_upper = 0; - BN_free(bnhigh); - bnhigh = bnhigh2; - bnhigh2 = NULL; - BN_free(bnlow); - bnlow = bnlow2; - bnlow2 = NULL; - } - else if (BN_cmp(&bnfloor, bnlow) > 0) { - /* Low prefix is partly below the floor */ - BN_copy(bnlow, &bnfloor); - } - } - - } else { - BN_copy(bnhigh, &bnceil); - BN_clear(bnlow); - } - - /* Limit the prefix to the address type */ - BN_clear(&bntmp); - BN_set_word(&bntmp, addrtype); - BN_lshift(&bntmp2, &bntmp, 192); - - if (check_upper) { - if (BN_cmp(&bntmp2, bnhigh2) > 0) { - check_upper = 0; - BN_free(bnhigh2); - bnhigh2 = NULL; - BN_free(bnlow2); - bnlow2 = NULL; - } - else if (BN_cmp(&bntmp2, bnlow2) > 0) - BN_copy(bnlow2, &bntmp2); - } - - if (BN_cmp(&bntmp2, bnhigh) > 0) { - if (!check_upper) { - printf("Prefix '%s' not possible\n", pfx); - goto out; - } - check_upper = 0; - BN_free(bnhigh); - bnhigh = bnhigh2; - bnhigh2 = NULL; - BN_free(bnlow); - bnlow = bnlow2; - bnlow2 = NULL; - } - else if (BN_cmp(&bntmp2, bnlow) > 0) { - BN_copy(bnlow, &bntmp2); - } - - BN_set_word(&bntmp, addrtype + 1); - BN_lshift(&bntmp2, &bntmp, 192); - - if (check_upper) { - if (BN_cmp(&bntmp2, bnlow2) < 0) { - check_upper = 0; - BN_free(bnhigh2); - bnhigh2 = NULL; - BN_free(bnlow2); - bnlow2 = NULL; - } - else if (BN_cmp(&bntmp2, bnhigh2) < 0) - BN_copy(bnlow2, &bntmp2); - } - - if (BN_cmp(&bntmp2, bnlow) < 0) { - if (!check_upper) { - printf("Prefix '%s' not possible\n", pfx); - goto out; - } - check_upper = 0; - BN_free(bnhigh); - bnhigh = bnhigh2; - bnhigh2 = NULL; - BN_free(bnlow); - bnlow = bnlow2; - bnlow2 = NULL; - } - else if (BN_cmp(&bntmp2, bnhigh) < 0) { - BN_copy(bnhigh, &bntmp2); - } - - /* Address ranges are complete */ - assert(check_upper || ((bnlow2 == NULL) && (bnhigh2 == NULL))); - result[0] = bnlow; - result[1] = bnhigh; - result[2] = bnlow2; - result[3] = bnhigh2; - bnlow = NULL; - bnhigh = NULL; - bnlow2 = NULL; - bnhigh2 = NULL; - ret = 1; - -out: - BN_clear_free(&bntarg); - BN_clear_free(&bnceil); - BN_clear_free(&bnfloor); - BN_clear_free(&bnbase); - BN_clear_free(&bntmp); - BN_clear_free(&bntmp2); - if (bnhigh) - BN_free(bnhigh); - if (bnlow) - BN_free(bnlow); - if (bnhigh2) - BN_free(bnhigh2); - if (bnlow2) - BN_free(bnlow2); - - return ret; -} - -/* - * AVL tree implementation - */ - -typedef enum { CENT = 1, LEFT = 0, RIGHT = 2 } avl_balance_t; - -typedef struct _avl_item_s { - struct _avl_item_s *ai_left, *ai_right, *ai_up; - avl_balance_t ai_balance; -#ifndef NDEBUG - int ai_indexed; -#endif -} avl_item_t; - -typedef struct _avl_root_s { - avl_item_t *ar_root; -} avl_root_t; - -INLINE void -avl_root_init(avl_root_t *rootp) -{ - rootp->ar_root = NULL; -} - -INLINE int -avl_root_empty(avl_root_t *rootp) -{ - return (rootp->ar_root == NULL) ? 1 : 0; -} - -INLINE void -avl_item_init(avl_item_t *itemp) -{ - memset(itemp, 0, sizeof(*itemp)); - itemp->ai_balance = CENT; -} - -#define container_of(ptr, type, member) \ - (((type*) (((unsigned char *)ptr) - \ - (size_t)&(((type *)((unsigned char *)0))->member)))) - -#define avl_item_entry(ptr, type, member) \ - container_of(ptr, type, member) - - - -INLINE void -_avl_rotate_ll(avl_root_t *rootp, avl_item_t *itemp) -{ - avl_item_t *tmp; - tmp = itemp->ai_left; - itemp->ai_left = tmp->ai_right; - if (itemp->ai_left) - itemp->ai_left->ai_up = itemp; - tmp->ai_right = itemp; - - if (itemp->ai_up) { - if (itemp->ai_up->ai_left == itemp) { - itemp->ai_up->ai_left = tmp; - } else { - assert(itemp->ai_up->ai_right == itemp); - itemp->ai_up->ai_right = tmp; - } - } else { - rootp->ar_root = tmp; - } - tmp->ai_up = itemp->ai_up; - itemp->ai_up = tmp; -} - -INLINE void -_avl_rotate_lr(avl_root_t *rootp, avl_item_t *itemp) -{ - avl_item_t *rcp, *rlcp; - rcp = itemp->ai_left; - rlcp = rcp->ai_right; - if (itemp->ai_up) { - if (itemp == itemp->ai_up->ai_left) { - itemp->ai_up->ai_left = rlcp; - } else { - assert(itemp == itemp->ai_up->ai_right); - itemp->ai_up->ai_right = rlcp; - } - } else { - rootp->ar_root = rlcp; - } - rlcp->ai_up = itemp->ai_up; - rcp->ai_right = rlcp->ai_left; - if (rcp->ai_right) - rcp->ai_right->ai_up = rcp; - itemp->ai_left = rlcp->ai_right; - if (itemp->ai_left) - itemp->ai_left->ai_up = itemp; - rlcp->ai_left = rcp; - rlcp->ai_right = itemp; - rcp->ai_up = rlcp; - itemp->ai_up = rlcp; -} - -INLINE void -_avl_rotate_rr(avl_root_t *rootp, avl_item_t *itemp) -{ - avl_item_t *tmp; - tmp = itemp->ai_right; - itemp->ai_right = tmp->ai_left; - if (itemp->ai_right) - itemp->ai_right->ai_up = itemp; - tmp->ai_left = itemp; - - if (itemp->ai_up) { - if (itemp->ai_up->ai_right == itemp) { - itemp->ai_up->ai_right = tmp; - } else { - assert(itemp->ai_up->ai_left == itemp); - itemp->ai_up->ai_left = tmp; - } - } else { - rootp->ar_root = tmp; - } - tmp->ai_up = itemp->ai_up; - itemp->ai_up = tmp; -} - -INLINE void -_avl_rotate_rl(avl_root_t *rootp, avl_item_t *itemp) -{ - avl_item_t *rcp, *rlcp; - rcp = itemp->ai_right; - rlcp = rcp->ai_left; - if (itemp->ai_up) { - if (itemp == itemp->ai_up->ai_right) { - itemp->ai_up->ai_right = rlcp; - } else { - assert(itemp == itemp->ai_up->ai_left); - itemp->ai_up->ai_left = rlcp; - } - } else { - rootp->ar_root = rlcp; - } - rlcp->ai_up = itemp->ai_up; - rcp->ai_left = rlcp->ai_right; - if (rcp->ai_left) - rcp->ai_left->ai_up = rcp; - itemp->ai_right = rlcp->ai_left; - if (itemp->ai_right) - itemp->ai_right->ai_up = itemp; - rlcp->ai_right = rcp; - rlcp->ai_left = itemp; - rcp->ai_up = rlcp; - itemp->ai_up = rlcp; -} - -void -avl_delete_fix(avl_root_t *rootp, avl_item_t *itemp, avl_item_t *parentp) -{ - avl_item_t *childp; - - if ((parentp->ai_left == NULL) && - (parentp->ai_right == NULL)) { - assert(itemp == NULL); - parentp->ai_balance = CENT; - itemp = parentp; - parentp = itemp->ai_up; - } - - while (parentp) { - if (itemp == parentp->ai_right) { - itemp = parentp->ai_left; - if (parentp->ai_balance == LEFT) { - /* Parent was left-heavy, now worse */ - if (itemp->ai_balance == LEFT) { - /* If left child is also - * left-heavy, LL fixes it. */ - _avl_rotate_ll(rootp, parentp); - itemp->ai_balance = CENT; - parentp->ai_balance = CENT; - parentp = itemp; - } else if (itemp->ai_balance == CENT) { - _avl_rotate_ll(rootp, parentp); - itemp->ai_balance = RIGHT; - parentp->ai_balance = LEFT; - break; - } else { - childp = itemp->ai_right; - _avl_rotate_lr(rootp, parentp); - itemp->ai_balance = CENT; - parentp->ai_balance = CENT; - if (childp->ai_balance == RIGHT) - itemp->ai_balance = LEFT; - if (childp->ai_balance == LEFT) - parentp->ai_balance = RIGHT; - childp->ai_balance = CENT; - parentp = childp; - } - } else if (parentp->ai_balance == CENT) { - parentp->ai_balance = LEFT; - break; - } else { - parentp->ai_balance = CENT; - } - - } else { - itemp = parentp->ai_right; - if (parentp->ai_balance == RIGHT) { - if (itemp->ai_balance == RIGHT) { - _avl_rotate_rr(rootp, parentp); - itemp->ai_balance = CENT; - parentp->ai_balance = CENT; - parentp = itemp; - } else if (itemp->ai_balance == CENT) { - _avl_rotate_rr(rootp, parentp); - itemp->ai_balance = LEFT; - parentp->ai_balance = RIGHT; - break; - } else { - childp = itemp->ai_left; - _avl_rotate_rl(rootp, parentp); - - itemp->ai_balance = CENT; - parentp->ai_balance = CENT; - if (childp->ai_balance == RIGHT) - parentp->ai_balance = LEFT; - if (childp->ai_balance == LEFT) - itemp->ai_balance = RIGHT; - childp->ai_balance = CENT; - parentp = childp; - } - } else if (parentp->ai_balance == CENT) { - parentp->ai_balance = RIGHT; - break; - } else { - parentp->ai_balance = CENT; - } - } - - itemp = parentp; - parentp = itemp->ai_up; - } -} - -void -avl_insert_fix(avl_root_t *rootp, avl_item_t *itemp) -{ - avl_item_t *childp, *parentp = itemp->ai_up; - itemp->ai_left = itemp->ai_right = NULL; -#ifndef NDEBUG - assert(!itemp->ai_indexed); - itemp->ai_indexed = 1; -#endif - while (parentp) { - if (itemp == parentp->ai_left) { - if (parentp->ai_balance == LEFT) { - /* Parent was left-heavy, now worse */ - if (itemp->ai_balance == LEFT) { - /* If left child is also - * left-heavy, LL fixes it. */ - _avl_rotate_ll(rootp, parentp); - itemp->ai_balance = CENT; - parentp->ai_balance = CENT; - break; - } else { - assert(itemp->ai_balance != CENT); - childp = itemp->ai_right; - _avl_rotate_lr(rootp, parentp); - itemp->ai_balance = CENT; - parentp->ai_balance = CENT; - if (childp->ai_balance == RIGHT) - itemp->ai_balance = LEFT; - if (childp->ai_balance == LEFT) - parentp->ai_balance = RIGHT; - childp->ai_balance = CENT; - break; - } - } else if (parentp->ai_balance == CENT) { - parentp->ai_balance = LEFT; - } else { - parentp->ai_balance = CENT; - return; - } - } else { - if (parentp->ai_balance == RIGHT) { - if (itemp->ai_balance == RIGHT) { - _avl_rotate_rr(rootp, parentp); - itemp->ai_balance = CENT; - parentp->ai_balance = CENT; - break; - } else { - assert(itemp->ai_balance != CENT); - childp = itemp->ai_left; - _avl_rotate_rl(rootp, parentp); - itemp->ai_balance = CENT; - parentp->ai_balance = CENT; - if (childp->ai_balance == RIGHT) - parentp->ai_balance = LEFT; - if (childp->ai_balance == LEFT) - itemp->ai_balance = RIGHT; - childp->ai_balance = CENT; - break; - } - } else if (parentp->ai_balance == CENT) { - parentp->ai_balance = RIGHT; - } else { - parentp->ai_balance = CENT; - break; - } - } - - itemp = parentp; - parentp = itemp->ai_up; - } -} - -INLINE avl_item_t * -avl_next(avl_item_t *itemp) -{ - if (itemp->ai_right) { - itemp = itemp->ai_right; - while (itemp->ai_left) - itemp = itemp->ai_left; - return itemp; - } - - while (itemp->ai_up && (itemp == itemp->ai_up->ai_right)) - itemp = itemp->ai_up; - - if (!itemp->ai_up) - return NULL; - - return itemp->ai_up; -} - -void -avl_remove(avl_root_t *rootp, avl_item_t *itemp) -{ - avl_item_t *relocp, *replacep, *parentp = NULL; -#ifndef NDEBUG - assert(itemp->ai_indexed); - itemp->ai_indexed = 0; -#endif - /* If the item is directly replaceable, do it. */ - if ((itemp->ai_left == NULL) || (itemp->ai_right == NULL)) { - parentp = itemp->ai_up; - replacep = itemp->ai_left; - if (replacep == NULL) - replacep = itemp->ai_right; - if (replacep != NULL) - replacep->ai_up = parentp; - if (parentp == NULL) { - rootp->ar_root = replacep; - } else { - if (itemp == parentp->ai_left) - parentp->ai_left = replacep; - else - parentp->ai_right = replacep; - - avl_delete_fix(rootp, replacep, parentp); - } - return; - } - - /* - * Otherwise we do an indirect replacement with - * the item's leftmost right descendant. - */ - relocp = avl_next(itemp); - assert(relocp); - assert(relocp->ai_up != NULL); - assert(relocp->ai_left == NULL); - replacep = relocp->ai_right; - relocp->ai_left = itemp->ai_left; - if (relocp->ai_left != NULL) - relocp->ai_left->ai_up = relocp; - if (itemp->ai_up == NULL) - rootp->ar_root = relocp; - else { - if (itemp == itemp->ai_up->ai_left) - itemp->ai_up->ai_left = relocp; - else - itemp->ai_up->ai_right = relocp; - } - if (relocp == relocp->ai_up->ai_left) { - assert(relocp->ai_up != itemp); - relocp->ai_up->ai_left = replacep; - parentp = relocp->ai_up; - if (replacep != NULL) - replacep->ai_up = relocp->ai_up; - relocp->ai_right = itemp->ai_right; - } else { - assert(relocp->ai_up == itemp); - relocp->ai_right = replacep; - parentp = relocp; - } - if (relocp->ai_right != NULL) - relocp->ai_right->ai_up = relocp; - relocp->ai_up = itemp->ai_up; - relocp->ai_balance = itemp->ai_balance; - avl_delete_fix(rootp, replacep, parentp); -} - - - -/* - * Address prefix AVL tree node - */ - -typedef struct _vg_prefix_s { - avl_item_t vp_item; - struct _vg_prefix_s *vp_sibling; - const char *vp_pattern; - BIGNUM *vp_low; - BIGNUM *vp_high; -} vg_prefix_t; - -void -vg_prefix_free(vg_prefix_t *vp) -{ - if (vp->vp_low) - BN_free(vp->vp_low); - if (vp->vp_high) - BN_free(vp->vp_high); - free(vp); -} - -vg_prefix_t * -vg_prefix_avl_search(avl_root_t *rootp, BIGNUM *targ) -{ - vg_prefix_t *vp; - avl_item_t *itemp = rootp->ar_root; - - while (itemp) { - vp = avl_item_entry(itemp, vg_prefix_t, vp_item); - if (BN_cmp(vp->vp_low, targ) > 0) { - itemp = itemp->ai_left; - } else { - if (BN_cmp(vp->vp_high, targ) < 0) { - itemp = itemp->ai_right; - } else - return vp; - } - } - return NULL; -} - -vg_prefix_t * -vg_prefix_avl_insert(avl_root_t *rootp, vg_prefix_t *vpnew) -{ - vg_prefix_t *vp; - avl_item_t *itemp = NULL; - avl_item_t **ptrp = &rootp->ar_root; - while (*ptrp) { - itemp = *ptrp; - vp = avl_item_entry(itemp, vg_prefix_t, vp_item); - if (BN_cmp(vp->vp_low, vpnew->vp_high) > 0) { - ptrp = &itemp->ai_left; - } else { - if (BN_cmp(vp->vp_high, vpnew->vp_low) < 0) { - ptrp = &itemp->ai_right; - } else - return vp; - } - } - vpnew->vp_item.ai_up = itemp; - itemp = &vpnew->vp_item; - *ptrp = itemp; - avl_insert_fix(rootp, itemp); - return NULL; -} - -vg_prefix_t * -vg_prefix_add(avl_root_t *rootp, const char *pattern, BIGNUM *low, BIGNUM *high) -{ - vg_prefix_t *vp, *vp2; - assert(BN_cmp(low, high) < 0); - vp = (vg_prefix_t *) malloc(sizeof(*vp)); - if (vp) { - avl_item_init(&vp->vp_item); - vp->vp_sibling = NULL; - vp->vp_pattern = pattern; - vp->vp_low = low; - vp->vp_high = high; - vp2 = vg_prefix_avl_insert(rootp, vp); - if (vp2 != NULL) { - printf("Prefix '%s' ignored, overlaps '%s'\n", - pattern, vp2->vp_pattern); - vg_prefix_free(vp); - vp = NULL; - } - } - return vp; -} - -void -vg_prefix_delete(avl_root_t *rootp, vg_prefix_t *vp) -{ - vg_prefix_t *sibp, *delp; - - avl_remove(rootp, &vp->vp_item); - sibp = vp->vp_sibling; - while (sibp && sibp != vp) { - avl_remove(rootp, &sibp->vp_item); - delp = sibp; - sibp = sibp->vp_sibling; - vg_prefix_free(delp); - } - vg_prefix_free(vp); + assert(!vtcp->vt_stop); + assert(!vtcp->vt_mode); + vtcp->vt_mode = 1; } -vg_prefix_t * -vg_prefix_add_ranges(avl_root_t *rootp, const char *pattern, BIGNUM **ranges, - vg_prefix_t *master) +void +vg_thread_context_init(vg_context_t *vcp, vg_thread_context_t *vtcp) { - vg_prefix_t *vp, *vp2 = NULL; - - assert(ranges[0]); - vp = vg_prefix_add(rootp, pattern, ranges[0], ranges[1]); - if (!vp) - return NULL; - - if (ranges[2]) { - vp2 = vg_prefix_add(rootp, pattern, ranges[2], ranges[3]); - if (!vp2) { - vg_prefix_delete(rootp, vp); - return NULL; - } - } + vtcp->vt_mode = 0; + vtcp->vt_stop = 0; - if (!master) { - vp->vp_sibling = vp2; - if (vp2) - vp2->vp_sibling = vp; - } else if (vp2) { - vp->vp_sibling = vp2; - vp2->vp_sibling = (master->vp_sibling ? - master->vp_sibling : - master); - master->vp_sibling = vp; - } else { - vp->vp_sibling = (master->vp_sibling ? - master->vp_sibling : - master); - master->vp_sibling = vp; - } - return vp; + pthread_mutex_lock(&vg_thread_lock); + vg_exec_context_init(vcp, &vtcp->base); + vtcp->vt_next = vg_threads; + vg_threads = vtcp; + __vg_thread_yield(vtcp); + pthread_mutex_unlock(&vg_thread_lock); } void -vg_prefix_range_sum(vg_prefix_t *vp, BIGNUM *result, BIGNUM *tmp1, BIGNUM *tmp2) +vg_thread_context_del(vg_thread_context_t *vtcp) { - vg_prefix_t *startp; - - BIGNUM *bnptr = result; + vg_thread_context_t *tp, **pprev; - startp = vp; - BN_clear(result); - do { - BN_sub(tmp1, vp->vp_high, vp->vp_low); - if (bnptr == result) { - BN_add(tmp2, bnptr, tmp1); - bnptr = tmp2; - } else { - BN_add(result, bnptr, tmp1); - bnptr = result; - } - vp = vp->vp_sibling; - } while (vp && (vp != startp)); + if (vtcp->vt_mode == 2) + vg_exec_downgrade_lock(&vtcp->base); - if (bnptr != result) - BN_copy(result, bnptr); -} + pthread_mutex_lock(&vg_thread_lock); + assert(vtcp->vt_mode == 1); + vtcp->vt_mode = 0; + for (pprev = &vg_threads, tp = *pprev; + (tp != vtcp) && (tp != NULL); + pprev = &tp->vt_next, tp = *pprev); -typedef struct _prefix_case_iter_s { - char ci_prefix[32]; - char ci_case_map[32]; - char ci_nbits; - int ci_value; -} prefix_case_iter_t; - -const unsigned char b58_case_map[256] = { - 0, 0, 0, 0, 0, 0, 0, 0, 0, 0, 0, 0, 0, 0, 0, 0, - 0, 0, 0, 0, 0, 0, 0, 0, 0, 0, 0, 0, 0, 0, 0, 0, - 0, 0, 0, 0, 0, 0, 0, 0, 0, 0, 0, 0, 0, 0, 0, 0, - 0, 0, 0, 0, 0, 0, 0, 0, 0, 0, 0, 0, 0, 0, 0, 0, - 0, 1, 1, 1, 1, 1, 1, 1, 1, 0, 1, 1, 0, 1, 1, 0, - 1, 1, 1, 1, 1, 1, 1, 1, 1, 1, 1, 0, 0, 0, 0, 0, - 0, 1, 1, 1, 1, 1, 1, 1, 1, 0, 1, 1, 0, 1, 1, 0, - 1, 1, 1, 1, 1, 1, 1, 1, 1, 1, 1, 0, 0, 0, 0, 0, -}; + assert(tp == vtcp); + *pprev = tp->vt_next; -int -prefix_case_iter_init(prefix_case_iter_t *cip, const char *pfx) -{ - int i; + if (tp->vt_stop) + pthread_cond_signal(&vg_thread_upcond); - cip->ci_nbits = 0; - cip->ci_value = 0; - for (i = 0; pfx[i]; i++) { - if (i > sizeof(cip->ci_prefix)) - return 0; - if (!b58_case_map[(int)pfx[i]]) { - cip->ci_prefix[i] = pfx[i]; - continue; - } - cip->ci_prefix[i] = pfx[i] | 0x20; - cip->ci_case_map[(int)cip->ci_nbits] = i; - cip->ci_nbits++; - } - cip->ci_prefix[i] = '\0'; - return 1; + vg_exec_context_del(&vtcp->base); + pthread_mutex_unlock(&vg_thread_lock); } -int -prefix_case_iter_next(prefix_case_iter_t *cip) +void +vg_thread_yield(vg_thread_context_t *vtcp) { - unsigned long val, max, mask; - int i, nbits; + if (vtcp->vt_mode == 2) + vg_exec_downgrade_lock(&vtcp->base); - nbits = cip->ci_nbits; - max = (1UL << nbits) - 1; - val = cip->ci_value + 1; - if (val >= max) - return 0; - - for (i = 0, mask = 1; i < nbits; i++, mask <<= 1) { - if (val & mask) - cip->ci_prefix[(int)cip->ci_case_map[i]] &= 0xdf; - else - cip->ci_prefix[(int)cip->ci_case_map[i]] |= 0x20; + else if (vtcp->vt_stop) { + assert(vtcp->vt_mode == 1); + pthread_mutex_lock(&vg_thread_lock); + __vg_thread_yield(vtcp); + pthread_mutex_unlock(&vg_thread_lock); } - cip->ci_value = val; - return 1; -} + assert(vtcp->vt_mode == 1); +} -typedef struct _vg_prefix_context_s { - vg_context_t base; - avl_root_t vcp_avlroot; - BIGNUM vcp_difficulty; -} vg_prefix_context_t; - -void -vg_prefix_context_free(vg_prefix_context_t *vcpp) -{ - vg_prefix_t *vp; - unsigned long npfx_left = 0; - - while (!avl_root_empty(&vcpp->vcp_avlroot)) { - vp = avl_item_entry(vcpp->vcp_avlroot.ar_root, - vg_prefix_t, vp_item); - vg_prefix_delete(&vcpp->vcp_avlroot, vp); - npfx_left++; - } - assert(npfx_left == vcpp->base.vc_npatterns); - BN_clear_free(&vcpp->vcp_difficulty); - free(vcpp); -} void -vg_prefix_context_next_difficulty(vg_prefix_context_t *vcpp, - BIGNUM *bntmp, BIGNUM *bntmp2, BN_CTX *bnctx) +vg_exec_downgrade_lock(vg_exec_context_t *vxcp) { - char *dbuf; - - BN_clear(bntmp); - BN_set_bit(bntmp, 192); - BN_div(bntmp2, NULL, bntmp, &vcpp->vcp_difficulty, bnctx); + vg_thread_context_t *vtcp = (vg_thread_context_t *) vxcp; + pthread_mutex_lock(&vg_thread_lock); - dbuf = BN_bn2dec(bntmp2); - if (verbose > 0) { - if (vcpp->base.vc_npatterns > 1) - printf("Next match difficulty: %s (%ld prefixes)\n", - dbuf, vcpp->base.vc_npatterns); - else - printf("Difficulty: %s\n", dbuf); + assert(vtcp->vt_mode == 2); + assert(!vtcp->vt_stop); + if (!--vg_thread_excl) { + vtcp->vt_mode = 1; + pthread_cond_broadcast(&vg_thread_rdcond); + pthread_mutex_unlock(&vg_thread_lock); + return; } - vcpp->base.vc_chance = atof(dbuf); - OPENSSL_free(dbuf); + pthread_cond_signal(&vg_thread_wrcond); + __vg_thread_yield(vtcp); + pthread_mutex_unlock(&vg_thread_lock); } int -vg_prefix_context_add_patterns(vg_prefix_context_t *vcpp, - char ** const patterns, int npatterns, - int caseinsensitive) +vg_exec_upgrade_lock(vg_exec_context_t *vxcp) { - prefix_case_iter_t caseiter; - vg_prefix_t *vp, *vp2; - BN_CTX *bnctx; - BIGNUM bntmp, bntmp2, bntmp3; - BIGNUM *ranges[4]; - int ret = 0; - int i, fail; - unsigned long npfx; - char *dbuf; - - bnctx = BN_CTX_new(); - BN_init(&bntmp); - BN_init(&bntmp2); - BN_init(&bntmp3); - - npfx = 0; - fail = 0; - for (i = 0; i < npatterns; i++) { - if (!caseinsensitive) { - vp = NULL; - if (get_prefix_ranges(vcpp->base.vc_addrtype, - patterns[i], - ranges, bnctx)) { - vp = vg_prefix_add_ranges(&vcpp->vcp_avlroot, - patterns[i], - ranges, NULL); - } + vg_thread_context_t *vtcp = (vg_thread_context_t *) vxcp; + vg_thread_context_t *tp; - } else { - /* Case-enumerate the prefix */ - if (!prefix_case_iter_init(&caseiter, patterns[i])) { - printf("Prefix '%s' is too long\n", - patterns[i]); - continue; - } + if (vtcp->vt_mode == 2) + return 0; - if (caseiter.ci_nbits > 16) { - printf("WARNING: Prefix '%s' has " - "2^%d case-varied derivitives\n", - patterns[i], caseiter.ci_nbits); - } + pthread_mutex_lock(&vg_thread_lock); - vp = NULL; - fail = 0; - do { - if (!get_prefix_ranges(vcpp->base.vc_addrtype, - caseiter.ci_prefix, - ranges, bnctx)) { - fail = 1; - break; - } - vp2 = vg_prefix_add_ranges(&vcpp->vcp_avlroot, - patterns[i], - ranges, - vp); - if (!vp2) { - fail = 1; - break; - } - if (!vp) - vp = vp2; + assert(vtcp->vt_mode == 1); + vtcp->vt_mode = 0; - } while (prefix_case_iter_next(&caseiter)); + if (vg_thread_excl++) { + assert(vtcp->vt_stop); + vtcp->vt_stop = 0; + pthread_cond_signal(&vg_thread_upcond); + pthread_cond_wait(&vg_thread_wrcond, &vg_thread_lock); - if (fail && vp) { - vg_prefix_delete(&vcpp->vcp_avlroot, vp); - vp = NULL; - } + for (tp = vg_threads; tp != NULL; tp = tp->vt_next) { + assert(!tp->vt_mode); + assert(!tp->vt_stop); } - if (!vp) - continue; - - npfx++; - - /* Determine the probability of finding a match */ - vg_prefix_range_sum(vp, &bntmp, &bntmp2, &bntmp3); - BN_add(&bntmp2, &vcpp->vcp_difficulty, &bntmp); - BN_copy(&vcpp->vcp_difficulty, &bntmp2); - - if (verbose > 1) { - BN_clear(&bntmp2); - BN_set_bit(&bntmp2, 192); - BN_div(&bntmp3, NULL, &bntmp2, &bntmp, bnctx); - - dbuf = BN_bn2dec(&bntmp3); - printf("Prefix difficulty: %20s %s\n", - dbuf, patterns[i]); - OPENSSL_free(dbuf); + } else { + for (tp = vg_threads; tp != NULL; tp = tp->vt_next) { + if (tp->vt_mode) { + assert(tp->vt_mode != 2); + tp->vt_stop = 1; + } } - } - - vcpp->base.vc_npatterns += npfx; - vcpp->base.vc_npatterns_start += npfx; - if (avl_root_empty(&vcpp->vcp_avlroot)) { - printf("No prefix patterns to search\n"); - vg_prefix_context_free(vcpp); - vcpp = NULL; - goto out; + do { + for (tp = vg_threads; tp != NULL; tp = tp->vt_next) { + if (tp->vt_mode) { + assert(tp->vt_mode != 2); + pthread_cond_wait(&vg_thread_upcond, + &vg_thread_lock); + break; + } + } + } while (tp); } - assert(npfx); - ret = 1; - - vg_prefix_context_next_difficulty(vcpp, &bntmp, &bntmp2, bnctx); - -out: - BN_clear_free(&bntmp); - BN_clear_free(&bntmp2); - BN_clear_free(&bntmp3); - BN_CTX_free(bnctx); - return ret; + vtcp->vt_mode = 2; + pthread_mutex_unlock(&vg_thread_lock); + return 1; } +/* + * Address search thread main loop + */ -int -vg_prefix_test(vg_context_t *vcp, vg_thread_context_t *vctp) -{ - vg_prefix_context_t *vcpp = (vg_prefix_context_t *) vcp; - vg_prefix_t *vp; - int res = 0; - - /* - * We constrain the prefix so that we can check for - * a match without generating the lower four byte - * check code. - */ - - BN_bin2bn(vctp->vct_binres, sizeof(vctp->vct_binres), - &vctp->vct_bntarg); - -retry: - vp = vg_prefix_avl_search(&vcpp->vcp_avlroot, &vctp->vct_bntarg); - if (vp) { - if (vg_thread_upgrade_lock(vctp)) - goto retry; - - vg_thread_consolidate_key(vctp); - output_match(vctp->vct_key, vp->vp_pattern, &vcpp->base); - - vcpp->base.vc_found++; - - if (remove_on_match) { - /* Subtract the range from the difficulty */ - vg_prefix_range_sum(vp, - &vctp->vct_bntarg, - &vctp->vct_bntmp, - &vctp->vct_bntmp2); - BN_sub(&vctp->vct_bntmp, - &vcpp->vcp_difficulty, - &vctp->vct_bntarg); - BN_copy(&vcpp->vcp_difficulty, &vctp->vct_bntmp); - - vg_prefix_delete(&vcpp->vcp_avlroot,vp); - vcpp->base.vc_npatterns--; - - if (!avl_root_empty(&vcpp->vcp_avlroot)) - vg_prefix_context_next_difficulty( - vcpp, &vctp->vct_bntmp, - &vctp->vct_bntmp2, - vctp->vct_bnctx); - } - res = 1; - } - if (avl_root_empty(&vcpp->vcp_avlroot)) { - return 2; - } - return res; -} - -vg_prefix_context_t * -vg_prefix_context_new(int addrtype, int privtype) +void * +vg_thread_loop(void *arg) { - vg_prefix_context_t *vcpp; - - vcpp = (vg_prefix_context_t *) malloc(sizeof(*vcpp)); - if (vcpp) { - vcpp->base.vc_addrtype = addrtype; - vcpp->base.vc_privtype = privtype; - vcpp->base.vc_npatterns = 0; - vcpp->base.vc_npatterns_start = 0; - vcpp->base.vc_found = 0; - vcpp->base.vc_chance = 0.0; - vcpp->base.vc_test = vg_prefix_test; - avl_root_init(&vcpp->vcp_avlroot); - BN_init(&vcpp->vcp_difficulty); - } - return vcpp; -} + unsigned char eckey_buf[128]; + unsigned char hash1[32]; + int i, c, len, output_interval; + const BN_ULONG rekey_max = 10000000; + BN_ULONG npoints, rekey_at, nbatch; + vg_context_t *vcp = (vg_context_t *) arg; + EC_KEY *pkey = NULL; + const EC_GROUP *pgroup; + const EC_POINT *pgen; + const int ptarraysize = 256; + EC_POINT *ppnt[ptarraysize]; + EC_POINT *pbatchinc; -typedef struct _vg_regex_context_s { - vg_context_t base; - pcre **vcr_regex; - pcre_extra **vcr_regex_extra; - const char **vcr_regex_pat; - unsigned long vcr_nalloc; -} vg_regex_context_t; + vg_test_func_t test_func = vcp->vc_test; + vg_thread_context_t ctx; + vg_exec_context_t *vxcp; -int -vg_regex_context_add_patterns(vg_regex_context_t *vcrp, - char ** const patterns, int npatterns) -{ - const char *pcre_errptr; - int pcre_erroffset; - unsigned long i, nres, count; - void **mem; + struct timeval tvstart; - if (!npatterns) - return 1; - if (npatterns > (vcrp->vcr_nalloc - vcrp->base.vc_npatterns)) { - count = npatterns + vcrp->base.vc_npatterns; - if (count < (2 * vcrp->vcr_nalloc)) { - count = (2 * vcrp->vcr_nalloc); - } - if (count < 16) { - count = 16; - } - mem = (void **) malloc(3 * count * sizeof(void*)); - if (!mem) - return 0; + memset(&ctx, 0, sizeof(ctx)); + vxcp = &ctx.base; - for (i = 0; i < vcrp->base.vc_npatterns; i++) { - mem[i] = vcrp->vcr_regex[i]; - mem[count + i] = vcrp->vcr_regex_extra[i]; - mem[(2 * count) + i] = (void *) vcrp->vcr_regex_pat[i]; - } + vg_thread_context_init(vcp, &ctx); - if (vcrp->vcr_nalloc) - free(vcrp->vcr_regex); - vcrp->vcr_regex = (pcre **) mem; - vcrp->vcr_regex_extra = (pcre_extra **) &mem[count]; - vcrp->vcr_regex_pat = (const char **) &mem[2 * count]; - vcrp->vcr_nalloc = count; - } + pkey = vxcp->vxc_key; + pgroup = EC_KEY_get0_group(pkey); + pgen = EC_GROUP_get0_generator(pgroup); - nres = vcrp->base.vc_npatterns; - for (i = 0; i < npatterns; i++) { - vcrp->vcr_regex[nres] = - pcre_compile(patterns[i], 0, - &pcre_errptr, &pcre_erroffset, NULL); - if (!vcrp->vcr_regex[nres]) { - const char *spaces = " "; - printf("%s\n", patterns[i]); - while (pcre_erroffset > 16) { - printf("%s", spaces); - pcre_erroffset -= 16; - } - if (pcre_erroffset > 0) - printf("%s", &spaces[16 - pcre_erroffset]); - printf("^\nRegex error: %s\n", pcre_errptr); - continue; - } - vcrp->vcr_regex_extra[nres] = - pcre_study(vcrp->vcr_regex[nres], 0, &pcre_errptr); - if (pcre_errptr) { - printf("Regex error: %s\n", pcre_errptr); - pcre_free(vcrp->vcr_regex[nres]); - continue; + for (i = 0; i < ptarraysize; i++) { + ppnt[i] = EC_POINT_new(pgroup); + if (!ppnt[i]) { + printf("ERROR: out of memory?\n"); + exit(1); } - vcrp->vcr_regex_pat[nres] = patterns[i]; - nres += 1; + } + pbatchinc = EC_POINT_new(pgroup); + if (!pbatchinc) { + printf("ERROR: out of memory?\n"); + exit(1); } - if (nres == vcrp->base.vc_npatterns) - return 0; + BN_set_word(&vxcp->vxc_bntmp, ptarraysize); + EC_POINT_mul(pgroup, pbatchinc, &vxcp->vxc_bntmp, NULL, NULL, + vxcp->vxc_bnctx); + EC_POINT_make_affine(pgroup, pbatchinc, vxcp->vxc_bnctx); - vcrp->base.vc_npatterns_start += (nres - vcrp->base.vc_npatterns); - vcrp->base.vc_npatterns = nres; - return 1; -} + npoints = 0; + rekey_at = 0; + nbatch = 0; + vxcp->vxc_key = pkey; + vxcp->vxc_binres[0] = vcp->vc_addrtype; + c = 0; + output_interval = 1000; + gettimeofday(&tvstart, NULL); -void -vg_regex_context_free(vg_regex_context_t *vcrp) -{ - int i; - for (i = 0; i < vcrp->base.vc_npatterns; i++) { - if (vcrp->vcr_regex_extra[i]) - pcre_free(vcrp->vcr_regex_extra[i]); - pcre_free(vcrp->vcr_regex[i]); - } - if (vcrp->vcr_nalloc) - free(vcrp->vcr_regex); - free(vcrp); -} + while (1) { + if (++npoints >= rekey_at) { + pthread_mutex_lock(&vg_thread_lock); + /* Generate a new random private key */ + EC_KEY_generate_key(pkey); + npoints = 0; -int -vg_regex_test(vg_context_t *vcp, vg_thread_context_t *vctp) -{ - vg_regex_context_t *vcrp = (vg_regex_context_t *) vcp; - - unsigned char hash1[32], hash2[32]; - int i, zpfx, p, d, nres, re_vec[9]; - char b58[40]; - BIGNUM bnrem; - BIGNUM *bn, *bndiv, *bnptmp; - int res = 0; - - pcre *re; - - BN_init(&bnrem); - - /* Hash the hash and write the four byte check code */ - SHA256(vctp->vct_binres, 21, hash1); - SHA256(hash1, sizeof(hash1), hash2); - memcpy(&vctp->vct_binres[21], hash2, 4); - - bn = &vctp->vct_bntmp; - bndiv = &vctp->vct_bntmp2; - - BN_bin2bn(vctp->vct_binres, sizeof(vctp->vct_binres), bn); - - /* Compute the complete encoded address */ - for (zpfx = 0; zpfx < 25 && vctp->vct_binres[zpfx] == 0; zpfx++); - p = sizeof(b58) - 1; - b58[p] = '\0'; - while (!BN_is_zero(bn)) { - BN_div(bndiv, &bnrem, bn, &vctp->vct_bnbase, vctp->vct_bnctx); - bnptmp = bn; - bn = bndiv; - bndiv = bnptmp; - d = BN_get_word(&bnrem); - b58[--p] = b58_alphabet[d]; - } - while (zpfx--) { - b58[--p] = b58_alphabet[0]; - } + /* Determine rekey interval */ + EC_GROUP_get_order(pgroup, &vxcp->vxc_bntmp, + vxcp->vxc_bnctx); + BN_sub(&vxcp->vxc_bntmp2, + &vxcp->vxc_bntmp, + EC_KEY_get0_private_key(pkey)); + rekey_at = BN_get_word(&vxcp->vxc_bntmp2); + if ((rekey_at == BN_MASK2) || (rekey_at > rekey_max)) + rekey_at = rekey_max; + assert(rekey_at > 0); - /* - * Run the regular expressions on it - * SLOW, runs in linear time with the number of REs - */ -restart_loop: - nres = vcrp->base.vc_npatterns; - if (!nres) { - res = 2; - goto out; - } - for (i = 0; i < nres; i++) { - d = pcre_exec(vcrp->vcr_regex[i], - vcrp->vcr_regex_extra[i], - &b58[p], (sizeof(b58) - 1) - p, 0, - 0, - re_vec, sizeof(re_vec)/sizeof(re_vec[0])); - - if (d <= 0) { - if (d != PCRE_ERROR_NOMATCH) { - printf("PCRE error: %d\n", d); - res = 2; - goto out; + EC_POINT_copy(ppnt[0], EC_KEY_get0_public_key(pkey)); + pthread_mutex_unlock(&vg_thread_lock); + + npoints++; + vxcp->vxc_delta = 0; + + for (nbatch = 1; + (nbatch < ptarraysize) && (npoints < rekey_at); + nbatch++, npoints++) { + EC_POINT_add(pgroup, + ppnt[nbatch], + ppnt[nbatch-1], + pgen, vxcp->vxc_bnctx); } - continue; - } - re = vcrp->vcr_regex[i]; - - if (vg_thread_upgrade_lock(vctp) && - ((i >= vcrp->base.vc_npatterns) || - (vcrp->vcr_regex[i] != re))) - goto restart_loop; - - vg_thread_consolidate_key(vctp); - output_match(vctp->vct_key, - vcrp->vcr_regex_pat[i], - &vcrp->base); - vcrp->base.vc_found++; - - if (remove_on_match) { - pcre_free(vcrp->vcr_regex[i]); - if (vcrp->vcr_regex_extra[i]) - pcre_free(vcrp->vcr_regex_extra[i]); - nres -= 1; - vcrp->base.vc_npatterns = nres; - if (!nres) { - res = 2; - goto out; + } else { + /* + * Common case + * + * EC_POINT_add() can skip a few multiplies if + * one or both inputs are affine (Z_is_one). + * This is the case for every point in ppnt, as + * well as pbatchinc. + */ + assert(nbatch == ptarraysize); + for (nbatch = 0; + (nbatch < ptarraysize) && (npoints < rekey_at); + nbatch++, npoints++) { + EC_POINT_add(pgroup, + ppnt[nbatch], + ppnt[nbatch], + pbatchinc, + vxcp->vxc_bnctx); } - vcrp->vcr_regex[i] = vcrp->vcr_regex[nres]; - vcrp->vcr_regex_extra[i] = - vcrp->vcr_regex_extra[nres]; - vcrp->vcr_regex_pat[i] = vcrp->vcr_regex_pat[nres]; - vcrp->base.vc_npatterns = nres; } - res = 1; - } -out: - BN_clear_free(&bnrem); - return res; -} -vg_regex_context_t * -vg_regex_context_new(int addrtype, int privtype) -{ - vg_regex_context_t *vcrp; - - vcrp = (vg_regex_context_t *) malloc(sizeof(*vcrp)); - if (vcrp) { - vcrp->base.vc_addrtype = addrtype; - vcrp->base.vc_privtype = privtype; - vcrp->base.vc_npatterns = 0; - vcrp->base.vc_npatterns_start = 0; - vcrp->base.vc_found = 0; - vcrp->base.vc_chance = 0.0; - vcrp->base.vc_test = vg_regex_test; - vcrp->vcr_regex = NULL; - vcrp->vcr_nalloc = 0; - } - return vcrp; -} + /* + * The single most expensive operation performed in this + * loop is modular inversion of ppnt->Z. There is an + * algorithm implemented in OpenSSL to do batched inversion + * that only does one actual BN_mod_inverse(), and saves + * a _lot_ of time. + * + * To take advantage of this, we batch up a few points, + * and feed them to EC_POINTs_make_affine() below. + */ + EC_POINTs_make_affine(pgroup, nbatch, ppnt, vxcp->vxc_bnctx); -int -read_file(FILE *fp, char ***result, int *rescount) -{ - int ret = 1; + for (i = 0; i < nbatch; i++, vxcp->vxc_delta++) { + /* Hash the public key */ + len = EC_POINT_point2oct(pgroup, ppnt[i], + POINT_CONVERSION_UNCOMPRESSED, + eckey_buf, + sizeof(eckey_buf), + vxcp->vxc_bnctx); - char **patterns; - char *buf = NULL, *obuf, *pat; - const int blksize = 16*1024; - int nalloc = 16; - int npatterns = 0; - int count, pos; + SHA256(eckey_buf, len, hash1); + RIPEMD160(hash1, sizeof(hash1), &vxcp->vxc_binres[1]); - patterns = (char**) malloc(sizeof(char*) * nalloc); - count = 0; - pos = 0; + vxcp->vxc_point = ppnt[i]; - while (1) { - obuf = buf; - buf = (char *) malloc(blksize); - if (!buf) { - ret = 0; - break; - } - if (pos < count) { - memcpy(buf, &obuf[pos], count - pos); - } - pos = count - pos; - count = fread(&buf[pos], 1, blksize - pos, fp); - if (count < 0) { - printf("Error reading file: %s\n", strerror(errno)); - ret = 0; - } - if (count <= 0) - break; - count += pos; - pat = buf; - - while (pos < count) { - if ((buf[pos] == '\r') || (buf[pos] == '\n')) { - buf[pos] = '\0'; - if (pat) { - if (npatterns == nalloc) { - nalloc *= 2; - patterns = (char**) - realloc(patterns, - sizeof(char*) * - nalloc); - } - patterns[npatterns] = pat; - npatterns++; - pat = NULL; - } - } - else if (!pat) { - pat = &buf[pos]; + switch (test_func(vxcp)) { + case 1: + npoints = 0; + rekey_at = 0; + i = nbatch; + break; + case 2: + goto out; + default: + break; } - pos++; } - pos = pat ? (pat - buf) : count; - } + c += (i + 1); + if (c >= output_interval) { + output_interval = vg_output_timing(vcp, c, &tvstart); + c = 0; + } - *result = patterns; - *rescount = npatterns; + vg_thread_yield(&ctx); + } + +out: + vg_thread_context_del(&ctx); - return ret; + for (i = 0; i < ptarraysize; i++) + if (ppnt[i]) + EC_POINT_free(ppnt[i]); + if (pbatchinc) + EC_POINT_free(pbatchinc); + return NULL; } + #if !defined(_WIN32) int count_processors(void) @@ -2243,7 +399,7 @@ count_processors(void) #endif int -start_threads(void *(*func)(void *), void *arg, int nthreads) +start_threads(vg_context_t *vcp, int nthreads) { pthread_t thread; @@ -2256,16 +412,16 @@ start_threads(void *(*func)(void *), void *arg, int nthreads) } } - if (verbose > 1) { + if (vcp->vc_verbose > 1) { printf("Using %d worker thread(s)\n", nthreads); } while (--nthreads) { - if (pthread_create(&thread, NULL, func, arg)) + if (pthread_create(&thread, NULL, vg_thread_loop, vcp)) return 0; } - func(arg); + vg_thread_loop(vcp); return 1; } @@ -2306,13 +462,16 @@ main(int argc, char **argv) int privtype = 128; int regex = 0; int caseinsensitive = 0; + int verbose = 1; + int remove_on_match = 1; int opt; char *seedfile = NULL; FILE *fp = NULL; + const char *result_file = NULL; char **patterns; int npatterns = 0; int nthreads = 0; - void *thread_arg = NULL; + vg_context_t *vcp = NULL; while ((opt = getopt(argc, argv, "vqrikNTt:h?f:o:s:")) != -1) { switch (opt) { @@ -2413,7 +572,7 @@ main(int argc, char **argv) } if (fp) { - if (!read_file(fp, &patterns, &npatterns)) { + if (!vg_read_file(fp, &patterns, &npatterns)) { printf("Failed to load pattern file\n"); return 1; } @@ -2430,26 +589,29 @@ main(int argc, char **argv) } if (regex) { - vg_regex_context_t *vcrp; - vcrp = vg_regex_context_new(addrtype, privtype); - if (!vg_regex_context_add_patterns(vcrp, patterns, npatterns)) - return 1; - if ((verbose > 0) && (vcrp->base.vc_npatterns > 1)) - printf("Regular expressions: %ld\n", - vcrp->base.vc_npatterns); - thread_arg = vcrp; + vcp = vg_regex_context_new(addrtype, privtype); } else { - vg_prefix_context_t *vcpp; - vcpp = vg_prefix_context_new(addrtype, privtype); - if (!vg_prefix_context_add_patterns(vcpp, patterns, npatterns, - caseinsensitive)) - return 1; - thread_arg = vcpp; + vcp = vg_prefix_context_new(addrtype, privtype, + caseinsensitive); + } + + vcp->vc_verbose = verbose; + vcp->vc_result_file = result_file; + vcp->vc_remove_on_match = remove_on_match; + + if (!vg_context_add_patterns(vcp, patterns, npatterns)) + return 1; + + if (!vcp->vc_npatterns) { + printf("No patterns to search\n"); + return 1; } - if (!start_threads((void *(*)(void*))vg_thread_loop, - thread_arg, nthreads)) + if ((verbose > 0) && regex && (vcp->vc_npatterns > 1)) + printf("Regular expressions: %ld\n", vcp->vc_npatterns); + + if (!start_threads(vcp, nthreads)) return 1; return 0; } diff --git a/winglue.c b/winglue.c index dc22d71..e7bd9f9 100644 --- a/winglue.c +++ b/winglue.c @@ -1,10 +1,25 @@ -#include -#include -#include - -#define INLINE -#define snprintf _snprintf +/* + * Vanitygen, vanity bitcoin address generator + * Copyright (C) 2011 + * + * Vanitygen is free software: you can redistribute it and/or modify + * it under the terms of the GNU Affero General Public License as published by + * the Free Software Foundation, either version 3 of the License, or + * any later version. + * + * Vanitygen is distributed in the hope that it will be useful, + * but WITHOUT ANY WARRANTY; without even the implied warranty of + * MERCHANTABILITY or FITNESS FOR A PARTICULAR PURPOSE. See the + * GNU Affero General Public License for more details. + * + * You should have received a copy of the GNU Affero General Public License + * along with Vanitygen. If not, see . + */ +#include +#include +#include +#include "winglue.h" int count_processors(void) @@ -64,8 +79,6 @@ count_processors(void) #define DELTA_EPOCH_IN_MICROSECS 11644473600000000ULL #endif -struct timezone; - int gettimeofday(struct timeval *tv, struct timezone *tz) { @@ -104,7 +117,7 @@ timersub(struct timeval *a, struct timeval *b, struct timeval *result) * getopt() for Win32 -- public domain ripped from codeproject.com */ -TCHAR *optarg; +TCHAR *optarg = NULL; int optind = 0; int getopt(int argc, TCHAR *argv[], TCHAR *optstring) diff --git a/winglue.h b/winglue.h new file mode 100644 index 0000000..f94866d --- /dev/null +++ b/winglue.h @@ -0,0 +1,42 @@ +/* + * Vanitygen, vanity bitcoin address generator + * Copyright (C) 2011 + * + * Vanitygen is free software: you can redistribute it and/or modify + * it under the terms of the GNU Affero General Public License as published by + * the Free Software Foundation, either version 3 of the License, or + * any later version. + * + * Vanitygen is distributed in the hope that it will be useful, + * but WITHOUT ANY WARRANTY; without even the implied warranty of + * MERCHANTABILITY or FITNESS FOR A PARTICULAR PURPOSE. See the + * GNU Affero General Public License for more details. + * + * You should have received a copy of the GNU Affero General Public License + * along with Vanitygen. If not, see . + */ + +#if !defined (__VG_WINGLUE_H__) +#define __VG_WINGLUE_H__ + +#include +#include +#include + +#define INLINE +#define snprintf _snprintf + +struct timezone; + +extern int gettimeofday(struct timeval *tv, struct timezone *tz); +extern void timersub(struct timeval *a, struct timeval *b, + struct timeval *result); + +extern TCHAR *optarg; +extern int optind; + +extern int getopt(int argc, TCHAR *argv[], TCHAR *optstring); + +extern int count_processors(void); + +#endif /* !defined (__VG_WINGLUE_H__) */