From 8ab297ae6a4fa1bae423f0fd11b65b845f5a9405 Mon Sep 17 00:00:00 2001 From: Philippe Teuwen Date: Sat, 4 Apr 2020 14:03:21 +0200 Subject: [PATCH] add new ht2 tools, thanks to anonymous donator --- tools/hitag2crack/README.md | 50 +- tools/hitag2crack/crack5/HardwareProfile.h | 524 ++++++++++++ tools/hitag2crack/crack5/Makefile | 18 + tools/hitag2crack/crack5/README.md | 23 + tools/hitag2crack/crack5/hitagcrypto.c | 373 +++++++++ tools/hitag2crack/crack5/hitagcrypto.h | 167 ++++ tools/hitag2crack/crack5/ht2crack2utils.c | 172 ++++ tools/hitag2crack/crack5/ht2crack2utils.h | 35 + tools/hitag2crack/crack5/ht2crack5.c | 751 ++++++++++++++++++ tools/hitag2crack/crack5/rfidler.h | 412 ++++++++++ tools/hitag2crack/crack5/util.h | 147 ++++ tools/hitag2crack/crack5/utilpart.c | 180 +++++ tools/hitag2crack/crack5gpu/HardwareProfile.h | 524 ++++++++++++ tools/hitag2crack/crack5gpu/Makefile | 24 + tools/hitag2crack/crack5gpu/README.md | 27 + tools/hitag2crack/crack5gpu/hitagcrypto.c | 373 +++++++++ tools/hitag2crack/crack5gpu/hitagcrypto.h | 167 ++++ tools/hitag2crack/crack5gpu/ht2crack2utils.c | 172 ++++ tools/hitag2crack/crack5gpu/ht2crack2utils.h | 35 + tools/hitag2crack/crack5gpu/ht2crack5.c | 421 ++++++++++ .../hitag2crack/crack5gpu/ht2crack5kernel.cl | 429 ++++++++++ tools/hitag2crack/crack5gpu/rfidler.h | 412 ++++++++++ tools/hitag2crack/crack5gpu/util.h | 147 ++++ tools/hitag2crack/crack5gpu/utilpart.c | 180 +++++ 24 files changed, 5762 insertions(+), 1 deletion(-) create mode 100644 tools/hitag2crack/crack5/HardwareProfile.h create mode 100644 tools/hitag2crack/crack5/Makefile create mode 100644 tools/hitag2crack/crack5/README.md create mode 100644 tools/hitag2crack/crack5/hitagcrypto.c create mode 100644 tools/hitag2crack/crack5/hitagcrypto.h create mode 100644 tools/hitag2crack/crack5/ht2crack2utils.c create mode 100644 tools/hitag2crack/crack5/ht2crack2utils.h create mode 100644 tools/hitag2crack/crack5/ht2crack5.c create mode 100644 tools/hitag2crack/crack5/rfidler.h create mode 100644 tools/hitag2crack/crack5/util.h create mode 100644 tools/hitag2crack/crack5/utilpart.c create mode 100644 tools/hitag2crack/crack5gpu/HardwareProfile.h create mode 100644 tools/hitag2crack/crack5gpu/Makefile create mode 100644 tools/hitag2crack/crack5gpu/README.md create mode 100644 tools/hitag2crack/crack5gpu/hitagcrypto.c create mode 100644 tools/hitag2crack/crack5gpu/hitagcrypto.h create mode 100644 tools/hitag2crack/crack5gpu/ht2crack2utils.c create mode 100644 tools/hitag2crack/crack5gpu/ht2crack2utils.h create mode 100644 tools/hitag2crack/crack5gpu/ht2crack5.c create mode 100644 tools/hitag2crack/crack5gpu/ht2crack5kernel.cl create mode 100644 tools/hitag2crack/crack5gpu/rfidler.h create mode 100644 tools/hitag2crack/crack5gpu/util.h create mode 100644 tools/hitag2crack/crack5gpu/utilpart.c diff --git a/tools/hitag2crack/README.md b/tools/hitag2crack/README.md index a43d7a1fd..9f31f6d62 100644 --- a/tools/hitag2crack/README.md +++ b/tools/hitag2crack/README.md @@ -1,7 +1,10 @@ HiTag2 Cracking Suite --------------------- -Author: Kevin Sheldrake +Authors: + +* Attacks 1, 2, 3, 4 : Kevin Sheldrake +* Attacks 5, 5gpu : anonymous, based on https://github.com/factoritbv/hitag2hell by FactorIT B.V. Introduction ------------ @@ -89,6 +92,21 @@ encrypted nonces and the keystream they should produce. Each guess is then expanded by 1 bit and the process iterates, with only the best guesses taken forward to the next iteration. +Attack 5 +-------- + +Attack 5 is heavily based on the HiTag2 Hell CPU implementation from https://github.com/factoritbv/hitag2hell by FactorIT B.V., +with the following changes: + +* Main takes a UID and 2 {nR},{aR} pairs as arguments and searches for states producing the first aR sample, reconstructs the corresponding key candidates and tests them against the second nR,aR pair; +* Reuses the Hitag helping functions of the other attacks. + +Attack 5gpu +----------- + +Attack 5gpu is identical to attack 5, simply the code has been ported to OpenCL +to run on GPUs and is therefore much faster than attack 5. + Usage details: Attack 1 ----------------------- @@ -172,6 +190,36 @@ Stop once you got enough pairs. Start with -N 16 and -t 500000. If the attack fails to find the key, double the table size and try again, repeating if it still fails. +Usage details: Attack 5 +----------------------- + +Attack 5 requires two encrypted nonce and challenge +response value pairs (nR, aR) for the tag's UID. + +``` +pm3 --> lf hitag sniff +``` +Stop once you got two pairs. + +``` +$ ./ht2crack5 +``` + +Usage details: Attack 5gpu +-------------------------- + +Attack 5gpu requires two encrypted nonce and challenge +response value pairs (nR, aR) for the tag's UID. + +``` +pm3 --> lf hitag sniff +``` +Stop once you got two pairs. + +``` +$ ./ht2crack5gpu +``` + Usage details: Next steps ------------------------- diff --git a/tools/hitag2crack/crack5/HardwareProfile.h b/tools/hitag2crack/crack5/HardwareProfile.h new file mode 100644 index 000000000..bce139042 --- /dev/null +++ b/tools/hitag2crack/crack5/HardwareProfile.h @@ -0,0 +1,524 @@ +/*************************************************************************** + * A copy of the GNU GPL is appended to this file. * + * * + * This licence is based on the nmap licence, and we express our gratitude * + * for the work that went into producing it. There is no other connection * + * between RFIDler and nmap either expressed or implied. * + * * + ********************** IMPORTANT RFIDler LICENSE TERMS ******************** + * * + * * + * All references to RFIDler herein imply all it's derivatives, namely: * + * * + * o RFIDler-LF Standard * + * o RFIDler-LF Lite * + * o RFIDler-LF Nekkid * + * * + * * + * RFIDler is (C) 2013-2014 Aperture Labs Ltd. * + * * + * This program is free software; you may redistribute and/or modify it * + * under the terms of the GNU General Public License as published by the * + * Free Software Foundation; Version 2 ("GPL"), BUT ONLY WITH ALL OF THE * + * CLARIFICATIONS AND EXCEPTIONS DESCRIBED HEREIN. This guarantees your * + * right to use, modify, and redistribute this software under certain * + * conditions. If you wish to embed RFIDler technology into proprietary * + * software or hardware, we sell alternative licenses * + * (contact sales@aperturelabs.com). * + * * + * Note that the GPL places important restrictions on "derivative works", * + * yet it does not provide a detailed definition of that term. To avoid * + * misunderstandings, we interpret that term as broadly as copyright law * + * allows. For example, we consider an application to constitute a * + * derivative work for the purpose of this license if it does any of the * + * following with any software or content covered by this license * + * ("Covered Software"): * + * * + * o Integrates source code from Covered Software. * + * * + * o Is designed specifically to execute Covered Software and parse the * + * results (as opposed to typical shell or execution-menu apps, which will * + * execute anything you tell them to). * + * * + * o Includes Covered Software in a proprietary executable installer. The * + * installers produced by InstallShield are an example of this. Including * + * RFIDler with other software in compressed or archival form does not * + * trigger this provision, provided appropriate open source decompression * + * or de-archiving software is widely available for no charge. For the * + * purposes of this license, an installer is considered to include Covered * + * Software even if it actually retrieves a copy of Covered Software from * + * another source during runtime (such as by downloading it from the * + * Internet). * + * * + * o Links (statically or dynamically) to a library which does any of the * + * above. * + * * + * o Executes a helper program, module, or script to do any of the above. * + * * + * This list is not exclusive, but is meant to clarify our interpretation * + * of derived works with some common examples. Other people may interpret * + * the plain GPL differently, so we consider this a special exception to * + * the GPL that we apply to Covered Software. Works which meet any of * + * these conditions must conform to all of the terms of this license, * + * particularly including the GPL Section 3 requirements of providing * + * source code and allowing free redistribution of the work as a whole. * + * * + * As another special exception to the GPL terms, Aperture Labs Ltd. grants* + * permission to link the code of this program with any version of the * + * OpenSSL library which is distributed under a license identical to that * + * listed in the included docs/licenses/OpenSSL.txt file, and distribute * + * linked combinations including the two. * + * * + * Any redistribution of Covered Software, including any derived works, * + * must obey and carry forward all of the terms of this license, including * + * obeying all GPL rules and restrictions. For example, source code of * + * the whole work must be provided and free redistribution must be * + * allowed. All GPL references to "this License", are to be treated as * + * including the terms and conditions of this license text as well. * + * * + * Because this license imposes special exceptions to the GPL, Covered * + * Work may not be combined (even as part of a larger work) with plain GPL * + * software. The terms, conditions, and exceptions of this license must * + * be included as well. This license is incompatible with some other open * + * source licenses as well. In some cases we can relicense portions of * + * RFIDler or grant special permissions to use it in other open source * + * software. Please contact sales@aperturelabs.com with any such requests.* + * Similarly, we don't incorporate incompatible open source software into * + * Covered Software without special permission from the copyright holders. * + * * + * If you have any questions about the licensing restrictions on using * + * RFIDler in other works, are happy to help. As mentioned above, we also * + * offer alternative license to integrate RFIDler into proprietary * + * applications and appliances. These contracts have been sold to dozens * + * of software vendors, and generally include a perpetual license as well * + * as providing for priority support and updates. They also fund the * + * continued development of RFIDler. Please email sales@aperturelabs.com * + * for further information. * + * If you have received a written license agreement or contract for * + * Covered Software stating terms other than these, you may choose to use * + * and redistribute Covered Software under those terms instead of these. * + * * + * Source is provided to this software because we believe users have a * + * right to know exactly what a program is going to do before they run it. * + * This also allows you to audit the software for security holes (none * + * have been found so far). * + * * + * Source code also allows you to port RFIDler to new platforms, fix bugs, * + * and add new features. You are highly encouraged to send your changes * + * to the RFIDler mailing list for possible incorporation into the * + * main distribution. By sending these changes to Aperture Labs Ltd. or * + * one of the Aperture Labs Ltd. development mailing lists, or checking * + * them into the RFIDler source code repository, it is understood (unless * + * you specify otherwise) that you are offering the RFIDler Project * + * (Aperture Labs Ltd.) the unlimited, non-exclusive right to reuse, * + * modify, and relicense the code. RFIDler will always be available Open * + * Source, but this is important because the inability to relicense code * + * has caused devastating problems for other Free Software projects (such * + * as KDE and NASM). We also occasionally relicense the code to third * + * parties as discussed above. If you wish to specify special license * + * conditions of your contributions, just say so when you send them. * + * * + * This program 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 RFIDler * + * license file for more details (it's in a COPYING file included with * + * RFIDler, and also available from * + * https://github.com/ApertureLabsLtd/RFIDler/COPYING * + * * + ***************************************************************************/ + +// Author: Adam Laurie + + + +#ifndef HARDWARE_PROFILE_UBW32_H +#define HARDWARE_PROFILE_UBW32_H + +//#include "plib.h" +typedef char BOOL; +typedef char BYTE; +typedef int rtccTime; +typedef int rtccDate; + + +#ifndef __PIC32MX__ +#define __PIC32MX__ +#endif + +#define GetSystemClock() (80000000ul) +#define GetPeripheralClock() (GetSystemClock()) +#define GetInstructionClock() (GetSystemClock()) + +//#define USE_SELF_POWER_SENSE_IO +#define tris_self_power TRISAbits.TRISA2 // Input +#define self_power 1 + +//#define USE_USB_BUS_SENSE_IO +#define tris_usb_bus_sense TRISBbits.TRISB5 // Input +#define USB_BUS_SENSE 1 + +// LEDs +#define mLED_1 LATEbits.LATE3 + +#define mLED_2 LATEbits.LATE2 +#define mLED_Comms mLED_2 + +#define mLED_3 LATEbits.LATE1 +#define mLED_Clock mLED_3 + +#define mLED_4 LATEbits.LATE0 +#define mLED_Emulate mLED_4 + +#define mLED_5 LATGbits.LATG6 +#define mLED_Read mLED_5 + +#define mLED_6 LATAbits.LATA15 +#define mLED_User mLED_6 + +#define mLED_7 LATDbits.LATD11 +#define mLED_Error mLED_7 + +// active low +#define mLED_ON 0 +#define mLED_OFF 1 + +#define mGetLED_1() mLED_1 +#define mGetLED_USB() mLED_1 +#define mGetLED_2() mLED_2 +#define mGetLED_Comms() mLED_2 +#define mGetLED_3() mLED_3 +#define mGetLED_Clock() mLED_3 +#define mGetLED_4() mLED_4 +#define mGetLED_Emulate() mLED_4 +#define mGetLED_5() mLED_5 +#define mGetLED_Read() mLED_5 +#define mGetLED_6() mLED_6 +#define mGetLED_User() mLED_6 +#define mGetLED_7() mLED_7 +#define mGetLED_Error() mLED_7 + +#define mLED_1_On() mLED_1 = mLED_ON +#define mLED_USB_On() mLED_1_On() +#define mLED_2_On() mLED_2 = mLED_ON +#define mLED_Comms_On() mLED_2_On() +#define mLED_3_On() mLED_3 = mLED_ON +#define mLED_Clock_On() mLED_3_On() +#define mLED_4_On() mLED_4 = mLED_ON +#define mLED_Emulate_On() mLED_4_On() +#define mLED_5_On() mLED_5 = mLED_ON +#define mLED_Read_On() mLED_5_On() +#define mLED_6_On() mLED_6 = mLED_ON +#define mLED_User_On() mLED_6_On() +#define mLED_7_On() mLED_7 = mLED_ON +#define mLED_Error_On() mLED_7_On() + +#define mLED_1_Off() mLED_1 = mLED_OFF +#define mLED_USB_Off() mLED_1_Off() +#define mLED_2_Off() mLED_2 = mLED_OFF +#define mLED_Comms_Off() mLED_2_Off() +#define mLED_3_Off() mLED_3 = mLED_OFF +#define mLED_Clock_Off() mLED_3_Off() +#define mLED_4_Off() mLED_4 = mLED_OFF +#define mLED_Emulate_Off() mLED_4_Off() +#define mLED_5_Off() mLED_5 = mLED_OFF +#define mLED_Read_Off() mLED_5_Off() +#define mLED_6_Off() mLED_6 = mLED_OFF +#define mLED_User_Off() mLED_6_Off() +#define mLED_7_Off() mLED_7 = mLED_OFF +#define mLED_Error_Off() mLED_7_Off() + +#define mLED_1_Toggle() mLED_1 = !mLED_1 +#define mLED_USB_Toggle() mLED_1_Toggle() +#define mLED_2_Toggle() mLED_2 = !mLED_2 +#define mLED_Comms_Toggle() mLED_2_Toggle() +#define mLED_3_Toggle() mLED_3 = !mLED_3 +#define mLED_Clock_Toggle() mLED_3_Toggle() +#define mLED_4_Toggle() mLED_4 = !mLED_4 +#define mLED_Emulate_Toggle() mLED_4_Toggle() +#define mLED_5_Toggle() mLED_5 = !mLED_5 +#define mLED_Read_Toggle( ) mLED_5_Toggle() +#define mLED_6_Toggle() mLED_6 = !mLED_6 +#define mLED_User_Toggle() mLED_6_Toggle() +#define mLED_7_Toggle() mLED_7 = !mLED_7 +#define mLED_Error_Toggle() mLED_7_Toggle() + +#define mLED_All_On() { mLED_1_On(); mLED_2_On(); mLED_3_On(); mLED_4_On(); mLED_5_On(); mLED_6_On(); mLED_7_On(); } +#define mLED_All_Off() { mLED_1_Off(); mLED_2_Off(); mLED_3_Off(); mLED_4_Off(); mLED_5_Off(); mLED_6_Off(); mLED_7_Off(); } + +// usb status lights +#define mLED_Both_Off() {mLED_USB_Off();mLED_Comms_Off();} +#define mLED_Both_On() {mLED_USB_On();mLED_Comms_On();} +#define mLED_Only_USB_On() {mLED_USB_On();mLED_Comms_Off();} +#define mLED_Only_Comms_On() {mLED_USB_Off();mLED_Comms_On();} + +/** SWITCH *********************************************************/ +#define swBootloader PORTEbits.RE7 +#define swUser PORTEbits.RE6 + +/** I/O pin definitions ********************************************/ +#define INPUT_PIN 1 +#define OUTPUT_PIN 0 + +#define TRUE 1 +#define FALSE 0 + +#define ENABLE 1 +#define DISABE 0 + +#define EVEN 0 +#define ODD 1 + +#define LOW FALSE +#define HIGH TRUE + +#define CLOCK_ON LOW +#define CLOCK_OFF HIGH + +// output coil control - select between reader/emulator circuits +#define COIL_MODE LATBbits.LATB4 +#define COIL_MODE_READER() COIL_MODE= LOW +#define COIL_MODE_EMULATOR() COIL_MODE= HIGH + +// coil for emulation +#define COIL_OUT LATGbits.LATG9 +#define COIL_OUT_HIGH() COIL_OUT=HIGH +#define COIL_OUT_LOW() COIL_OUT=LOW + +// door relay (active low) +#define DOOR_RELAY LATAbits.LATA14 +#define DOOR_RELAY_OPEN() DOOR_RELAY= HIGH +#define DOOR_RELAY_CLOSE() DOOR_RELAY= LOW + +// inductance/capacitance freq +#define IC_FREQUENCY PORTAbits.RA2 + +#define SNIFFER_COIL PORTDbits.RD12 // external reader clock detect +#define READER_ANALOGUE PORTBbits.RB11 // reader coil analogue +#define DIV_LOW_ANALOGUE PORTBbits.RB12 // voltage divider LOW analogue +#define DIV_HIGH_ANALOGUE PORTBbits.RB13 // voltage divider HIGH analogue + +// clock coil (normally controlled by OC Module, but defined here so we can force it high or low) +#define CLOCK_COIL PORTDbits.RD4 +#define CLOCK_COIL_MOVED PORTDbits.RD0 // temporary for greenwire + +// digital output after analogue reader circuit +#define READER_DATA PORTDbits.RD8 + +// trace / debug +#define DEBUG_PIN_1 LATCbits.LATC1 +#define DEBUG_PIN_1_TOGGLE() DEBUG_PIN_1= !DEBUG_PIN_1 +#define DEBUG_PIN_2 LATCbits.LATC2 +#define DEBUG_PIN_2_TOGGLE() DEBUG_PIN_2= !DEBUG_PIN_2 +#define DEBUG_PIN_3 LATCbits.LATC3 +#define DEBUG_PIN_3_TOGGLE() DEBUG_PIN_3= !DEBUG_PIN_3 +#define DEBUG_PIN_4 LATEbits.LATE5 +#define DEBUG_PIN_4_TOGGLE() DEBUG_PIN_4= !DEBUG_PIN_4 + +// spi (sdi1) for sd card (not directly referenced) +//#define SD_CARD_RX LATCbits.LATC4 +//#define SD_CARD_TX LATDbits.LATD0 +//#define SD_CARD_CLK LATDbits.LATD10 +//#define SD_CARD_SS LATDbits.LATD9 +// spi for SD card +#define SD_CARD_DET LATFbits.LATF0 +#define SD_CARD_WE LATFbits.LATF1 // write enable - unused for microsd but allocated anyway as library checks it +// (held LOW by default - cut solder bridge to GND to free pin if required) +#define SPI_SD SPI_CHANNEL1 +#define SPI_SD_BUFF SPI1BUF +#define SPI_SD_STAT SPI1STATbits +// see section below for more defines! + +// iso 7816 smartcard +// microchip SC module defines pins so we don't need to, but +// they are listed here to help avoid conflicts +#define ISO_7816_RX LATBbits.LATF2 // RX +#define ISO_7816_TX LATBbits.LATF8 // TX +#define ISO_7816_VCC LATBbits.LATB9 // Power +#define ISO_7816_CLK LATCbits.LATD1 // Clock +#define ISO_7816_RST LATEbits.LATE8 // Reset + +// user LED +#define USER_LED LATDbits.LATD7 +#define USER_LED_ON() LATDbits.LATD7=1 +#define USER_LED_OFF() LATDbits.LATD7=0 + +// LCR +#define LCR_CALIBRATE LATBbits.LATB5 + +// wiegand / clock & data +#define WIEGAND_IN_0 PORTDbits.RD5 +#define WIEGAND_IN_0_PULLUP CNPUEbits.CNPUE14 +#define WIEGAND_IN_0_PULLDOWN CNPDbits.CNPD14 +#define WIEGAND_IN_1 PORTDbits.RD6 +#define WIEGAND_IN_1_PULLUP CNPUEbits.CNPUE15 +#define WIEGAND_IN_1_PULLDOWN CNPDbits.CNPD15 +#define CAND_IN_DATA WIEGAND_IN_0 +#define CAND_IN_CLOCK WIEGAND_IN_1 + +#define WIEGAND_OUT_0 LATDbits.LATD3 +#define WIEGAND_OUT_1 LATDbits.LATD2 +#define WIEGAND_OUT_0_TRIS TRISDbits.TRISD3 +#define WIEGAND_OUT_1_TRIS TRISDbits.TRISD2 +#define CAND_OUT_DATA WIEGAND_OUT_0 +#define CAND_OUT_CLOCK WIEGAND_OUT_1 + +// connect/disconnect reader clock from coil - used to send RWD signals by creating gaps in carrier +#define READER_CLOCK_ENABLE LATEbits.LATE9 +#define READER_CLOCK_ENABLE_ON() READER_CLOCK_ENABLE=CLOCK_ON +#define READER_CLOCK_ENABLE_OFF(x) {READER_CLOCK_ENABLE=CLOCK_OFF; COIL_OUT=x;} + +// these input pins must NEVER bet set to output or they will cause short circuits! +// they can be used to see data from reader before it goes into or gate +#define OR_IN_A PORTAbits.RA4 +#define OR_IN_B PORTAbits.RA5 + + +// CNCON and CNEN are set to allow wiegand input pin weak pullups to be switched on +#define Init_GPIO() { \ + CNCONbits.ON= TRUE; \ + CNENbits.CNEN14= TRUE; \ + CNENbits.CNEN15= TRUE; \ + TRISAbits.TRISA2= INPUT_PIN; \ + TRISAbits.TRISA4= INPUT_PIN; \ + TRISAbits.TRISA5= INPUT_PIN; \ + TRISAbits.TRISA14= OUTPUT_PIN; \ + TRISAbits.TRISA15= OUTPUT_PIN; \ + TRISBbits.TRISB4= OUTPUT_PIN; \ + TRISBbits.TRISB5= OUTPUT_PIN; \ + TRISBbits.TRISB9= OUTPUT_PIN; \ + TRISBbits.TRISB11= INPUT_PIN; \ + TRISBbits.TRISB12= INPUT_PIN; \ + TRISBbits.TRISB13= INPUT_PIN; \ + TRISCbits.TRISC1= OUTPUT_PIN; \ + TRISCbits.TRISC2= OUTPUT_PIN; \ + TRISCbits.TRISC3= OUTPUT_PIN; \ + TRISCbits.TRISC4= INPUT_PIN; \ + TRISDbits.TRISD0= INPUT_PIN; \ + TRISDbits.TRISD1= OUTPUT_PIN; \ + TRISDbits.TRISD2= OUTPUT_PIN; \ + TRISDbits.TRISD3= OUTPUT_PIN; \ + TRISDbits.TRISD4= OUTPUT_PIN; \ + TRISDbits.TRISD5= INPUT_PIN; \ + TRISDbits.TRISD6= INPUT_PIN; \ + TRISDbits.TRISD7= OUTPUT_PIN; \ + TRISDbits.TRISD8= INPUT_PIN; \ + TRISDbits.TRISD11= OUTPUT_PIN; \ + TRISDbits.TRISD12= INPUT_PIN; \ + TRISEbits.TRISE0= OUTPUT_PIN; \ + TRISEbits.TRISE1= OUTPUT_PIN; \ + TRISEbits.TRISE2= OUTPUT_PIN; \ + TRISEbits.TRISE3= OUTPUT_PIN; \ + TRISEbits.TRISE5= OUTPUT_PIN; \ + TRISEbits.TRISE6= INPUT_PIN; \ + TRISEbits.TRISE7= INPUT_PIN; \ + TRISEbits.TRISE8= OUTPUT_PIN; \ + TRISEbits.TRISE9= OUTPUT_PIN; \ + TRISFbits.TRISF0= INPUT_PIN; \ + TRISFbits.TRISF1= INPUT_PIN; \ + TRISFbits.TRISF2= INPUT_PIN; \ + TRISFbits.TRISF8= OUTPUT_PIN; \ + TRISGbits.TRISG6= OUTPUT_PIN; \ + TRISGbits.TRISG12= INPUT_PIN; \ + TRISGbits.TRISG13= INPUT_PIN; \ + TRISGbits.TRISG9= OUTPUT_PIN; \ + LATBbits.LATB9= LOW; \ + LATCbits.LATC1= LOW; \ + LATCbits.LATC2= LOW; \ + LATCbits.LATC3= LOW; \ + LATDbits.LATD2= WIEGAND_IN_1; \ + LATDbits.LATD3= WIEGAND_IN_0; \ + LATEbits.LATE5= LOW; \ + LATEbits.LATE9= HIGH; \ + } + +// uart3 (CLI/API) speed +#define BAUDRATE3 115200UL +#define BRG_DIV3 4 +#define BRGH3 1 + +// spi for potentiometer +#define SPI_POT SPI_CHANNEL4 +#define SPI_POT_BUFF SPI4BUF +#define SPI_POT_STAT SPI4STATbits + +// spi for sd card - defines required for Microchip SD-SPI libs +// define interface type +#define USE_SD_INTERFACE_WITH_SPI + +#define MDD_USE_SPI_1 +#define SPI_START_CFG_1 (PRI_PRESCAL_64_1 | SEC_PRESCAL_8_1 | MASTER_ENABLE_ON | SPI_CKE_ON | SPI_SMP_ON) +#define SPI_START_CFG_2 (SPI_ENABLE) +// Define the SPI frequency +#define SPI_FREQUENCY (20000000) +// Description: SD-SPI Card Detect Input bit +#define SD_CD PORTFbits.RF0 +// Description: SD-SPI Card Detect TRIS bit +#define SD_CD_TRIS TRISFbits.TRISF0 +// Description: SD-SPI Write Protect Check Input bit +#define SD_WE PORTFbits.RF1 +// Description: SD-SPI Write Protect Check TRIS bit +#define SD_WE_TRIS TRISFbits.TRISF1 +// Description: The main SPI control register +#define SPICON1 SPI1CON +// Description: The SPI status register +#define SPISTAT SPI1STAT +// Description: The SPI Buffer +#define SPIBUF SPI1BUF +// Description: The receive buffer full bit in the SPI status register +#define SPISTAT_RBF SPI1STATbits.SPIRBF +// Description: The bitwise define for the SPI control register (i.e. _____bits) +#define SPICON1bits SPI1CONbits +// Description: The bitwise define for the SPI status register (i.e. _____bits) +#define SPISTATbits SPI1STATbits +// Description: The enable bit for the SPI module +#define SPIENABLE SPICON1bits.ON +// Description: The definition for the SPI baud rate generator register (PIC32) +#define SPIBRG SPI1BRG +// Description: The TRIS bit for the SCK pin +#define SPICLOCK TRISDbits.TRISD10 +// Description: The TRIS bit for the SDI pin +#define SPIIN TRISCbits.TRISC4 +// Description: The TRIS bit for the SDO pin +#define SPIOUT TRISDbits.TRISD0 +#define SD_CS LATDbits.LATD9 +// Description: SD-SPI Chip Select TRIS bit +#define SD_CS_TRIS TRISDbits.TRISD9 +//SPI library functions +#define putcSPI putcSPI1 +#define getcSPI getcSPI1 +#define OpenSPI(config1, config2) OpenSPI1(config1, config2) + +// Define setup parameters for OpenADC10 function +// Turn module on | Ouput in integer format | Trigger mode auto | Enable autosample +#define ADC_CONFIG1 (ADC_FORMAT_INTG | ADC_CLK_AUTO | ADC_AUTO_SAMPLING_ON) +// ADC ref external | Disable offset test | Disable scan mode | Perform 2 samples | Use dual buffers | Use alternate mode +#define ADC_CONFIG2 (ADC_VREF_AVDD_AVSS | ADC_OFFSET_CAL_DISABLE | ADC_SCAN_OFF | ADC_SAMPLES_PER_INT_1 | ADC_ALT_BUF_ON | ADC_ALT_INPUT_ON) + +// Use ADC internal clock | Set sample time +#define ADC_CONFIG3 (ADC_CONV_CLK_INTERNAL_RC | ADC_SAMPLE_TIME_0) + +// slow sample rate for tuning coils +#define ADC_CONFIG2_SLOW (ADC_VREF_AVDD_AVSS | ADC_OFFSET_CAL_DISABLE | ADC_SCAN_OFF | ADC_SAMPLES_PER_INT_16 | ADC_ALT_BUF_ON | ADC_ALT_INPUT_ON) +#define ADC_CONFIG3_SLOW (ADC_CONV_CLK_INTERNAL_RC | ADC_SAMPLE_TIME_31) + +// use AN11 +#define ADC_CONFIGPORT ENABLE_AN11_ANA +// Do not assign channels to scan +#define ADC_CONFIGSCAN SKIP_SCAN_ALL + +#define ADC_TO_VOLTS 0.003208F + + +// flash memory - int myvar = *(int*)(myflashmemoryaddress); + +// memory is 0x9D005000 to 0x9D07FFFF + +#define NVM_MEMORY_END 0x9D07FFFF +#define NVM_PAGE_SIZE 4096 +#define NVM_PAGES 2 // config & VTAG +#define RFIDLER_NVM_ADDRESS (NVM_MEMORY_END - (NVM_PAGE_SIZE * NVM_PAGES)) + +// UART timeout in us +#define SERIAL_TIMEOUT 100 + +#endif diff --git a/tools/hitag2crack/crack5/Makefile b/tools/hitag2crack/crack5/Makefile new file mode 100644 index 000000000..e38c821d1 --- /dev/null +++ b/tools/hitag2crack/crack5/Makefile @@ -0,0 +1,18 @@ +CFLAGS?=-Wall +LIBS=-lpthread + +all: ht2crack5.c utilpart.o ht2crack2utils.o hitagcrypto.o + $(CC) $(CFLAGS) -O3 ht2crack5.c -o ht2crack5 utilpart.o ht2crack2utils.o hitagcrypto.o $(LIBS) + +utilpart.o: util.h utilpart.c + $(CC) $(CFLAGS) -c utilpart.c + +hitagcrypto.o: hitagcrypto.h hitagcrypto.c + $(CC) $(CFLAGS) -c hitagcrypto.c + +ht2crack2utils.o: ht2crack2utils.h ht2crack2utils.c + $(CC) $(CFLAGS) -c ht2crack2utils.c + +clean: + rm -f *.o ht2crack5 +fresh: clean all diff --git a/tools/hitag2crack/crack5/README.md b/tools/hitag2crack/crack5/README.md new file mode 100644 index 000000000..5c900122b --- /dev/null +++ b/tools/hitag2crack/crack5/README.md @@ -0,0 +1,23 @@ +ht2crack5 + + + +Build +----- + +``` +make clean +make +``` + +Run +--- + +You'll need just two nR aR pairs. These are the +encrypted nonces and challenge response values. They should be in hex. + +``` +./ht2crack5 +``` + +UID is the UID of the tag that you used to gather the nR aR values. diff --git a/tools/hitag2crack/crack5/hitagcrypto.c b/tools/hitag2crack/crack5/hitagcrypto.c new file mode 100644 index 000000000..47449c3e3 --- /dev/null +++ b/tools/hitag2crack/crack5/hitagcrypto.c @@ -0,0 +1,373 @@ +/*************************************************************************** + * A copy of the GNU GPL is appended to this file. * + * * + * This licence is based on the nmap licence, and we express our gratitude * + * for the work that went into producing it. There is no other connection * + * between RFIDler and nmap either expressed or implied. * + * * + ********************** IMPORTANT RFIDler LICENSE TERMS ******************** + * * + * * + * All references to RFIDler herein imply all it's derivatives, namely: * + * * + * o RFIDler-LF Standard * + * o RFIDler-LF Lite * + * o RFIDler-LF Nekkid * + * * + * * + * RFIDler is (C) 2013-2015 Aperture Labs Ltd. * + * * + * This program is free software; you may redistribute and/or modify it * + * under the terms of the GNU General Public License as published by the * + * Free Software Foundation; Version 2 ("GPL"), BUT ONLY WITH ALL OF THE * + * CLARIFICATIONS AND EXCEPTIONS DESCRIBED HEREIN. This guarantees your * + * right to use, modify, and redistribute this software under certain * + * conditions. If you wish to embed RFIDler technology into proprietary * + * software or hardware, we sell alternative licenses * + * (contact sales@aperturelabs.com). * + * * + * Note that the GPL places important restrictions on "derivative works", * + * yet it does not provide a detailed definition of that term. To avoid * + * misunderstandings, we interpret that term as broadly as copyright law * + * allows. For example, we consider an application to constitute a * + * derivative work for the purpose of this license if it does any of the * + * following with any software or content covered by this license * + * ("Covered Software"): * + * * + * o Integrates source code from Covered Software. * + * * + * o Is designed specifically to execute Covered Software and parse the * + * results (as opposed to typical shell or execution-menu apps, which will * + * execute anything you tell them to). * + * * + * o Includes Covered Software in a proprietary executable installer. The * + * installers produced by InstallShield are an example of this. Including * + * RFIDler with other software in compressed or archival form does not * + * trigger this provision, provided appropriate open source decompression * + * or de-archiving software is widely available for no charge. For the * + * purposes of this license, an installer is considered to include Covered * + * Software even if it actually retrieves a copy of Covered Software from * + * another source during runtime (such as by downloading it from the * + * Internet). * + * * + * o Links (statically or dynamically) to a library which does any of the * + * above. * + * * + * o Executes a helper program, module, or script to do any of the above. * + * * + * This list is not exclusive, but is meant to clarify our interpretation * + * of derived works with some common examples. Other people may interpret * + * the plain GPL differently, so we consider this a special exception to * + * the GPL that we apply to Covered Software. Works which meet any of * + * these conditions must conform to all of the terms of this license, * + * particularly including the GPL Section 3 requirements of providing * + * source code and allowing free redistribution of the work as a whole. * + * * + * As another special exception to the GPL terms, Aperture Labs Ltd. grants* + * permission to link the code of this program with any version of the * + * OpenSSL library which is distributed under a license identical to that * + * listed in the included docs/licenses/OpenSSL.txt file, and distribute * + * linked combinations including the two. * + * * + * Any redistribution of Covered Software, including any derived works, * + * must obey and carry forward all of the terms of this license, including * + * obeying all GPL rules and restrictions. For example, source code of * + * the whole work must be provided and free redistribution must be * + * allowed. All GPL references to "this License", are to be treated as * + * including the terms and conditions of this license text as well. * + * * + * Because this license imposes special exceptions to the GPL, Covered * + * Work may not be combined (even as part of a larger work) with plain GPL * + * software. The terms, conditions, and exceptions of this license must * + * be included as well. This license is incompatible with some other open * + * source licenses as well. In some cases we can relicense portions of * + * RFIDler or grant special permissions to use it in other open source * + * software. Please contact sales@aperturelabs.com with any such requests.* + * Similarly, we don't incorporate incompatible open source software into * + * Covered Software without special permission from the copyright holders. * + * * + * If you have any questions about the licensing restrictions on using * + * RFIDler in other works, are happy to help. As mentioned above, we also * + * offer alternative license to integrate RFIDler into proprietary * + * applications and appliances. These contracts have been sold to dozens * + * of software vendors, and generally include a perpetual license as well * + * as providing for priority support and updates. They also fund the * + * continued development of RFIDler. Please email sales@aperturelabs.com * + * for further information. * + * If you have received a written license agreement or contract for * + * Covered Software stating terms other than these, you may choose to use * + * and redistribute Covered Software under those terms instead of these. * + * * + * Source is provided to this software because we believe users have a * + * right to know exactly what a program is going to do before they run it. * + * This also allows you to audit the software for security holes (none * + * have been found so far). * + * * + * Source code also allows you to port RFIDler to new platforms, fix bugs, * + * and add new features. You are highly encouraged to send your changes * + * to the RFIDler mailing list for possible incorporation into the * + * main distribution. By sending these changes to Aperture Labs Ltd. or * + * one of the Aperture Labs Ltd. development mailing lists, or checking * + * them into the RFIDler source code repository, it is understood (unless * + * you specify otherwise) that you are offering the RFIDler Project * + * (Aperture Labs Ltd.) the unlimited, non-exclusive right to reuse, * + * modify, and relicense the code. RFIDler will always be available Open * + * Source, but this is important because the inability to relicense code * + * has caused devastating problems for other Free Software projects (such * + * as KDE and NASM). We also occasionally relicense the code to third * + * parties as discussed above. If you wish to specify special license * + * conditions of your contributions, just say so when you send them. * + * * + * This program 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 RFIDler * + * license file for more details (it's in a COPYING file included with * + * RFIDler, and also available from * + * https://github.com/ApertureLabsLtd/RFIDler/COPYING * + * * + ***************************************************************************/ + +// Author: unknown. +// Modifications for RFIDler: Tony Naggs , Adam Laurie + +// uncomment this to build file as a standalone crypto test program +// #define UNIT_TEST +// also uncomment to include verbose debug prints +// #define TEST_DEBUG + +//#include +#include "HardwareProfile.h" +#include "rfidler.h" +#include "hitagcrypto.h" +#include "util.h" + +#ifdef UNIT_TEST +#include +#endif + +#if defined(UNIT_TEST) && defined(TEST_DEBUG) +// Note that printf format %I64x prints 64 bit ints in MS Visual C/C++. +// This may need changing for other compilers/platforms. +#define DEBUG_PRINTF(...) printf(__VA_ARGS__) +#else +#define DEBUG_PRINTF(...) +#endif + + +/* Brief info about NXP Hitag 1, Hitag 2, Hitag S and Hitag u (mu) + + Hitag 125kHz RFID was created by a company called Mikron (Mikron Gesellschaft + fur Integrierte Mikroelektronik Mbh), of Austria, for micropayment applications. + At about the same time, late 1980s to early 1990s, Mikron developed the + similarly featured Mifare micropayment card for 13.56MHz RFID. + (Mikron's European Patent EP 0473569 A2 was filed 23 August 1991, with a + priority date of 23 Aug 1990.) + Mikron was subsequently acquired by Philips Semiconductors in 1995. + Philips Semiconductors divsion subsequently became NXP. + + + Modulation read/write device -> transponder: 100 % ASK and binary pulse + length coding + + Modulation transponder -> read/write device: Strong ASK modulation, + selectable Manchester or Biphase coding + + Hitag S, Hitag u; anti-collision procedure + + Fast anti-collision protocol + + Hitag u; optional Cyclic Redundancy Check (CRC) + + Reader Talks First mode + + Hitag 2 & later; Transponder Talks First (TTF) mode + + Temporary switch from Transponder Talks First into Reader Talks First + (RTF) Mode + + Data rate read/write device to transponder: 5.2 kbit/s + + Data rates transponder to read/write device: 2 kbit/s, 4 kbit/s, 8 kbit/s + + 32-bit password feature + + Hitag 2, S = 32-bit Unique Identifier + + Hitag u = 48-bit Unique Identifier + + Selectable password modes for reader / tag mutual authentication + (Hitag 1 has 2 pairs of keys, later versions have 1 pair) + + Hitag 2 & Hitag S; Selectable encrypted mode, 48 bit key + + Known tag types: + + HITAG 1 2048 bits total memory + + HITAG 2 256 Bit total memory Read/Write + 8 pages of 32 bits, inc UID (32), + secret key (64), password (24), config (8) + + HITAG S 32 32 bits Unique Identifier Read Only + HITAG S 256 256 bits total memory Read/Write + HITAG S 2048 2048 bits total memory Read/Write + + HITAG u RO64 64 bits total memory Read Only + HITAG u 128 bits total memory Read/Write + HITAG u Advanced 512 bits total memory Read/Write + HITAG u Advanced+ 1760 bits total memory Read/Write + + Default 48-bit key for Hitag 2, S encryption: + "MIKRON" = O N M I K R + Key = 4F 4E 4D 49 4B 52 + +*/ + + +// We want the crypto functions to be as fast as possible, so optimize! +// The best compiler optimization in Microchip's free XC32 edition is -O1 +#pragma GCC optimize("O1") + +// private, nonlinear function to generate 1 crypto bit +static uint32_t hitag2_crypt(uint64_t x); + + +// macros to pick out 4 bits in various patterns of 1s & 2s & make a new number +#define pickbits2_2(S, A, B) ( ((S >> A) & 3) | ((S >> (B - 2)) & 0xC) ) +#define pickbits1x4(S, A, B, C, D) ( ((S >> A) & 1) | ((S >> (B - 1)) & 2) | \ + ((S >> (C - 2)) & 4) | ((S >> (D - 3)) & 8) ) +#define pickbits1_1_2(S, A, B, C) ( ((S >> A) & 1) | ((S >> (B - 1)) & 2) | \ + ((S >> (C - 2)) & 0xC) ) +#define pickbits2_1_1(S, A, B, C) ( ((S >> A) & 3) | ((S >> (B - 2)) & 4) | \ + ((S >> (C - 3)) & 8) ) +#define pickbits1_2_1(S, A, B, C) ( ((S >> A) & 1) | ((S >> (B - 1)) & 6) | \ + ((S >> (C - 3)) & 8) ) + + +static uint32_t hitag2_crypt(uint64_t x) { + const uint32_t ht2_function4a = 0x2C79; // 0010 1100 0111 1001 + const uint32_t ht2_function4b = 0x6671; // 0110 0110 0111 0001 + const uint32_t ht2_function5c = 0x7907287B; // 0111 1001 0000 0111 0010 1000 0111 1011 + uint32_t bitindex; + + bitindex = (ht2_function4a >> pickbits2_2(x, 1, 4)) & 1; + bitindex |= ((ht2_function4b << 1) >> pickbits1_1_2(x, 7, 11, 13)) & 0x02; + bitindex |= ((ht2_function4b << 2) >> pickbits1x4(x, 16, 20, 22, 25)) & 0x04; + bitindex |= ((ht2_function4b << 3) >> pickbits2_1_1(x, 27, 30, 32)) & 0x08; + bitindex |= ((ht2_function4a << 4) >> pickbits1_2_1(x, 33, 42, 45)) & 0x10; + + DEBUG_PRINTF("hitag2_crypt bitindex = %02x\n", bitindex); + return (ht2_function5c >> bitindex) & 1; +} + +/* + * Parameters: + * Hitag_State* pstate - output, internal state after initialisation + * uint64_t sharedkey - 48 bit key shared between reader & tag + * uint32_t serialnum - 32 bit tag serial number + * uint32_t initvector - 32 bit random IV from reader, part of tag authentication + */ +void hitag2_init(Hitag_State *pstate, uint64_t sharedkey, uint32_t serialnum, uint32_t initvector) { + // init state, from serial number and lowest 16 bits of shared key + uint64_t state = ((sharedkey & 0xFFFF) << 32) | serialnum; + + // mix the initialisation vector and highest 32 bits of the shared key + initvector ^= (uint32_t)(sharedkey >> 16); + + // move 16 bits from (IV xor Shared Key) to top of uint64_t state + // these will be XORed in turn with output of the crypto function + state |= (uint64_t) initvector << 48; + initvector >>= 16; + + // unrolled loop is faster on PIC32 (MIPS), do 32 times + // shift register, then calc new bit + state >>= 1; + state = (state >> 1) ^ (uint64_t) hitag2_crypt(state) << 46; + state = (state >> 1) ^ (uint64_t) hitag2_crypt(state) << 46; + state = (state >> 1) ^ (uint64_t) hitag2_crypt(state) << 46; + state = (state >> 1) ^ (uint64_t) hitag2_crypt(state) << 46; + state = (state >> 1) ^ (uint64_t) hitag2_crypt(state) << 46; + state = (state >> 1) ^ (uint64_t) hitag2_crypt(state) << 46; + state = (state >> 1) ^ (uint64_t) hitag2_crypt(state) << 46; + state = (state >> 1) ^ (uint64_t) hitag2_crypt(state) << 46; + + state = (state >> 1) ^ (uint64_t) hitag2_crypt(state) << 46; + state = (state >> 1) ^ (uint64_t) hitag2_crypt(state) << 46; + state = (state >> 1) ^ (uint64_t) hitag2_crypt(state) << 46; + state = (state >> 1) ^ (uint64_t) hitag2_crypt(state) << 46; + state = (state >> 1) ^ (uint64_t) hitag2_crypt(state) << 46; + state = (state >> 1) ^ (uint64_t) hitag2_crypt(state) << 46; + state = (state >> 1) ^ (uint64_t) hitag2_crypt(state) << 46; + state = (state >> 1) ^ (uint64_t) hitag2_crypt(state) << 46; + + // highest 16 bits of IV XOR Shared Key + state |= (uint64_t) initvector << 47; + + state = (state >> 1) ^ (uint64_t) hitag2_crypt(state) << 46; + state = (state >> 1) ^ (uint64_t) hitag2_crypt(state) << 46; + state = (state >> 1) ^ (uint64_t) hitag2_crypt(state) << 46; + state = (state >> 1) ^ (uint64_t) hitag2_crypt(state) << 46; + state = (state >> 1) ^ (uint64_t) hitag2_crypt(state) << 46; + state = (state >> 1) ^ (uint64_t) hitag2_crypt(state) << 46; + state = (state >> 1) ^ (uint64_t) hitag2_crypt(state) << 46; + state = (state >> 1) ^ (uint64_t) hitag2_crypt(state) << 46; + + state = (state >> 1) ^ (uint64_t) hitag2_crypt(state) << 46; + state = (state >> 1) ^ (uint64_t) hitag2_crypt(state) << 46; + state = (state >> 1) ^ (uint64_t) hitag2_crypt(state) << 46; + state = (state >> 1) ^ (uint64_t) hitag2_crypt(state) << 46; + state = (state >> 1) ^ (uint64_t) hitag2_crypt(state) << 46; + state = (state >> 1) ^ (uint64_t) hitag2_crypt(state) << 46; + state = (state >> 1) ^ (uint64_t) hitag2_crypt(state) << 46; + state ^= (uint64_t) hitag2_crypt(state) << 47; + + DEBUG_PRINTF("hitag2_init result = %012I64x\n", state); + pstate->shiftreg = state; + /* naive version for reference, LFSR has 16 taps + pstate->lfsr = state ^ (state >> 2) ^ (state >> 3) ^ (state >> 6) + ^ (state >> 7) ^ (state >> 8) ^ (state >> 16) ^ (state >> 22) + ^ (state >> 23) ^ (state >> 26) ^ (state >> 30) ^ (state >> 41) + ^ (state >> 42) ^ (state >> 43) ^ (state >> 46) ^ (state >> 47); + */ + { + // optimise with one 64-bit intermediate + uint64_t temp = state ^ (state >> 1); + pstate->lfsr = state ^ (state >> 6) ^ (state >> 16) + ^ (state >> 26) ^ (state >> 30) ^ (state >> 41) + ^ (temp >> 2) ^ (temp >> 7) ^ (temp >> 22) + ^ (temp >> 42) ^ (temp >> 46); + } +} + + +/* + * Return up to 32 crypto bits. + * Last bit is in least significant bit, earlier bits are shifted left. + * Note that the Hitag transmission protocol is least significant bit, + * so we may want to change this, or add a function, that returns the + * crypto output bits in the other order. + * + * Parameters: + * Hitag_State* pstate - in/out, internal cipher state after initialisation + * uint32_t steps - number of bits requested, (capped at 32) + */ +uint32_t hitag2_nstep(Hitag_State *pstate, uint32_t steps) { + uint64_t state = pstate->shiftreg; + uint32_t result = 0; + uint64_t lfsr = pstate->lfsr; + + if (steps == 0) + return 0; + +// if (steps > 32) +// steps = 32; + + do { + // update shift registers + if (lfsr & 1) { + state = (state >> 1) | 0x800000000000; + lfsr = (lfsr >> 1) ^ 0xB38083220073; + + // accumulate next bit of crypto + result = (result << 1) | hitag2_crypt(state); + } else { + state >>= 1; + lfsr >>= 1; + + result = (result << 1) | hitag2_crypt(state); + } + } while (--steps); + + DEBUG_PRINTF("hitag2_nstep state = %012I64x, result %02x\n", state, result); + pstate->shiftreg = state; + pstate->lfsr = lfsr; + return result; +} + +// end of crypto core, revert to default optimization level +#pragma GCC reset_options diff --git a/tools/hitag2crack/crack5/hitagcrypto.h b/tools/hitag2crack/crack5/hitagcrypto.h new file mode 100644 index 000000000..274d3d82c --- /dev/null +++ b/tools/hitag2crack/crack5/hitagcrypto.h @@ -0,0 +1,167 @@ +/*************************************************************************** + * A copy of the GNU GPL is appended to this file. * + * * + * This licence is based on the nmap licence, and we express our gratitude * + * for the work that went into producing it. There is no other connection * + * between RFIDler and nmap either expressed or implied. * + * * + ********************** IMPORTANT RFIDler LICENSE TERMS ******************** + * * + * * + * All references to RFIDler herein imply all it's derivatives, namely: * + * * + * o RFIDler-LF Standard * + * o RFIDler-LF Lite * + * o RFIDler-LF Nekkid * + * * + * * + * RFIDler is (C) 2013-2014 Aperture Labs Ltd. * + * * + * This program is free software; you may redistribute and/or modify it * + * under the terms of the GNU General Public License as published by the * + * Free Software Foundation; Version 2 ("GPL"), BUT ONLY WITH ALL OF THE * + * CLARIFICATIONS AND EXCEPTIONS DESCRIBED HEREIN. This guarantees your * + * right to use, modify, and redistribute this software under certain * + * conditions. If you wish to embed RFIDler technology into proprietary * + * software or hardware, we sell alternative licenses * + * (contact sales@aperturelabs.com). * + * * + * Note that the GPL places important restrictions on "derivative works", * + * yet it does not provide a detailed definition of that term. To avoid * + * misunderstandings, we interpret that term as broadly as copyright law * + * allows. For example, we consider an application to constitute a * + * derivative work for the purpose of this license if it does any of the * + * following with any software or content covered by this license * + * ("Covered Software"): * + * * + * o Integrates source code from Covered Software. * + * * + * o Is designed specifically to execute Covered Software and parse the * + * results (as opposed to typical shell or execution-menu apps, which will * + * execute anything you tell them to). * + * * + * o Includes Covered Software in a proprietary executable installer. The * + * installers produced by InstallShield are an example of this. Including * + * RFIDler with other software in compressed or archival form does not * + * trigger this provision, provided appropriate open source decompression * + * or de-archiving software is widely available for no charge. For the * + * purposes of this license, an installer is considered to include Covered * + * Software even if it actually retrieves a copy of Covered Software from * + * another source during runtime (such as by downloading it from the * + * Internet). * + * * + * o Links (statically or dynamically) to a library which does any of the * + * above. * + * * + * o Executes a helper program, module, or script to do any of the above. * + * * + * This list is not exclusive, but is meant to clarify our interpretation * + * of derived works with some common examples. Other people may interpret * + * the plain GPL differently, so we consider this a special exception to * + * the GPL that we apply to Covered Software. Works which meet any of * + * these conditions must conform to all of the terms of this license, * + * particularly including the GPL Section 3 requirements of providing * + * source code and allowing free redistribution of the work as a whole. * + * * + * As another special exception to the GPL terms, Aperture Labs Ltd. grants* + * permission to link the code of this program with any version of the * + * OpenSSL library which is distributed under a license identical to that * + * listed in the included docs/licenses/OpenSSL.txt file, and distribute * + * linked combinations including the two. * + * * + * Any redistribution of Covered Software, including any derived works, * + * must obey and carry forward all of the terms of this license, including * + * obeying all GPL rules and restrictions. For example, source code of * + * the whole work must be provided and free redistribution must be * + * allowed. All GPL references to "this License", are to be treated as * + * including the terms and conditions of this license text as well. * + * * + * Because this license imposes special exceptions to the GPL, Covered * + * Work may not be combined (even as part of a larger work) with plain GPL * + * software. The terms, conditions, and exceptions of this license must * + * be included as well. This license is incompatible with some other open * + * source licenses as well. In some cases we can relicense portions of * + * RFIDler or grant special permissions to use it in other open source * + * software. Please contact sales@aperturelabs.com with any such requests.* + * Similarly, we don't incorporate incompatible open source software into * + * Covered Software without special permission from the copyright holders. * + * * + * If you have any questions about the licensing restrictions on using * + * RFIDler in other works, are happy to help. As mentioned above, we also * + * offer alternative license to integrate RFIDler into proprietary * + * applications and appliances. These contracts have been sold to dozens * + * of software vendors, and generally include a perpetual license as well * + * as providing for priority support and updates. They also fund the * + * continued development of RFIDler. Please email sales@aperturelabs.com * + * for further information. * + * If you have received a written license agreement or contract for * + * Covered Software stating terms other than these, you may choose to use * + * and redistribute Covered Software under those terms instead of these. * + * * + * Source is provided to this software because we believe users have a * + * right to know exactly what a program is going to do before they run it. * + * This also allows you to audit the software for security holes (none * + * have been found so far). * + * * + * Source code also allows you to port RFIDler to new platforms, fix bugs, * + * and add new features. You are highly encouraged to send your changes * + * to the RFIDler mailing list for possible incorporation into the * + * main distribution. By sending these changes to Aperture Labs Ltd. or * + * one of the Aperture Labs Ltd. development mailing lists, or checking * + * them into the RFIDler source code repository, it is understood (unless * + * you specify otherwise) that you are offering the RFIDler Project * + * (Aperture Labs Ltd.) the unlimited, non-exclusive right to reuse, * + * modify, and relicense the code. RFIDler will always be available Open * + * Source, but this is important because the inability to relicense code * + * has caused devastating problems for other Free Software projects (such * + * as KDE and NASM). We also occasionally relicense the code to third * + * parties as discussed above. If you wish to specify special license * + * conditions of your contributions, just say so when you send them. * + * * + * This program 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 RFIDler * + * license file for more details (it's in a COPYING file included with * + * RFIDler, and also available from * + * https://github.com/ApertureLabsLtd/RFIDler/COPYING * + * * + ***************************************************************************/ + +// Author: unknown. +// Modifications for RFIDler: Tony Naggs , Adam Laurie + + +#ifndef HITAGCRYPTO_H +#define HITAGCRYPTO_H + +#include + +/* + Our model of Hitag 2 crypto uses 2 parallel shift registers: + a. 48 bit Feedback Shift Register, required for inputs to the nonlinear function. + b. 48 bit Linear Feedback Shift Register (LFSR). + A transform of initial register (a) value, which is then run in parallel. + Enables much faster calculation of the feedback values. + + API: + void hitag2_init(Hitag_State* pstate, uint64_t sharedkey, uint32_t serialnum, + uint32_t initvector); + Initialise state from 48 bit shared (secret) reader/tag key, + 32 bit tag serial number and 32 bit initialisation vector from reader. + + uint32_t hitag2_nstep(Hitag_State* pstate, uint32_t steps); + update shift register state and generate N cipher bits (N should be <= 32) + */ + + +typedef struct { + uint64_t shiftreg; // naive shift register, required for nonlinear fn input + uint64_t lfsr; // fast lfsr, used to make software faster +} Hitag_State; + +void hitag2_init(Hitag_State *pstate, uint64_t sharedkey, uint32_t serialnum, uint32_t initvector); + +uint32_t hitag2_nstep(Hitag_State *pstate, uint32_t steps); + +#endif /* HITAGCRYPTO_H */ + diff --git a/tools/hitag2crack/crack5/ht2crack2utils.c b/tools/hitag2crack/crack5/ht2crack2utils.c new file mode 100644 index 000000000..75d4c27a9 --- /dev/null +++ b/tools/hitag2crack/crack5/ht2crack2utils.c @@ -0,0 +1,172 @@ +#include "ht2crack2utils.h" + +// writes a value into a buffer as a series of bytes +void writebuf(unsigned char *buf, uint64_t val, unsigned int len) { + int i; + char c; + + for (i = len - 1; i >= 0; i--) { + c = val & 0xff; + buf[i] = c; + val = val >> 8; + } + +} + + +/* simple hexdump for testing purposes */ +void shexdump(unsigned char *data, int data_len) { + int i; + + if (!data || (data_len <= 0)) { + printf("shexdump: invalid parameters\n"); + return; + } + + printf("Hexdump from %p:\n", data); + + for (i = 0; i < data_len; i++) { + if ((i % HEX_PER_ROW) == 0) { + printf("\n0x%04x: ", i); + } + printf("%02x ", data[i]); + } + printf("\n\n"); +} + + + +void printbin(unsigned char *c) { + int i, j; + unsigned char x; + + if (!c) { + printf("printbin: invalid params\n"); + return; + } + + for (i = 0; i < 6; i++) { + x = c[i]; + for (j = 0; j < 8; j++) { + printf("%d", (x & 0x80) >> 7); + x = x << 1; + } + } + printf("\n"); +} + + +void printbin2(uint64_t val, unsigned int size) { + int i; + uint64_t mask = 1; + + mask = mask << (size - 1); + + for (i = 0; i < size; i++) { + if (val & mask) { + printf("1"); + } else { + printf("0"); + } + val = val << 1; + } +} + + +void printstate(Hitag_State *hstate) { + printf("shiftreg =\t"); + printbin2(hstate->shiftreg, 48); + printf("\n"); +} + + + + +// convert hex char to binary +unsigned char hex2bin(unsigned char c) { + if ((c >= '0') && (c <= '9')) { + return (c - '0'); + } else if ((c >= 'a') && (c <= 'f')) { + return (c - 'a' + 10); + } else if ((c >= 'A') && (c <= 'F')) { + return (c - 'A' + 10); + } else { + return 0; + } +} + +// return a single bit from a value +int bitn(uint64_t x, int bit) { + uint64_t bitmask = 1; + + bitmask = bitmask << bit; + + if (x & bitmask) { + return 1; + } else { + return 0; + } +} + + +// the sub-function R that rollback depends upon +int fnR(uint64_t x) { + // renumbered bits because my state is 0-47, not 1-48 + return (bitn(x, 1) ^ bitn(x, 2) ^ bitn(x, 5) ^ bitn(x, 6) ^ bitn(x, 7) ^ + bitn(x, 15) ^ bitn(x, 21) ^ bitn(x, 22) ^ bitn(x, 25) ^ bitn(x, 29) ^ bitn(x, 40) ^ + bitn(x, 41) ^ bitn(x, 42) ^ bitn(x, 45) ^ bitn(x, 46) ^ bitn(x, 47)); +} + +// the rollback function that lets us go backwards in time +void rollback(Hitag_State *hstate, unsigned int steps) { + int i; + + for (i = 0; i < steps; i++) { + hstate->shiftreg = ((hstate->shiftreg << 1) & 0xffffffffffff) | fnR(hstate->shiftreg); + } + +} + + +// the three filter sub-functions that feed fnf +int fa(unsigned int i) { + return bitn(0x2C79, i); +} + +int fb(unsigned int i) { + return bitn(0x6671, i); +} + +int fc(unsigned int i) { + return bitn(0x7907287B, i); +} + +// the filter function that generates a bit of output from the prng state +int fnf(uint64_t s) { + unsigned int x1, x2, x3, x4, x5, x6; + + x1 = (bitn(s, 2) << 0) | (bitn(s, 3) << 1) | (bitn(s, 5) << 2) | (bitn(s, 6) << 3); + x2 = (bitn(s, 8) << 0) | (bitn(s, 12) << 1) | (bitn(s, 14) << 2) | (bitn(s, 15) << 3); + x3 = (bitn(s, 17) << 0) | (bitn(s, 21) << 1) | (bitn(s, 23) << 2) | (bitn(s, 26) << 3); + x4 = (bitn(s, 28) << 0) | (bitn(s, 29) << 1) | (bitn(s, 31) << 2) | (bitn(s, 33) << 3); + x5 = (bitn(s, 34) << 0) | (bitn(s, 43) << 1) | (bitn(s, 44) << 2) | (bitn(s, 46) << 3); + + x6 = (fa(x1) << 0) | (fb(x2) << 1) | (fb(x3) << 2) | (fb(x4) << 3) | (fa(x5) << 4); + + return fc(x6); +} + +// builds the lfsr for the prng (quick calcs for hitag2_nstep()) +void buildlfsr(Hitag_State *hstate) { + uint64_t state = hstate->shiftreg; + uint64_t temp; + + temp = state ^ (state >> 1); + hstate->lfsr = state ^ (state >> 6) ^ (state >> 16) + ^ (state >> 26) ^ (state >> 30) ^ (state >> 41) + ^ (temp >> 2) ^ (temp >> 7) ^ (temp >> 22) + ^ (temp >> 42) ^ (temp >> 46); +} + + + diff --git a/tools/hitag2crack/crack5/ht2crack2utils.h b/tools/hitag2crack/crack5/ht2crack2utils.h new file mode 100644 index 000000000..33e0e3036 --- /dev/null +++ b/tools/hitag2crack/crack5/ht2crack2utils.h @@ -0,0 +1,35 @@ +#include +#include +#include +#include +#include +#include +#include +#include +#include +#include + +#include "HardwareProfile.h" +#include "rfidler.h" +#include "util.h" + +#include "hitagcrypto.h" + +#define HEX_PER_ROW 16 + + + +void writebuf(unsigned char *buf, uint64_t val, unsigned int len); +void shexdump(unsigned char *data, int data_len); +void printbin(unsigned char *c); +void printbin2(uint64_t val, unsigned int size); +void printstate(Hitag_State *hstate); +unsigned char hex2bin(unsigned char c); +int bitn(uint64_t x, int bit); +int fnR(uint64_t x); +void rollback(Hitag_State *hstate, unsigned int steps); +int fa(unsigned int i); +int fb(unsigned int i); +int fc(unsigned int i); +int fnf(uint64_t s); +void buildlfsr(Hitag_State *hstate); diff --git a/tools/hitag2crack/crack5/ht2crack5.c b/tools/hitag2crack/crack5/ht2crack5.c new file mode 100644 index 000000000..38e57ffa7 --- /dev/null +++ b/tools/hitag2crack/crack5/ht2crack5.c @@ -0,0 +1,751 @@ +/* ht2crack5.c + * + * This code is heavily based on the HiTag2 Hell CPU implementation + * from https://github.com/factoritbv/hitag2hell by FactorIT B.V., + * with the following changes: + * * Main takes a UID and 2 {nR},{aR} pairs as arguments + * and searches for states producing the first aR sample, + * reconstructs the corresponding key candidates + * and tests them against the second nR,aR pair; + * * Reuses the Hitag helping functions of the other attacks. + */ + +#include +#include +#include +#include +#include +#include +#include +#include +#include "ht2crack2utils.h" + +const uint8_t bits[9] = {20, 14, 4, 3, 1, 1, 1, 1, 1}; +#define lfsr_inv(state) (((state)<<1) | (__builtin_parityll((state) & ((0xce0044c101cd>>1)|(1ull<<(47)))))) +#define i4(x,a,b,c,d) ((uint32_t)((((x)>>(a))&1)<<3)|(((x)>>(b))&1)<<2|(((x)>>(c))&1)<<1|(((x)>>(d))&1)) +#define f(state) ((0xdd3929b >> ( (((0x3c65 >> i4(state, 2, 3, 5, 6) ) & 1) <<4) \ + | ((( 0xee5 >> i4(state, 8,12,14,15) ) & 1) <<3) \ + | ((( 0xee5 >> i4(state,17,21,23,26) ) & 1) <<2) \ + | ((( 0xee5 >> i4(state,28,29,31,33) ) & 1) <<1) \ + | (((0x3c65 >> i4(state,34,43,44,46) ) & 1) ))) & 1) + +#define MAX_BITSLICES 256 +#define VECTOR_SIZE (MAX_BITSLICES/8) + +typedef unsigned int __attribute__((aligned(VECTOR_SIZE))) __attribute__((vector_size(VECTOR_SIZE))) bitslice_value_t; +typedef union { + bitslice_value_t value; + uint64_t bytes64[MAX_BITSLICES / 64]; + uint8_t bytes[MAX_BITSLICES / 8]; +} bitslice_t; + +// we never actually set or use the lowest 2 bits the initial state, so we can save 2 bitslices everywhere +__thread bitslice_t state[-2 + 32 + 48]; + +bitslice_t keystream[32]; +bitslice_t bs_zeroes, bs_ones; + +#define f_a_bs(a,b,c,d) (~(((a|b)&c)^(a|d)^b)) // 6 ops +#define f_b_bs(a,b,c,d) (~(((d|c)&(a^b))^(d|a|b))) // 7 ops +#define f_c_bs(a,b,c,d,e) (~((((((c^e)|d)&a)^b)&(c^b))^(((d^e)|a)&((d^b)|c)))) // 13 ops +#define lfsr_bs(i) (state[-2+i+ 0].value ^ state[-2+i+ 2].value ^ state[-2+i+ 3].value ^ state[-2+i+ 6].value ^ \ + state[-2+i+ 7].value ^ state[-2+i+ 8].value ^ state[-2+i+16].value ^ state[-2+i+22].value ^ \ + state[-2+i+23].value ^ state[-2+i+26].value ^ state[-2+i+30].value ^ state[-2+i+41].value ^ \ + state[-2+i+42].value ^ state[-2+i+43].value ^ state[-2+i+46].value ^ state[-2+i+47].value); +#define get_bit(n, word) ((word >> (n)) & 1) +#define get_vector_bit(slice, value) get_bit(slice&0x3f, value.bytes64[slice>>6]) + +const uint64_t expand(uint64_t mask, uint64_t value) { + uint64_t fill = 0; + for (uint64_t bit_index = 0; bit_index < 48; bit_index++) { + if (mask & 1) { + fill |= (value & 1) << bit_index; + value >>= 1; + } + mask >>= 1; + } + return fill; +} + +void bitslice(const uint64_t value, bitslice_t *restrict bitsliced_value, const size_t bit_len, bool reverse) { + size_t bit_idx; + for (bit_idx = 0; bit_idx < bit_len; bit_idx++) { + bool bit; + if (reverse) { + bit = get_bit(bit_len - 1 - bit_idx, value); + } else { + bit = get_bit(bit_idx, value); + } + if (bit) { + bitsliced_value[bit_idx].value = bs_ones.value; + } else { + bitsliced_value[bit_idx].value = bs_zeroes.value; + } + } +} + +const uint64_t unbitslice(const bitslice_t *restrict b, const uint8_t s, const uint8_t n) { + uint64_t result = 0; + for (uint8_t i = 0; i < n; ++i) { + result <<= 1; + result |= get_vector_bit(s, b[n - 1 - i]); + } + return result; +} + +uint32_t uid, nR1, aR1, nR2, aR2; + +uint64_t candidates[(1 << 20)]; +bitslice_t initial_bitslices[48]; +size_t filter_pos[20] = {4, 7, 9, 13, 16, 18, 22, 24, 27, 30, 32, 35, 45, 47 }; +size_t thread_count = 8; +size_t layer_0_found; +void *find_state(void *thread_d); +static void try_state(uint64_t s); + +int main(int argc, char *argv[]) { + // set constants + memset(bs_ones.bytes, 0xff, VECTOR_SIZE); + memset(bs_zeroes.bytes, 0x00, VECTOR_SIZE); + + uint32_t target = 0; + + if (argc < 6) { + printf("%s UID {nR1} {aR1} {nR2} {aR2}\n", argv[0]); + exit(1); + } + + if (!strncmp(argv[1], "0x", 2) || !strncmp(argv[1], "0X", 2)) { + uid = rev32(hexreversetoulong(argv[1] + 2)); + } else { + uid = rev32(hexreversetoulong(argv[1])); + } + + if (!strncmp(argv[2], "0x", 2) || !strncmp(argv[2], "0X", 2)) { + nR1 = rev32(hexreversetoulong(argv[2] + 2)); + } else { + nR1 = rev32(hexreversetoulong(argv[2])); + } + + aR1 = strtol(argv[3], NULL, 16); + + if (!strncmp(argv[4], "0x", 2) || !strncmp(argv[4], "0X", 2)) { + nR2 = rev32(hexreversetoulong(argv[4] + 2)); + } else { + nR2 = rev32(hexreversetoulong(argv[4])); + } + + aR2 = strtol(argv[5], NULL, 16); + + target = ~aR1; + // bitslice inverse target bits + bitslice(~target, keystream, 32, true); + + // bitslice all possible 256 values in the lowest 8 bits + memset(initial_bitslices[0].bytes, 0xaa, VECTOR_SIZE); + memset(initial_bitslices[1].bytes, 0xcc, VECTOR_SIZE); + memset(initial_bitslices[2].bytes, 0xf0, VECTOR_SIZE); + size_t interval = 1; + for (size_t bit = 3; bit < 8; bit++) { + for (size_t byte = 0; byte < VECTOR_SIZE;) { + for (size_t length = 0; length < interval; length++) { + initial_bitslices[bit].bytes[byte++] = 0x00; + } + for (size_t length = 0; length < interval; length++) { + initial_bitslices[bit].bytes[byte++] = 0xff; + } + } + interval <<= 1; + } + + // compute layer 0 output + for (size_t i0 = 0; i0 < 1 << 20; i0++) { + uint64_t state0 = expand(0x5806b4a2d16c, i0); + + if (f(state0) == target >> 31) { + candidates[layer_0_found++] = state0; + } + } + + // start threads and wait on them + pthread_t thread_handles[thread_count]; + for (size_t thread = 0; thread < thread_count; thread++) { + pthread_create(&thread_handles[thread], NULL, find_state, (void *) thread); + } + for (size_t thread = 0; thread < thread_count; thread++) { + pthread_join(thread_handles[thread], NULL); + } + + printf("Key not found\n"); + exit(1); +} + +void *find_state(void *thread_d) { + size_t thread = (size_t)thread_d; + + for (size_t index = thread; index < layer_0_found; index += thread_count) { + if (((index / thread_count) & 0xFF) == 0) + printf("Thread %lu slice %lu/%lu\n", thread, index / thread_count / 256 + 1, layer_0_found / thread_count / 256); + uint64_t state0 = candidates[index]; + bitslice(state0 >> 2, &state[0], 46, false); + for (size_t bit = 0; bit < 8; bit++) { + state[-2 + filter_pos[bit]] = initial_bitslices[bit]; + } + for (uint16_t i1 = 0; i1 < (1 << (bits[1] + 1) >> 8); i1++) { + state[-2 + 27].value = ((bool)(i1 & 0x1)) ? bs_ones.value : bs_zeroes.value; + state[-2 + 30].value = ((bool)(i1 & 0x2)) ? bs_ones.value : bs_zeroes.value; + state[-2 + 32].value = ((bool)(i1 & 0x4)) ? bs_ones.value : bs_zeroes.value; + state[-2 + 35].value = ((bool)(i1 & 0x8)) ? bs_ones.value : bs_zeroes.value; + state[-2 + 45].value = ((bool)(i1 & 0x10)) ? bs_ones.value : bs_zeroes.value; + state[-2 + 47].value = ((bool)(i1 & 0x20)) ? bs_ones.value : bs_zeroes.value; + state[-2 + 48].value = ((bool)(i1 & 0x40)) ? bs_ones.value : bs_zeroes.value; // guess lfsr output 0 + // 0xfc07fef3f9fe + const bitslice_value_t filter1_0 = f_a_bs(state[-2 + 3].value, state[-2 + 4].value, state[-2 + 6].value, state[-2 + 7].value); + const bitslice_value_t filter1_1 = f_b_bs(state[-2 + 9].value, state[-2 + 13].value, state[-2 + 15].value, state[-2 + 16].value); + const bitslice_value_t filter1_2 = f_b_bs(state[-2 + 18].value, state[-2 + 22].value, state[-2 + 24].value, state[-2 + 27].value); + const bitslice_value_t filter1_3 = f_b_bs(state[-2 + 29].value, state[-2 + 30].value, state[-2 + 32].value, state[-2 + 34].value); + const bitslice_value_t filter1_4 = f_a_bs(state[-2 + 35].value, state[-2 + 44].value, state[-2 + 45].value, state[-2 + 47].value); + const bitslice_value_t filter1 = f_c_bs(filter1_0, filter1_1, filter1_2, filter1_3, filter1_4); + bitslice_t results1; + results1.value = filter1 ^ keystream[1].value; + if (results1.bytes64[0] == 0 + && results1.bytes64[1] == 0 + && results1.bytes64[2] == 0 + && results1.bytes64[3] == 0 + ) { + continue; + } + const bitslice_value_t filter2_0 = f_a_bs(state[-2 + 4].value, state[-2 + 5].value, state[-2 + 7].value, state[-2 + 8].value); + const bitslice_value_t filter2_3 = f_b_bs(state[-2 + 30].value, state[-2 + 31].value, state[-2 + 33].value, state[-2 + 35].value); + const bitslice_value_t filter3_0 = f_a_bs(state[-2 + 5].value, state[-2 + 6].value, state[-2 + 8].value, state[-2 + 9].value); + const bitslice_value_t filter5_2 = f_b_bs(state[-2 + 22].value, state[-2 + 26].value, state[-2 + 28].value, state[-2 + 31].value); + const bitslice_value_t filter6_2 = f_b_bs(state[-2 + 23].value, state[-2 + 27].value, state[-2 + 29].value, state[-2 + 32].value); + const bitslice_value_t filter7_2 = f_b_bs(state[-2 + 24].value, state[-2 + 28].value, state[-2 + 30].value, state[-2 + 33].value); + const bitslice_value_t filter9_1 = f_b_bs(state[-2 + 17].value, state[-2 + 21].value, state[-2 + 23].value, state[-2 + 24].value); + const bitslice_value_t filter9_2 = f_b_bs(state[-2 + 26].value, state[-2 + 30].value, state[-2 + 32].value, state[-2 + 35].value); + const bitslice_value_t filter10_0 = f_a_bs(state[-2 + 12].value, state[-2 + 13].value, state[-2 + 15].value, state[-2 + 16].value); + const bitslice_value_t filter11_0 = f_a_bs(state[-2 + 13].value, state[-2 + 14].value, state[-2 + 16].value, state[-2 + 17].value); + const bitslice_value_t filter12_0 = f_a_bs(state[-2 + 14].value, state[-2 + 15].value, state[-2 + 17].value, state[-2 + 18].value); + for (uint16_t i2 = 0; i2 < (1 << (bits[2] + 1)); i2++) { + state[-2 + 10].value = ((bool)(i2 & 0x1)) ? bs_ones.value : bs_zeroes.value; + state[-2 + 19].value = ((bool)(i2 & 0x2)) ? bs_ones.value : bs_zeroes.value; + state[-2 + 25].value = ((bool)(i2 & 0x4)) ? bs_ones.value : bs_zeroes.value; + state[-2 + 36].value = ((bool)(i2 & 0x8)) ? bs_ones.value : bs_zeroes.value; + state[-2 + 49].value = ((bool)(i2 & 0x10)) ? bs_ones.value : bs_zeroes.value; // guess lfsr output 1 + // 0xfe07fffbfdff + const bitslice_value_t filter2_1 = f_b_bs(state[-2 + 10].value, state[-2 + 14].value, state[-2 + 16].value, state[-2 + 17].value); + const bitslice_value_t filter2_2 = f_b_bs(state[-2 + 19].value, state[-2 + 23].value, state[-2 + 25].value, state[-2 + 28].value); + const bitslice_value_t filter2_4 = f_a_bs(state[-2 + 36].value, state[-2 + 45].value, state[-2 + 46].value, state[-2 + 48].value); + const bitslice_value_t filter2 = f_c_bs(filter2_0, filter2_1, filter2_2, filter2_3, filter2_4); + bitslice_t results2; + results2.value = results1.value & (filter2 ^ keystream[2].value); + if (results2.bytes64[0] == 0 + && results2.bytes64[1] == 0 + && results2.bytes64[2] == 0 + && results2.bytes64[3] == 0 + ) { + continue; + } + state[-2 + 50].value = lfsr_bs(2); + const bitslice_value_t filter3_3 = f_b_bs(state[-2 + 31].value, state[-2 + 32].value, state[-2 + 34].value, state[-2 + 36].value); + const bitslice_value_t filter4_0 = f_a_bs(state[-2 + 6].value, state[-2 + 7].value, state[-2 + 9].value, state[-2 + 10].value); + const bitslice_value_t filter4_1 = f_b_bs(state[-2 + 12].value, state[-2 + 16].value, state[-2 + 18].value, state[-2 + 19].value); + const bitslice_value_t filter4_2 = f_b_bs(state[-2 + 21].value, state[-2 + 25].value, state[-2 + 27].value, state[-2 + 30].value); + const bitslice_value_t filter7_0 = f_a_bs(state[-2 + 9].value, state[-2 + 10].value, state[-2 + 12].value, state[-2 + 13].value); + const bitslice_value_t filter7_1 = f_b_bs(state[-2 + 15].value, state[-2 + 19].value, state[-2 + 21].value, state[-2 + 22].value); + const bitslice_value_t filter8_2 = f_b_bs(state[-2 + 25].value, state[-2 + 29].value, state[-2 + 31].value, state[-2 + 34].value); + const bitslice_value_t filter10_1 = f_b_bs(state[-2 + 18].value, state[-2 + 22].value, state[-2 + 24].value, state[-2 + 25].value); + const bitslice_value_t filter10_2 = f_b_bs(state[-2 + 27].value, state[-2 + 31].value, state[-2 + 33].value, state[-2 + 36].value); + const bitslice_value_t filter11_1 = f_b_bs(state[-2 + 19].value, state[-2 + 23].value, state[-2 + 25].value, state[-2 + 26].value); + for (uint8_t i3 = 0; i3 < (1 << bits[3]); i3++) { + state[-2 + 11].value = ((bool)(i3 & 0x1)) ? bs_ones.value : bs_zeroes.value; + state[-2 + 20].value = ((bool)(i3 & 0x2)) ? bs_ones.value : bs_zeroes.value; + state[-2 + 37].value = ((bool)(i3 & 0x4)) ? bs_ones.value : bs_zeroes.value; + // 0xff07ffffffff + const bitslice_value_t filter3_1 = f_b_bs(state[-2 + 11].value, state[-2 + 15].value, state[-2 + 17].value, state[-2 + 18].value); + const bitslice_value_t filter3_2 = f_b_bs(state[-2 + 20].value, state[-2 + 24].value, state[-2 + 26].value, state[-2 + 29].value); + const bitslice_value_t filter3_4 = f_a_bs(state[-2 + 37].value, state[-2 + 46].value, state[-2 + 47].value, state[-2 + 49].value); + const bitslice_value_t filter3 = f_c_bs(filter3_0, filter3_1, filter3_2, filter3_3, filter3_4); + bitslice_t results3; + results3.value = results2.value & (filter3 ^ keystream[3].value); + if (results3.bytes64[0] == 0 + && results3.bytes64[1] == 0 + && results3.bytes64[2] == 0 + && results3.bytes64[3] == 0 + ) { + continue; + } + state[-2 + 51].value = lfsr_bs(3); + state[-2 + 52].value = lfsr_bs(4); + state[-2 + 53].value = lfsr_bs(5); + state[-2 + 54].value = lfsr_bs(6); + state[-2 + 55].value = lfsr_bs(7); + const bitslice_value_t filter4_3 = f_b_bs(state[-2 + 32].value, state[-2 + 33].value, state[-2 + 35].value, state[-2 + 37].value); + const bitslice_value_t filter5_0 = f_a_bs(state[-2 + 7].value, state[-2 + 8].value, state[-2 + 10].value, state[-2 + 11].value); + const bitslice_value_t filter5_1 = f_b_bs(state[-2 + 13].value, state[-2 + 17].value, state[-2 + 19].value, state[-2 + 20].value); + const bitslice_value_t filter6_0 = f_a_bs(state[-2 + 8].value, state[-2 + 9].value, state[-2 + 11].value, state[-2 + 12].value); + const bitslice_value_t filter6_1 = f_b_bs(state[-2 + 14].value, state[-2 + 18].value, state[-2 + 20].value, state[-2 + 21].value); + const bitslice_value_t filter8_0 = f_a_bs(state[-2 + 10].value, state[-2 + 11].value, state[-2 + 13].value, state[-2 + 14].value); + const bitslice_value_t filter8_1 = f_b_bs(state[-2 + 16].value, state[-2 + 20].value, state[-2 + 22].value, state[-2 + 23].value); + const bitslice_value_t filter9_0 = f_a_bs(state[-2 + 11].value, state[-2 + 12].value, state[-2 + 14].value, state[-2 + 15].value); + const bitslice_value_t filter9_4 = f_a_bs(state[-2 + 43].value, state[-2 + 52].value, state[-2 + 53].value, state[-2 + 55].value); + const bitslice_value_t filter11_2 = f_b_bs(state[-2 + 28].value, state[-2 + 32].value, state[-2 + 34].value, state[-2 + 37].value); + const bitslice_value_t filter12_1 = f_b_bs(state[-2 + 20].value, state[-2 + 24].value, state[-2 + 26].value, state[-2 + 27].value); + for (uint8_t i4 = 0; i4 < (1 << bits[4]); i4++) { + state[-2 + 38].value = ((bool)(i4 & 0x1)) ? bs_ones.value : bs_zeroes.value; + // 0xff87ffffffff + const bitslice_value_t filter4_4 = f_a_bs(state[-2 + 38].value, state[-2 + 47].value, state[-2 + 48].value, state[-2 + 50].value); + const bitslice_value_t filter4 = f_c_bs(filter4_0, filter4_1, filter4_2, filter4_3, filter4_4); + bitslice_t results4; + results4.value = results3.value & (filter4 ^ keystream[4].value); + if (results4.bytes64[0] == 0 + && results4.bytes64[1] == 0 + && results4.bytes64[2] == 0 + && results4.bytes64[3] == 0 + ) { + continue; + } + const bitslice_value_t filter5_3 = f_b_bs(state[-2 + 33].value, state[-2 + 34].value, state[-2 + 36].value, state[-2 + 38].value); + const bitslice_value_t filter12_2 = f_b_bs(state[-2 + 29].value, state[-2 + 33].value, state[-2 + 35].value, state[-2 + 38].value); + for (uint8_t i5 = 0; i5 < (1 << bits[5]); i5++) { + state[-2 + 39].value = ((bool)(i5 & 0x1)) ? bs_ones.value : bs_zeroes.value; + // 0xffc7ffffffff + const bitslice_value_t filter5_4 = f_a_bs(state[-2 + 39].value, state[-2 + 48].value, state[-2 + 49].value, state[-2 + 51].value); + const bitslice_value_t filter5 = f_c_bs(filter5_0, filter5_1, filter5_2, filter5_3, filter5_4); + bitslice_t results5; + results5.value = results4.value & (filter5 ^ keystream[5].value); + if (results5.bytes64[0] == 0 + && results5.bytes64[1] == 0 + && results5.bytes64[2] == 0 + && results5.bytes64[3] == 0 + ) { + continue; + } + const bitslice_value_t filter6_3 = f_b_bs(state[-2 + 34].value, state[-2 + 35].value, state[-2 + 37].value, state[-2 + 39].value); + for (uint8_t i6 = 0; i6 < (1 << bits[6]); i6++) { + state[-2 + 40].value = ((bool)(i6 & 0x1)) ? bs_ones.value : bs_zeroes.value; + // 0xffe7ffffffff + const bitslice_value_t filter6_4 = f_a_bs(state[-2 + 40].value, state[-2 + 49].value, state[-2 + 50].value, state[-2 + 52].value); + const bitslice_value_t filter6 = f_c_bs(filter6_0, filter6_1, filter6_2, filter6_3, filter6_4); + bitslice_t results6; + results6.value = results5.value & (filter6 ^ keystream[6].value); + if (results6.bytes64[0] == 0 + && results6.bytes64[1] == 0 + && results6.bytes64[2] == 0 + && results6.bytes64[3] == 0 + ) { + continue; + } + const bitslice_value_t filter7_3 = f_b_bs(state[-2 + 35].value, state[-2 + 36].value, state[-2 + 38].value, state[-2 + 40].value); + for (uint8_t i7 = 0; i7 < (1 << bits[7]); i7++) { + state[-2 + 41].value = ((bool)(i7 & 0x1)) ? bs_ones.value : bs_zeroes.value; + // 0xfff7ffffffff + const bitslice_value_t filter7_4 = f_a_bs(state[-2 + 41].value, state[-2 + 50].value, state[-2 + 51].value, state[-2 + 53].value); + const bitslice_value_t filter7 = f_c_bs(filter7_0, filter7_1, filter7_2, filter7_3, filter7_4); + bitslice_t results7; + results7.value = results6.value & (filter7 ^ keystream[7].value); + if (results7.bytes64[0] == 0 + && results7.bytes64[1] == 0 + && results7.bytes64[2] == 0 + && results7.bytes64[3] == 0 + ) { + continue; + } + const bitslice_value_t filter8_3 = f_b_bs(state[-2 + 36].value, state[-2 + 37].value, state[-2 + 39].value, state[-2 + 41].value); + const bitslice_value_t filter10_3 = f_b_bs(state[-2 + 38].value, state[-2 + 39].value, state[-2 + 41].value, state[-2 + 43].value); + const bitslice_value_t filter12_3 = f_b_bs(state[-2 + 40].value, state[-2 + 41].value, state[-2 + 43].value, state[-2 + 45].value); + for (uint8_t i8 = 0; i8 < (1 << bits[8]); i8++) { + state[-2 + 42].value = ((bool)(i8 & 0x1)) ? bs_ones.value : bs_zeroes.value; + // 0xffffffffffff + const bitslice_value_t filter8_4 = f_a_bs(state[-2 + 42].value, state[-2 + 51].value, state[-2 + 52].value, state[-2 + 54].value); + const bitslice_value_t filter9_3 = f_b_bs(state[-2 + 37].value, state[-2 + 38].value, state[-2 + 40].value, state[-2 + 42].value); + const bitslice_value_t filter11_3 = f_b_bs(state[-2 + 39].value, state[-2 + 40].value, state[-2 + 42].value, state[-2 + 44].value); + const bitslice_value_t filter8 = f_c_bs(filter8_0, filter8_1, filter8_2, filter8_3, filter8_4); + bitslice_t results8; + results8.value = results7.value & (filter8 ^ keystream[8].value); + if (results8.bytes64[0] == 0 + && results8.bytes64[1] == 0 + && results8.bytes64[2] == 0 + && results8.bytes64[3] == 0 + ) { + continue; + } + const bitslice_value_t filter9 = f_c_bs(filter9_0, filter9_1, filter9_2, filter9_3, filter9_4); + results8.value &= (filter9 ^ keystream[9].value); + if (results8.bytes64[0] == 0 + && results8.bytes64[1] == 0 + && results8.bytes64[2] == 0 + && results8.bytes64[3] == 0 + ) { + continue; + } + state[-2 + 56].value = lfsr_bs(8); + const bitslice_value_t filter10_4 = f_a_bs(state[-2 + 44].value, state[-2 + 53].value, state[-2 + 54].value, state[-2 + 56].value); + const bitslice_value_t filter10 = f_c_bs(filter10_0, filter10_1, filter10_2, filter10_3, filter10_4); + results8.value &= (filter10 ^ keystream[10].value); + if (results8.bytes64[0] == 0 + && results8.bytes64[1] == 0 + && results8.bytes64[2] == 0 + && results8.bytes64[3] == 0 + ) { + continue; + } + state[-2 + 57].value = lfsr_bs(9); + const bitslice_value_t filter11_4 = f_a_bs(state[-2 + 45].value, state[-2 + 54].value, state[-2 + 55].value, state[-2 + 57].value); + const bitslice_value_t filter11 = f_c_bs(filter11_0, filter11_1, filter11_2, filter11_3, filter11_4); + results8.value &= (filter11 ^ keystream[11].value); + if (results8.bytes64[0] == 0 + && results8.bytes64[1] == 0 + && results8.bytes64[2] == 0 + && results8.bytes64[3] == 0 + ) { + continue; + } + state[-2 + 58].value = lfsr_bs(10); + const bitslice_value_t filter12_4 = f_a_bs(state[-2 + 46].value, state[-2 + 55].value, state[-2 + 56].value, state[-2 + 58].value); + const bitslice_value_t filter12 = f_c_bs(filter12_0, filter12_1, filter12_2, filter12_3, filter12_4); + results8.value &= (filter12 ^ keystream[12].value); + if (results8.bytes64[0] == 0 + && results8.bytes64[1] == 0 + && results8.bytes64[2] == 0 + && results8.bytes64[3] == 0 + ) { + continue; + } + state[-2 + 59].value = lfsr_bs(11); + const bitslice_value_t filter13_0 = f_a_bs(state[-2 + 15].value, state[-2 + 16].value, state[-2 + 18].value, state[-2 + 19].value); + const bitslice_value_t filter13_1 = f_b_bs(state[-2 + 21].value, state[-2 + 25].value, state[-2 + 27].value, state[-2 + 28].value); + const bitslice_value_t filter13_2 = f_b_bs(state[-2 + 30].value, state[-2 + 34].value, state[-2 + 36].value, state[-2 + 39].value); + const bitslice_value_t filter13_3 = f_b_bs(state[-2 + 41].value, state[-2 + 42].value, state[-2 + 44].value, state[-2 + 46].value); + const bitslice_value_t filter13_4 = f_a_bs(state[-2 + 47].value, state[-2 + 56].value, state[-2 + 57].value, state[-2 + 59].value); + const bitslice_value_t filter13 = f_c_bs(filter13_0, filter13_1, filter13_2, filter13_3, filter13_4); + results8.value &= (filter13 ^ keystream[13].value); + if (results8.bytes64[0] == 0 + && results8.bytes64[1] == 0 + && results8.bytes64[2] == 0 + && results8.bytes64[3] == 0 + ) { + continue; + } + state[-2 + 60].value = lfsr_bs(12); + const bitslice_value_t filter14_0 = f_a_bs(state[-2 + 16].value, state[-2 + 17].value, state[-2 + 19].value, state[-2 + 20].value); + const bitslice_value_t filter14_1 = f_b_bs(state[-2 + 22].value, state[-2 + 26].value, state[-2 + 28].value, state[-2 + 29].value); + const bitslice_value_t filter14_2 = f_b_bs(state[-2 + 31].value, state[-2 + 35].value, state[-2 + 37].value, state[-2 + 40].value); + const bitslice_value_t filter14_3 = f_b_bs(state[-2 + 42].value, state[-2 + 43].value, state[-2 + 45].value, state[-2 + 47].value); + const bitslice_value_t filter14_4 = f_a_bs(state[-2 + 48].value, state[-2 + 57].value, state[-2 + 58].value, state[-2 + 60].value); + const bitslice_value_t filter14 = f_c_bs(filter14_0, filter14_1, filter14_2, filter14_3, filter14_4); + results8.value &= (filter14 ^ keystream[14].value); + if (results8.bytes64[0] == 0 + && results8.bytes64[1] == 0 + && results8.bytes64[2] == 0 + && results8.bytes64[3] == 0 + ) { + continue; + } + state[-2 + 61].value = lfsr_bs(13); + const bitslice_value_t filter15_0 = f_a_bs(state[-2 + 17].value, state[-2 + 18].value, state[-2 + 20].value, state[-2 + 21].value); + const bitslice_value_t filter15_1 = f_b_bs(state[-2 + 23].value, state[-2 + 27].value, state[-2 + 29].value, state[-2 + 30].value); + const bitslice_value_t filter15_2 = f_b_bs(state[-2 + 32].value, state[-2 + 36].value, state[-2 + 38].value, state[-2 + 41].value); + const bitslice_value_t filter15_3 = f_b_bs(state[-2 + 43].value, state[-2 + 44].value, state[-2 + 46].value, state[-2 + 48].value); + const bitslice_value_t filter15_4 = f_a_bs(state[-2 + 49].value, state[-2 + 58].value, state[-2 + 59].value, state[-2 + 61].value); + const bitslice_value_t filter15 = f_c_bs(filter15_0, filter15_1, filter15_2, filter15_3, filter15_4); + results8.value &= (filter15 ^ keystream[15].value); + if (results8.bytes64[0] == 0 + && results8.bytes64[1] == 0 + && results8.bytes64[2] == 0 + && results8.bytes64[3] == 0 + ) { + continue; + } + state[-2 + 62].value = lfsr_bs(14); + const bitslice_value_t filter16_0 = f_a_bs(state[-2 + 18].value, state[-2 + 19].value, state[-2 + 21].value, state[-2 + 22].value); + const bitslice_value_t filter16_1 = f_b_bs(state[-2 + 24].value, state[-2 + 28].value, state[-2 + 30].value, state[-2 + 31].value); + const bitslice_value_t filter16_2 = f_b_bs(state[-2 + 33].value, state[-2 + 37].value, state[-2 + 39].value, state[-2 + 42].value); + const bitslice_value_t filter16_3 = f_b_bs(state[-2 + 44].value, state[-2 + 45].value, state[-2 + 47].value, state[-2 + 49].value); + const bitslice_value_t filter16_4 = f_a_bs(state[-2 + 50].value, state[-2 + 59].value, state[-2 + 60].value, state[-2 + 62].value); + const bitslice_value_t filter16 = f_c_bs(filter16_0, filter16_1, filter16_2, filter16_3, filter16_4); + results8.value &= (filter16 ^ keystream[16].value); + if (results8.bytes64[0] == 0 + && results8.bytes64[1] == 0 + && results8.bytes64[2] == 0 + && results8.bytes64[3] == 0 + ) { + continue; + } + state[-2 + 63].value = lfsr_bs(15); + const bitslice_value_t filter17_0 = f_a_bs(state[-2 + 19].value, state[-2 + 20].value, state[-2 + 22].value, state[-2 + 23].value); + const bitslice_value_t filter17_1 = f_b_bs(state[-2 + 25].value, state[-2 + 29].value, state[-2 + 31].value, state[-2 + 32].value); + const bitslice_value_t filter17_2 = f_b_bs(state[-2 + 34].value, state[-2 + 38].value, state[-2 + 40].value, state[-2 + 43].value); + const bitslice_value_t filter17_3 = f_b_bs(state[-2 + 45].value, state[-2 + 46].value, state[-2 + 48].value, state[-2 + 50].value); + const bitslice_value_t filter17_4 = f_a_bs(state[-2 + 51].value, state[-2 + 60].value, state[-2 + 61].value, state[-2 + 63].value); + const bitslice_value_t filter17 = f_c_bs(filter17_0, filter17_1, filter17_2, filter17_3, filter17_4); + results8.value &= (filter17 ^ keystream[17].value); + if (results8.bytes64[0] == 0 + && results8.bytes64[1] == 0 + && results8.bytes64[2] == 0 + && results8.bytes64[3] == 0 + ) { + continue; + } + state[-2 + 64].value = lfsr_bs(16); + const bitslice_value_t filter18_0 = f_a_bs(state[-2 + 20].value, state[-2 + 21].value, state[-2 + 23].value, state[-2 + 24].value); + const bitslice_value_t filter18_1 = f_b_bs(state[-2 + 26].value, state[-2 + 30].value, state[-2 + 32].value, state[-2 + 33].value); + const bitslice_value_t filter18_2 = f_b_bs(state[-2 + 35].value, state[-2 + 39].value, state[-2 + 41].value, state[-2 + 44].value); + const bitslice_value_t filter18_3 = f_b_bs(state[-2 + 46].value, state[-2 + 47].value, state[-2 + 49].value, state[-2 + 51].value); + const bitslice_value_t filter18_4 = f_a_bs(state[-2 + 52].value, state[-2 + 61].value, state[-2 + 62].value, state[-2 + 64].value); + const bitslice_value_t filter18 = f_c_bs(filter18_0, filter18_1, filter18_2, filter18_3, filter18_4); + results8.value &= (filter18 ^ keystream[18].value); + if (results8.bytes64[0] == 0 + && results8.bytes64[1] == 0 + && results8.bytes64[2] == 0 + && results8.bytes64[3] == 0 + ) { + continue; + } + state[-2 + 65].value = lfsr_bs(17); + const bitslice_value_t filter19_0 = f_a_bs(state[-2 + 21].value, state[-2 + 22].value, state[-2 + 24].value, state[-2 + 25].value); + const bitslice_value_t filter19_1 = f_b_bs(state[-2 + 27].value, state[-2 + 31].value, state[-2 + 33].value, state[-2 + 34].value); + const bitslice_value_t filter19_2 = f_b_bs(state[-2 + 36].value, state[-2 + 40].value, state[-2 + 42].value, state[-2 + 45].value); + const bitslice_value_t filter19_3 = f_b_bs(state[-2 + 47].value, state[-2 + 48].value, state[-2 + 50].value, state[-2 + 52].value); + const bitslice_value_t filter19_4 = f_a_bs(state[-2 + 53].value, state[-2 + 62].value, state[-2 + 63].value, state[-2 + 65].value); + const bitslice_value_t filter19 = f_c_bs(filter19_0, filter19_1, filter19_2, filter19_3, filter19_4); + results8.value &= (filter19 ^ keystream[19].value); + if (results8.bytes64[0] == 0 + && results8.bytes64[1] == 0 + && results8.bytes64[2] == 0 + && results8.bytes64[3] == 0 + ) { + continue; + } + state[-2 + 66].value = lfsr_bs(18); + const bitslice_value_t filter20_0 = f_a_bs(state[-2 + 22].value, state[-2 + 23].value, state[-2 + 25].value, state[-2 + 26].value); + const bitslice_value_t filter20_1 = f_b_bs(state[-2 + 28].value, state[-2 + 32].value, state[-2 + 34].value, state[-2 + 35].value); + const bitslice_value_t filter20_2 = f_b_bs(state[-2 + 37].value, state[-2 + 41].value, state[-2 + 43].value, state[-2 + 46].value); + const bitslice_value_t filter20_3 = f_b_bs(state[-2 + 48].value, state[-2 + 49].value, state[-2 + 51].value, state[-2 + 53].value); + const bitslice_value_t filter20_4 = f_a_bs(state[-2 + 54].value, state[-2 + 63].value, state[-2 + 64].value, state[-2 + 66].value); + const bitslice_value_t filter20 = f_c_bs(filter20_0, filter20_1, filter20_2, filter20_3, filter20_4); + results8.value &= (filter20 ^ keystream[20].value); + if (results8.bytes64[0] == 0 + && results8.bytes64[1] == 0 + && results8.bytes64[2] == 0 + && results8.bytes64[3] == 0 + ) { + continue; + } + state[-2 + 67].value = lfsr_bs(19); + const bitslice_value_t filter21_0 = f_a_bs(state[-2 + 23].value, state[-2 + 24].value, state[-2 + 26].value, state[-2 + 27].value); + const bitslice_value_t filter21_1 = f_b_bs(state[-2 + 29].value, state[-2 + 33].value, state[-2 + 35].value, state[-2 + 36].value); + const bitslice_value_t filter21_2 = f_b_bs(state[-2 + 38].value, state[-2 + 42].value, state[-2 + 44].value, state[-2 + 47].value); + const bitslice_value_t filter21_3 = f_b_bs(state[-2 + 49].value, state[-2 + 50].value, state[-2 + 52].value, state[-2 + 54].value); + const bitslice_value_t filter21_4 = f_a_bs(state[-2 + 55].value, state[-2 + 64].value, state[-2 + 65].value, state[-2 + 67].value); + const bitslice_value_t filter21 = f_c_bs(filter21_0, filter21_1, filter21_2, filter21_3, filter21_4); + results8.value &= (filter21 ^ keystream[21].value); + if (results8.bytes64[0] == 0 + && results8.bytes64[1] == 0 + && results8.bytes64[2] == 0 + && results8.bytes64[3] == 0 + ) { + continue; + } + state[-2 + 68].value = lfsr_bs(20); + const bitslice_value_t filter22_0 = f_a_bs(state[-2 + 24].value, state[-2 + 25].value, state[-2 + 27].value, state[-2 + 28].value); + const bitslice_value_t filter22_1 = f_b_bs(state[-2 + 30].value, state[-2 + 34].value, state[-2 + 36].value, state[-2 + 37].value); + const bitslice_value_t filter22_2 = f_b_bs(state[-2 + 39].value, state[-2 + 43].value, state[-2 + 45].value, state[-2 + 48].value); + const bitslice_value_t filter22_3 = f_b_bs(state[-2 + 50].value, state[-2 + 51].value, state[-2 + 53].value, state[-2 + 55].value); + const bitslice_value_t filter22_4 = f_a_bs(state[-2 + 56].value, state[-2 + 65].value, state[-2 + 66].value, state[-2 + 68].value); + const bitslice_value_t filter22 = f_c_bs(filter22_0, filter22_1, filter22_2, filter22_3, filter22_4); + results8.value &= (filter22 ^ keystream[22].value); + if (results8.bytes64[0] == 0 + && results8.bytes64[1] == 0 + && results8.bytes64[2] == 0 + && results8.bytes64[3] == 0 + ) { + continue; + } + state[-2 + 69].value = lfsr_bs(21); + const bitslice_value_t filter23_0 = f_a_bs(state[-2 + 25].value, state[-2 + 26].value, state[-2 + 28].value, state[-2 + 29].value); + const bitslice_value_t filter23_1 = f_b_bs(state[-2 + 31].value, state[-2 + 35].value, state[-2 + 37].value, state[-2 + 38].value); + const bitslice_value_t filter23_2 = f_b_bs(state[-2 + 40].value, state[-2 + 44].value, state[-2 + 46].value, state[-2 + 49].value); + const bitslice_value_t filter23_3 = f_b_bs(state[-2 + 51].value, state[-2 + 52].value, state[-2 + 54].value, state[-2 + 56].value); + const bitslice_value_t filter23_4 = f_a_bs(state[-2 + 57].value, state[-2 + 66].value, state[-2 + 67].value, state[-2 + 69].value); + const bitslice_value_t filter23 = f_c_bs(filter23_0, filter23_1, filter23_2, filter23_3, filter23_4); + results8.value &= (filter23 ^ keystream[23].value); + if (results8.bytes64[0] == 0 + && results8.bytes64[1] == 0 + && results8.bytes64[2] == 0 + && results8.bytes64[3] == 0 + ) { + continue; + } + state[-2 + 70].value = lfsr_bs(22); + const bitslice_value_t filter24_0 = f_a_bs(state[-2 + 26].value, state[-2 + 27].value, state[-2 + 29].value, state[-2 + 30].value); + const bitslice_value_t filter24_1 = f_b_bs(state[-2 + 32].value, state[-2 + 36].value, state[-2 + 38].value, state[-2 + 39].value); + const bitslice_value_t filter24_2 = f_b_bs(state[-2 + 41].value, state[-2 + 45].value, state[-2 + 47].value, state[-2 + 50].value); + const bitslice_value_t filter24_3 = f_b_bs(state[-2 + 52].value, state[-2 + 53].value, state[-2 + 55].value, state[-2 + 57].value); + const bitslice_value_t filter24_4 = f_a_bs(state[-2 + 58].value, state[-2 + 67].value, state[-2 + 68].value, state[-2 + 70].value); + const bitslice_value_t filter24 = f_c_bs(filter24_0, filter24_1, filter24_2, filter24_3, filter24_4); + results8.value &= (filter24 ^ keystream[24].value); + if (results8.bytes64[0] == 0 + && results8.bytes64[1] == 0 + && results8.bytes64[2] == 0 + && results8.bytes64[3] == 0 + ) { + continue; + } + state[-2 + 71].value = lfsr_bs(23); + const bitslice_value_t filter25_0 = f_a_bs(state[-2 + 27].value, state[-2 + 28].value, state[-2 + 30].value, state[-2 + 31].value); + const bitslice_value_t filter25_1 = f_b_bs(state[-2 + 33].value, state[-2 + 37].value, state[-2 + 39].value, state[-2 + 40].value); + const bitslice_value_t filter25_2 = f_b_bs(state[-2 + 42].value, state[-2 + 46].value, state[-2 + 48].value, state[-2 + 51].value); + const bitslice_value_t filter25_3 = f_b_bs(state[-2 + 53].value, state[-2 + 54].value, state[-2 + 56].value, state[-2 + 58].value); + const bitslice_value_t filter25_4 = f_a_bs(state[-2 + 59].value, state[-2 + 68].value, state[-2 + 69].value, state[-2 + 71].value); + const bitslice_value_t filter25 = f_c_bs(filter25_0, filter25_1, filter25_2, filter25_3, filter25_4); + results8.value &= (filter25 ^ keystream[25].value); + if (results8.bytes64[0] == 0 + && results8.bytes64[1] == 0 + && results8.bytes64[2] == 0 + && results8.bytes64[3] == 0 + ) { + continue; + } + state[-2 + 72].value = lfsr_bs(24); + const bitslice_value_t filter26_0 = f_a_bs(state[-2 + 28].value, state[-2 + 29].value, state[-2 + 31].value, state[-2 + 32].value); + const bitslice_value_t filter26_1 = f_b_bs(state[-2 + 34].value, state[-2 + 38].value, state[-2 + 40].value, state[-2 + 41].value); + const bitslice_value_t filter26_2 = f_b_bs(state[-2 + 43].value, state[-2 + 47].value, state[-2 + 49].value, state[-2 + 52].value); + const bitslice_value_t filter26_3 = f_b_bs(state[-2 + 54].value, state[-2 + 55].value, state[-2 + 57].value, state[-2 + 59].value); + const bitslice_value_t filter26_4 = f_a_bs(state[-2 + 60].value, state[-2 + 69].value, state[-2 + 70].value, state[-2 + 72].value); + const bitslice_value_t filter26 = f_c_bs(filter26_0, filter26_1, filter26_2, filter26_3, filter26_4); + results8.value &= (filter26 ^ keystream[26].value); + if (results8.bytes64[0] == 0 + && results8.bytes64[1] == 0 + && results8.bytes64[2] == 0 + && results8.bytes64[3] == 0 + ) { + continue; + } + state[-2 + 73].value = lfsr_bs(25); + const bitslice_value_t filter27_0 = f_a_bs(state[-2 + 29].value, state[-2 + 30].value, state[-2 + 32].value, state[-2 + 33].value); + const bitslice_value_t filter27_1 = f_b_bs(state[-2 + 35].value, state[-2 + 39].value, state[-2 + 41].value, state[-2 + 42].value); + const bitslice_value_t filter27_2 = f_b_bs(state[-2 + 44].value, state[-2 + 48].value, state[-2 + 50].value, state[-2 + 53].value); + const bitslice_value_t filter27_3 = f_b_bs(state[-2 + 55].value, state[-2 + 56].value, state[-2 + 58].value, state[-2 + 60].value); + const bitslice_value_t filter27_4 = f_a_bs(state[-2 + 61].value, state[-2 + 70].value, state[-2 + 71].value, state[-2 + 73].value); + const bitslice_value_t filter27 = f_c_bs(filter27_0, filter27_1, filter27_2, filter27_3, filter27_4); + results8.value &= (filter27 ^ keystream[27].value); + if (results8.bytes64[0] == 0 + && results8.bytes64[1] == 0 + && results8.bytes64[2] == 0 + && results8.bytes64[3] == 0 + ) { + continue; + } + state[-2 + 74].value = lfsr_bs(26); + const bitslice_value_t filter28_0 = f_a_bs(state[-2 + 30].value, state[-2 + 31].value, state[-2 + 33].value, state[-2 + 34].value); + const bitslice_value_t filter28_1 = f_b_bs(state[-2 + 36].value, state[-2 + 40].value, state[-2 + 42].value, state[-2 + 43].value); + const bitslice_value_t filter28_2 = f_b_bs(state[-2 + 45].value, state[-2 + 49].value, state[-2 + 51].value, state[-2 + 54].value); + const bitslice_value_t filter28_3 = f_b_bs(state[-2 + 56].value, state[-2 + 57].value, state[-2 + 59].value, state[-2 + 61].value); + const bitslice_value_t filter28_4 = f_a_bs(state[-2 + 62].value, state[-2 + 71].value, state[-2 + 72].value, state[-2 + 74].value); + const bitslice_value_t filter28 = f_c_bs(filter28_0, filter28_1, filter28_2, filter28_3, filter28_4); + results8.value &= (filter28 ^ keystream[28].value); + if (results8.bytes64[0] == 0 + && results8.bytes64[1] == 0 + && results8.bytes64[2] == 0 + && results8.bytes64[3] == 0 + ) { + continue; + } + state[-2 + 75].value = lfsr_bs(27); + const bitslice_value_t filter29_0 = f_a_bs(state[-2 + 31].value, state[-2 + 32].value, state[-2 + 34].value, state[-2 + 35].value); + const bitslice_value_t filter29_1 = f_b_bs(state[-2 + 37].value, state[-2 + 41].value, state[-2 + 43].value, state[-2 + 44].value); + const bitslice_value_t filter29_2 = f_b_bs(state[-2 + 46].value, state[-2 + 50].value, state[-2 + 52].value, state[-2 + 55].value); + const bitslice_value_t filter29_3 = f_b_bs(state[-2 + 57].value, state[-2 + 58].value, state[-2 + 60].value, state[-2 + 62].value); + const bitslice_value_t filter29_4 = f_a_bs(state[-2 + 63].value, state[-2 + 72].value, state[-2 + 73].value, state[-2 + 75].value); + const bitslice_value_t filter29 = f_c_bs(filter29_0, filter29_1, filter29_2, filter29_3, filter29_4); + results8.value &= (filter29 ^ keystream[29].value); + if (results8.bytes64[0] == 0 + && results8.bytes64[1] == 0 + && results8.bytes64[2] == 0 + && results8.bytes64[3] == 0 + ) { + continue; + } + state[-2 + 76].value = lfsr_bs(28); + const bitslice_value_t filter30_0 = f_a_bs(state[-2 + 32].value, state[-2 + 33].value, state[-2 + 35].value, state[-2 + 36].value); + const bitslice_value_t filter30_1 = f_b_bs(state[-2 + 38].value, state[-2 + 42].value, state[-2 + 44].value, state[-2 + 45].value); + const bitslice_value_t filter30_2 = f_b_bs(state[-2 + 47].value, state[-2 + 51].value, state[-2 + 53].value, state[-2 + 56].value); + const bitslice_value_t filter30_3 = f_b_bs(state[-2 + 58].value, state[-2 + 59].value, state[-2 + 61].value, state[-2 + 63].value); + const bitslice_value_t filter30_4 = f_a_bs(state[-2 + 64].value, state[-2 + 73].value, state[-2 + 74].value, state[-2 + 76].value); + const bitslice_value_t filter30 = f_c_bs(filter30_0, filter30_1, filter30_2, filter30_3, filter30_4); + results8.value &= (filter30 ^ keystream[30].value); + if (results8.bytes64[0] == 0 + && results8.bytes64[1] == 0 + && results8.bytes64[2] == 0 + && results8.bytes64[3] == 0 + ) { + continue; + } + state[-2 + 77].value = lfsr_bs(29); + const bitslice_value_t filter31_0 = f_a_bs(state[-2 + 33].value, state[-2 + 34].value, state[-2 + 36].value, state[-2 + 37].value); + const bitslice_value_t filter31_1 = f_b_bs(state[-2 + 39].value, state[-2 + 43].value, state[-2 + 45].value, state[-2 + 46].value); + const bitslice_value_t filter31_2 = f_b_bs(state[-2 + 48].value, state[-2 + 52].value, state[-2 + 54].value, state[-2 + 57].value); + const bitslice_value_t filter31_3 = f_b_bs(state[-2 + 59].value, state[-2 + 60].value, state[-2 + 62].value, state[-2 + 64].value); + const bitslice_value_t filter31_4 = f_a_bs(state[-2 + 65].value, state[-2 + 74].value, state[-2 + 75].value, state[-2 + 77].value); + const bitslice_value_t filter31 = f_c_bs(filter31_0, filter31_1, filter31_2, filter31_3, filter31_4); + results8.value &= (filter31 ^ keystream[31].value); + if (results8.bytes64[0] == 0 + && results8.bytes64[1] == 0 + && results8.bytes64[2] == 0 + && results8.bytes64[3] == 0 + ) { + continue; + } + + for (size_t r = 0; r < MAX_BITSLICES; r++) { + if (!get_vector_bit(r, results8)) continue; + // take the state from layer 2 so we can recover the lowest 2 bits by inverting the LFSR + uint64_t state31 = unbitslice(&state[-2 + 2], r, 48); + state31 = lfsr_inv(state31); + state31 = lfsr_inv(state31); + try_state(state31 & ((1ull << 48) - 1)); + } + } // 8 + } // 7 + } // 6 + } // 5 + } // 4 + } // 3 + } // 2 + } // 1 + } // 0 + return NULL; +} + +static void try_state(uint64_t s) { + Hitag_State hstate; + uint64_t keyrev, key, nR1xk; + uint32_t b = 0; + + hstate.shiftreg = s; + + // recover key + keyrev = hstate.shiftreg & 0xffff; + nR1xk = (hstate.shiftreg >> 16) & 0xffffffff; + for (int i = 0; i < 32; i++) { + hstate.shiftreg = ((hstate.shiftreg) << 1) | ((uid >> (31 - i)) & 0x1); + b = (b << 1) | fnf(hstate.shiftreg); + } + keyrev |= (nR1xk ^ nR1 ^ b) << 16; + + // test key + hitag2_init(&hstate, keyrev, uid, nR2); + if ((aR2 ^ hitag2_nstep(&hstate, 32)) == 0xffffffff) { + + key = rev64(keyrev); + + printf("Key: "); + for (int i = 0; i < 6; i++) { + printf("%02X", (uint8_t)(key & 0xff)); + key = key >> 8; + } + printf("\n"); + exit(0); + } +} diff --git a/tools/hitag2crack/crack5/rfidler.h b/tools/hitag2crack/crack5/rfidler.h new file mode 100644 index 000000000..933547e6b --- /dev/null +++ b/tools/hitag2crack/crack5/rfidler.h @@ -0,0 +1,412 @@ +/*************************************************************************** + * A copy of the GNU GPL is appended to this file. * + * * + * This licence is based on the nmap licence, and we express our gratitude * + * for the work that went into producing it. There is no other connection * + * between RFIDler and nmap either expressed or implied. * + * * + ********************** IMPORTANT RFIDler LICENSE TERMS ******************** + * * + * * + * All references to RFIDler herein imply all it's derivatives, namely: * + * * + * o RFIDler-LF Standard * + * o RFIDler-LF Lite * + * o RFIDler-LF Nekkid * + * * + * * + * RFIDler is (C) 2013-2015 Aperture Labs Ltd. * + * * + * This program is free software; you may redistribute and/or modify it * + * under the terms of the GNU General Public License as published by the * + * Free Software Foundation; Version 2 ("GPL"), BUT ONLY WITH ALL OF THE * + * CLARIFICATIONS AND EXCEPTIONS DESCRIBED HEREIN. This guarantees your * + * right to use, modify, and redistribute this software under certain * + * conditions. If you wish to embed RFIDler technology into proprietary * + * software or hardware, we sell alternative licenses * + * (contact sales@aperturelabs.com). * + * * + * Note that the GPL places important restrictions on "derivative works", * + * yet it does not provide a detailed definition of that term. To avoid * + * misunderstandings, we interpret that term as broadly as copyright law * + * allows. For example, we consider an application to constitute a * + * derivative work for the purpose of this license if it does any of the * + * following with any software or content covered by this license * + * ("Covered Software"): * + * * + * o Integrates source code from Covered Software. * + * * + * o Is designed specifically to execute Covered Software and parse the * + * results (as opposed to typical shell or execution-menu apps, which will * + * execute anything you tell them to). * + * * + * o Includes Covered Software in a proprietary executable installer. The * + * installers produced by InstallShield are an example of this. Including * + * RFIDler with other software in compressed or archival form does not * + * trigger this provision, provided appropriate open source decompression * + * or de-archiving software is widely available for no charge. For the * + * purposes of this license, an installer is considered to include Covered * + * Software even if it actually retrieves a copy of Covered Software from * + * another source during runtime (such as by downloading it from the * + * Internet). * + * * + * o Links (statically or dynamically) to a library which does any of the * + * above. * + * * + * o Executes a helper program, module, or script to do any of the above. * + * * + * This list is not exclusive, but is meant to clarify our interpretation * + * of derived works with some common examples. Other people may interpret * + * the plain GPL differently, so we consider this a special exception to * + * the GPL that we apply to Covered Software. Works which meet any of * + * these conditions must conform to all of the terms of this license, * + * particularly including the GPL Section 3 requirements of providing * + * source code and allowing free redistribution of the work as a whole. * + * * + * As another special exception to the GPL terms, Aperture Labs Ltd. grants* + * permission to link the code of this program with any version of the * + * OpenSSL library which is distributed under a license identical to that * + * listed in the included docs/licenses/OpenSSL.txt file, and distribute * + * linked combinations including the two. * + * * + * Any redistribution of Covered Software, including any derived works, * + * must obey and carry forward all of the terms of this license, including * + * obeying all GPL rules and restrictions. For example, source code of * + * the whole work must be provided and free redistribution must be * + * allowed. All GPL references to "this License", are to be treated as * + * including the terms and conditions of this license text as well. * + * * + * Because this license imposes special exceptions to the GPL, Covered * + * Work may not be combined (even as part of a larger work) with plain GPL * + * software. The terms, conditions, and exceptions of this license must * + * be included as well. This license is incompatible with some other open * + * source licenses as well. In some cases we can relicense portions of * + * RFIDler or grant special permissions to use it in other open source * + * software. Please contact sales@aperturelabs.com with any such requests.* + * Similarly, we don't incorporate incompatible open source software into * + * Covered Software without special permission from the copyright holders. * + * * + * If you have any questions about the licensing restrictions on using * + * RFIDler in other works, are happy to help. As mentioned above, we also * + * offer alternative license to integrate RFIDler into proprietary * + * applications and appliances. These contracts have been sold to dozens * + * of software vendors, and generally include a perpetual license as well * + * as providing for priority support and updates. They also fund the * + * continued development of RFIDler. Please email sales@aperturelabs.com * + * for further information. * + * If you have received a written license agreement or contract for * + * Covered Software stating terms other than these, you may choose to use * + * and redistribute Covered Software under those terms instead of these. * + * * + * Source is provided to this software because we believe users have a * + * right to know exactly what a program is going to do before they run it. * + * This also allows you to audit the software for security holes (none * + * have been found so far). * + * * + * Source code also allows you to port RFIDler to new platforms, fix bugs, * + * and add new features. You are highly encouraged to send your changes * + * to the RFIDler mailing list for possible incorporation into the * + * main distribution. By sending these changes to Aperture Labs Ltd. or * + * one of the Aperture Labs Ltd. development mailing lists, or checking * + * them into the RFIDler source code repository, it is understood (unless * + * you specify otherwise) that you are offering the RFIDler Project * + * (Aperture Labs Ltd.) the unlimited, non-exclusive right to reuse, * + * modify, and relicense the code. RFIDler will always be available Open * + * Source, but this is important because the inability to relicense code * + * has caused devastating problems for other Free Software projects (such * + * as KDE and NASM). We also occasionally relicense the code to third * + * parties as discussed above. If you wish to specify special license * + * conditions of your contributions, just say so when you send them. * + * * + * This program 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 RFIDler * + * license file for more details (it's in a COPYING file included with * + * RFIDler, and also available from * + * https://github.com/ApertureLabsLtd/RFIDler/COPYING * + * * + ***************************************************************************/ + +// Author: Adam Laurie + +#include +#include + +// BCD hardware revision for usb descriptor (usb_descriptors.c) +#define RFIDLER_HW_VERSION 0x020 + +// max sizes in BITS +#define MAXBLOCKSIZE 512 +#define MAXTAGSIZE 4096 +#define MAXUID 512 + +#define TMP_LARGE_BUFF_LEN 2048 +#define TMP_SMALL_BUFF_LEN 256 +#define ANALOGUE_BUFF_LEN 8192 + +#define COMMS_BUFFER_SIZE 128 + +#define DETECT_BUFFER_SIZE 512 + +#define SAMPLEMASK ~(BIT_1 | BIT_0) // mask to remove two bottom bits from analogue sample - we will then use those for reader & bit period + +// globals + +extern BOOL WiegandOutput; // Output wiegand data whenenver UID is read +extern BYTE *EMU_Reset_Data; // Pointer to full array of bits as bytes, stored as 0x00/0x01, '*' terminated +extern BYTE *EMU_Data; // Pointer to current location in EMU_Reset_Data +extern BYTE EMU_ThisBit; // The next data bit to transmit +extern BYTE EMU_SubCarrier_T0; // Number of Frame Clocks for sub-carrier '0' +extern BYTE EMU_SubCarrier_T1; // Number of Frame Clocks for sub-carrier '1' +extern unsigned int EMU_Repeat; // Number of times to transmit full data set +extern BOOL EMU_Background; // Emulate in the background until told to stop +extern unsigned int EMU_DataBitRate; // Number of Frame Clocks per bit +extern BYTE TmpBits[TMP_LARGE_BUFF_LEN]; // Shared scratchpad +extern BYTE ReaderPeriod; // Flag for sample display +extern unsigned char Comms_In_Buffer[COMMS_BUFFER_SIZE]; // USB/Serial buffer +extern BYTE Interface; // user interface - CLI or API +extern BYTE CommsChannel; // user comms channel - USB or UART +extern BOOL FakeRead; // flag for analogue sampler to signal it wants access to buffers during read +extern BOOL PWD_Mode; // is this tag password protected? +extern BYTE Password[9]; // 32 bits as HEX string set with LOGIN +extern unsigned int Led_Count; // LED status counter, also used for entropy +extern unsigned long Reader_Bit_Count; // Reader ISR bit counter +extern char Previous; // Reader ISR previous bit type + +// RWD (read/write device) coil state +extern BYTE RWD_State; // current state of RWD coil +extern unsigned int RWD_Fc; // field clock in uS +extern unsigned int RWD_Gap_Period; // length of command gaps in OC5 ticks +extern unsigned int RWD_Zero_Period; // length of '0' in OC5 ticks +extern unsigned int RWD_One_Period; // length of '1' in OC5 ticks +extern unsigned int RWD_Sleep_Period; // length of initial sleep to reset tag in OC5 ticks +extern unsigned int RWD_Wake_Period; // length required for tag to restart in OC5 ticks +extern unsigned int RWD_Wait_Switch_TX_RX; // length to wait when switching from TX to RX in OC5 ticks +extern unsigned int RWD_Wait_Switch_RX_TX; // length to wait when switching from RX to TX in OC5 ticks +extern unsigned int RWD_Post_Wait; // low level ISR wait period in OC5 ticks +extern unsigned int RWD_OC5_config; // Output Compare Module settings +extern unsigned int RWD_OC5_r; // Output Compare Module primary compare value +extern unsigned int RWD_OC5_rs; // Output Compare Module secondary compare value +extern BYTE RWD_Command_Buff[TMP_SMALL_BUFF_LEN]; // Command buffer, array of bits as bytes, stored as 0x00/0x01, '*' terminated +extern BYTE *RWD_Command_ThisBit; // Current command bit +extern BOOL Reader_ISR_State; // current state of reader ISR + +// NVM variables +// timings etc. that want to survive a reboot should go here +typedef struct { + BYTE Name[7]; // will be set to "RFIDler" so we can test for new device + BYTE AutoRun[128]; // optional command to run at startup + unsigned char TagType; + unsigned int PSK_Quality; + unsigned int Timeout; + unsigned int Wiegand_Pulse; + unsigned int Wiegand_Gap; + BOOL Wiegand_IdleState; + unsigned int FrameClock; + unsigned char Modulation; + unsigned int DataRate; + unsigned int DataRateSub0; + unsigned int DataRateSub1; + unsigned int DataBits; + unsigned int DataBlocks; + unsigned int BlockSize; + unsigned char SyncBits; + BYTE Sync[4]; + BOOL BiPhase; + BOOL Invert; + BOOL Manchester; + BOOL HalfDuplex; + unsigned int Repeat; + unsigned int PotLow; + unsigned int PotHigh; + unsigned int RWD_Gap_Period; + unsigned int RWD_Zero_Period; + unsigned int RWD_One_Period; + unsigned int RWD_Sleep_Period; + unsigned int RWD_Wake_Period; + unsigned int RWD_Wait_Switch_TX_RX; + unsigned int RWD_Wait_Switch_RX_TX; +} StoredConfig; + +// somewhere to store TAG data. this will be interpreted according to the TAG +// type. +typedef struct { + BYTE TagType; // raw tag type + BYTE EmulatedTagType; // tag type this tag is configured to emulate + BYTE UID[MAXUID + 1]; // Null-terminated HEX string + BYTE Data[MAXTAGSIZE]; // raw data + unsigned char DataBlocks; // number of blocks in Data field + unsigned int BlockSize; // blocksize in bits +} VirtualTag; + +extern StoredConfig RFIDlerConfig; +extern VirtualTag RFIDlerVTag; +extern BYTE TmpBuff[NVM_PAGE_SIZE]; +extern BYTE DataBuff[ANALOGUE_BUFF_LEN]; +extern unsigned int DataBuffCount; +extern const BYTE *ModulationSchemes[]; +extern const BYTE *OnOff[]; +extern const BYTE *HighLow[]; +extern const BYTE *TagTypes[]; + +// globals for ISRs +extern BYTE EmulationMode; +extern unsigned long HW_Bits; +extern BYTE HW_Skip_Bits; +extern unsigned int PSK_Min_Pulse; +extern BOOL PSK_Read_Error; +extern BOOL Manchester_Error; +extern BOOL SnifferMode; +extern unsigned int Clock_Tick_Counter; +extern BOOL Clock_Tick_Counter_Reset; + +// smart card lib +#define MAX_ATR_LEN (BYTE)33 +extern BYTE scCardATR[MAX_ATR_LEN]; +extern BYTE scATRLength; + +// RTC +extern rtccTime RTC_time; // time structure +extern rtccDate RTC_date; // date structure + +// digital pots +#define POTLOW_DEFAULT 100 +#define POTHIGH_DEFAULT 150 +#define DC_OFFSET 60 // analogue circuit DC offset (as close as we can get without using 2 LSB) +#define VOLTS_TO_POT 0.019607843F + +// RWD/clock states +#define RWD_STATE_INACTIVE 0 // RWD not in use +#define RWD_STATE_GO_TO_SLEEP 1 // RWD coil shutdown request +#define RWD_STATE_SLEEPING 2 // RWD coil shutdown for sleep period +#define RWD_STATE_WAKING 3 // RWD active for pre-determined period after reset +#define RWD_STATE_START_SEND 4 // RWD starting send of data +#define RWD_STATE_SENDING_GAP 5 // RWD sending a gap +#define RWD_STATE_SENDING_BIT 6 // RWD sending a data bit +#define RWD_STATE_POST_WAIT 7 // RWD finished sending data, now in forced wait period +#define RWD_STATE_ACTIVE 8 // RWD finished, now just clocking a carrier + +// reader ISR states +#define READER_STOPPED 0 // reader not in use +#define READER_IDLING 1 // reader ISR running to preserve timing, but not reading +#define READER_RUNNING 2 // reader reading bits + + +// user interface types +#define INTERFACE_API 0 +#define INTERFACE_CLI 1 + +// comms channel +#define COMMS_NONE 0 +#define COMMS_USB 1 +#define COMMS_UART 2 + +#define MAX_HISTORY 2 // disable most of history for now - memory issue + +// tag write retries +#define TAG_WRITE_RETRY 5 + +// modulation modes - uppdate ModulationSchemes[] in tags.c if you change this +#define MOD_MODE_NONE 0 +#define MOD_MODE_ASK_OOK 1 +#define MOD_MODE_FSK1 2 +#define MOD_MODE_FSK2 3 +#define MOD_MODE_PSK1 4 +#define MOD_MODE_PSK2 5 +#define MOD_MODE_PSK3 6 + +// TAG types - update TagTypes[] in tags.c if you add to this list +#define TAG_TYPE_NONE 0 +#define TAG_TYPE_ASK_RAW 1 +#define TAG_TYPE_FSK1_RAW 2 +#define TAG_TYPE_FSK2_RAW 3 +#define TAG_TYPE_PSK1_RAW 4 +#define TAG_TYPE_PSK2_RAW 5 +#define TAG_TYPE_PSK3_RAW 6 +#define TAG_TYPE_HITAG1 7 +#define TAG_TYPE_HITAG2 8 +#define TAG_TYPE_EM4X02 9 +#define TAG_TYPE_Q5 10 +#define TAG_TYPE_HID_26 11 +#define TAG_TYPE_INDALA_64 12 +#define TAG_TYPE_INDALA_224 13 +#define TAG_TYPE_UNIQUE 14 +#define TAG_TYPE_FDXB 15 +#define TAG_TYPE_T55X7 16 // same as Q5 but different timings and no modulation-defeat +#define TAG_TYPE_AWID_26 17 +#define TAG_TYPE_EM4X05 18 +#define TAG_TYPE_TAMAGOTCHI 19 +#define TAG_TYPE_HDX 20 // same underlying data as FDX-B, but different modulation & telegram + +// various + +#define BINARY 0 +#define HEX 1 + +#define NO_ADDRESS -1 + +#define ACK TRUE +#define NO_ACK FALSE + +#define BLOCK TRUE +#define NO_BLOCK FALSE + +#define DATA TRUE +#define NO_DATA FALSE + +#define DEBUG_PIN_ON HIGH +#define DEBUG_PIN_OFF LOW + +#define FAST FALSE +#define SLOW TRUE + +#define NO_TRIGGER 0 + +#define LOCK TRUE +#define NO_LOCK FALSE + +#define NFC_MODE TRUE +#define NO_NFC_MODE FALSE + +#define ONESHOT_READ TRUE +#define NO_ONESHOT_READ FALSE + +#define RESET TRUE +#define NO_RESET FALSE + +#define SHUTDOWN_CLOCK TRUE +#define NO_SHUTDOWN_CLOCK FALSE + +#define SYNC TRUE +#define NO_SYNC FALSE + +#define VERIFY TRUE +#define NO_VERIFY FALSE + +#define VOLATILE FALSE +#define NON_VOLATILE TRUE + +#define NEWLINE TRUE +#define NO_NEWLINE FALSE + +#define WAIT TRUE +#define NO_WAIT FALSE + +#define WIPER_HIGH 0 +#define WIPER_LOW 1 + +// conversion for time to ticks +#define US_TO_TICKS 1000000L +#define US_OVER_10_TO_TICKS 10000000L +#define US_OVER_100_TO_TICKS 100000000L +// we can't get down to this level on pic, but we want to standardise on timings, so for now we fudge it +#define CONVERT_TO_TICKS(x) ((x / 10) * (GetSystemClock() / US_OVER_10_TO_TICKS)) +#define CONVERT_TICKS_TO_US(x) (x / (GetSystemClock() / US_TO_TICKS)) +#define TIMER5_PRESCALER 16 +#define MAX_TIMER5_TICKS (65535 * TIMER5_PRESCALER) + +// other conversions + +// bits to hex digits +#define HEXDIGITS(x) (x / 4) +#define HEXTOBITS(x) (x * 4) diff --git a/tools/hitag2crack/crack5/util.h b/tools/hitag2crack/crack5/util.h new file mode 100644 index 000000000..c2399c37c --- /dev/null +++ b/tools/hitag2crack/crack5/util.h @@ -0,0 +1,147 @@ +/*************************************************************************** + * A copy of the GNU GPL is appended to this file. * + * * + * This licence is based on the nmap licence, and we express our gratitude * + * for the work that went into producing it. There is no other connection * + * between RFIDler and nmap either expressed or implied. * + * * + ********************** IMPORTANT RFIDler LICENSE TERMS ******************** + * * + * * + * All references to RFIDler herein imply all it's derivatives, namely: * + * * + * o RFIDler-LF Standard * + * o RFIDler-LF Lite * + * o RFIDler-LF Nekkid * + * * + * * + * RFIDler is (C) 2013-2015 Aperture Labs Ltd. * + * * + * This program is free software; you may redistribute and/or modify it * + * under the terms of the GNU General Public License as published by the * + * Free Software Foundation; Version 2 ("GPL"), BUT ONLY WITH ALL OF THE * + * CLARIFICATIONS AND EXCEPTIONS DESCRIBED HEREIN. This guarantees your * + * right to use, modify, and redistribute this software under certain * + * conditions. If you wish to embed RFIDler technology into proprietary * + * software or hardware, we sell alternative licenses * + * (contact sales@aperturelabs.com). * + * * + * Note that the GPL places important restrictions on "derivative works", * + * yet it does not provide a detailed definition of that term. To avoid * + * misunderstandings, we interpret that term as broadly as copyright law * + * allows. For example, we consider an application to constitute a * + * derivative work for the purpose of this license if it does any of the * + * following with any software or content covered by this license * + * ("Covered Software"): * + * * + * o Integrates source code from Covered Software. * + * * + * o Is designed specifically to execute Covered Software and parse the * + * results (as opposed to typical shell or execution-menu apps, which will * + * execute anything you tell them to). * + * * + * o Includes Covered Software in a proprietary executable installer. The * + * installers produced by InstallShield are an example of this. Including * + * RFIDler with other software in compressed or archival form does not * + * trigger this provision, provided appropriate open source decompression * + * or de-archiving software is widely available for no charge. For the * + * purposes of this license, an installer is considered to include Covered * + * Software even if it actually retrieves a copy of Covered Software from * + * another source during runtime (such as by downloading it from the * + * Internet). * + * * + * o Links (statically or dynamically) to a library which does any of the * + * above. * + * * + * o Executes a helper program, module, or script to do any of the above. * + * * + * This list is not exclusive, but is meant to clarify our interpretation * + * of derived works with some common examples. Other people may interpret * + * the plain GPL differently, so we consider this a special exception to * + * the GPL that we apply to Covered Software. Works which meet any of * + * these conditions must conform to all of the terms of this license, * + * particularly including the GPL Section 3 requirements of providing * + * source code and allowing free redistribution of the work as a whole. * + * * + * As another special exception to the GPL terms, Aperture Labs Ltd. grants* + * permission to link the code of this program with any version of the * + * OpenSSL library which is distributed under a license identical to that * + * listed in the included docs/licenses/OpenSSL.txt file, and distribute * + * linked combinations including the two. * + * * + * Any redistribution of Covered Software, including any derived works, * + * must obey and carry forward all of the terms of this license, including * + * obeying all GPL rules and restrictions. For example, source code of * + * the whole work must be provided and free redistribution must be * + * allowed. All GPL references to "this License", are to be treated as * + * including the terms and conditions of this license text as well. * + * * + * Because this license imposes special exceptions to the GPL, Covered * + * Work may not be combined (even as part of a larger work) with plain GPL * + * software. The terms, conditions, and exceptions of this license must * + * be included as well. This license is incompatible with some other open * + * source licenses as well. In some cases we can relicense portions of * + * RFIDler or grant special permissions to use it in other open source * + * software. Please contact sales@aperturelabs.com with any such requests.* + * Similarly, we don't incorporate incompatible open source software into * + * Covered Software without special permission from the copyright holders. * + * * + * If you have any questions about the licensing restrictions on using * + * RFIDler in other works, are happy to help. As mentioned above, we also * + * offer alternative license to integrate RFIDler into proprietary * + * applications and appliances. These contracts have been sold to dozens * + * of software vendors, and generally include a perpetual license as well * + * as providing for priority support and updates. They also fund the * + * continued development of RFIDler. Please email sales@aperturelabs.com * + * for further information. * + * If you have received a written license agreement or contract for * + * Covered Software stating terms other than these, you may choose to use * + * and redistribute Covered Software under those terms instead of these. * + * * + * Source is provided to this software because we believe users have a * + * right to know exactly what a program is going to do before they run it. * + * This also allows you to audit the software for security holes (none * + * have been found so far). * + * * + * Source code also allows you to port RFIDler to new platforms, fix bugs, * + * and add new features. You are highly encouraged to send your changes * + * to the RFIDler mailing list for possible incorporation into the * + * main distribution. By sending these changes to Aperture Labs Ltd. or * + * one of the Aperture Labs Ltd. development mailing lists, or checking * + * them into the RFIDler source code repository, it is understood (unless * + * you specify otherwise) that you are offering the RFIDler Project * + * (Aperture Labs Ltd.) the unlimited, non-exclusive right to reuse, * + * modify, and relicense the code. RFIDler will always be available Open * + * Source, but this is important because the inability to relicense code * + * has caused devastating problems for other Free Software projects (such * + * as KDE and NASM). We also occasionally relicense the code to third * + * parties as discussed above. If you wish to specify special license * + * conditions of your contributions, just say so when you send them. * + * * + * This program 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 RFIDler * + * license file for more details (it's in a COPYING file included with * + * RFIDler, and also available from * + * https://github.com/ApertureLabsLtd/RFIDler/COPYING * + * * + ***************************************************************************/ + +// Author: Adam Laurie + +/* + * Hitag Crypto support macros + * These macros reverse the bit order in a byte, or *within* each byte of a + * 16 , 32 or 64 bit unsigned integer. (Not across the whole 16 etc bits.) + */ +#define rev8(X) ((((X) >> 7) &1) + (((X) >> 5) &2) + (((X) >> 3) &4) \ + + (((X) >> 1) &8) + (((X) << 1) &16) + (((X) << 3) &32) \ + + (((X) << 5) &64) + (((X) << 7) &128) ) +#define rev16(X) (rev8 (X) + (rev8 (X >> 8) << 8)) +#define rev32(X) (rev16(X) + (rev16(X >> 16) << 16)) +#define rev64(X) (rev32(X) + (rev32(X >> 32) << 32)) + + +unsigned long hexreversetoulong(BYTE *hex); +unsigned long long hexreversetoulonglong(BYTE *hex); + diff --git a/tools/hitag2crack/crack5/utilpart.c b/tools/hitag2crack/crack5/utilpart.c new file mode 100644 index 000000000..c46148491 --- /dev/null +++ b/tools/hitag2crack/crack5/utilpart.c @@ -0,0 +1,180 @@ +/*************************************************************************** + * A copy of the GNU GPL is appended to this file. * + * * + * This licence is based on the nmap licence, and we express our gratitude * + * for the work that went into producing it. There is no other connection * + * between RFIDler and nmap either expressed or implied. * + * * + ********************** IMPORTANT RFIDler LICENSE TERMS ******************** + * * + * * + * All references to RFIDler herein imply all it's derivatives, namely: * + * * + * o RFIDler-LF Standard * + * o RFIDler-LF Lite * + * o RFIDler-LF Nekkid * + * * + * * + * RFIDler is (C) 2013-2014 Aperture Labs Ltd. * + * * + * This program is free software; you may redistribute and/or modify it * + * under the terms of the GNU General Public License as published by the * + * Free Software Foundation; Version 2 ("GPL"), BUT ONLY WITH ALL OF THE * + * CLARIFICATIONS AND EXCEPTIONS DESCRIBED HEREIN. This guarantees your * + * right to use, modify, and redistribute this software under certain * + * conditions. If you wish to embed RFIDler technology into proprietary * + * software or hardware, we sell alternative licenses * + * (contact sales@aperturelabs.com). * + * * + * Note that the GPL places important restrictions on "derivative works", * + * yet it does not provide a detailed definition of that term. To avoid * + * misunderstandings, we interpret that term as broadly as copyright law * + * allows. For example, we consider an application to constitute a * + * derivative work for the purpose of this license if it does any of the * + * following with any software or content covered by this license * + * ("Covered Software"): * + * * + * o Integrates source code from Covered Software. * + * * + * o Is designed specifically to execute Covered Software and parse the * + * results (as opposed to typical shell or execution-menu apps, which will * + * execute anything you tell them to). * + * * + * o Includes Covered Software in a proprietary executable installer. The * + * installers produced by InstallShield are an example of this. Including * + * RFIDler with other software in compressed or archival form does not * + * trigger this provision, provided appropriate open source decompression * + * or de-archiving software is widely available for no charge. For the * + * purposes of this license, an installer is considered to include Covered * + * Software even if it actually retrieves a copy of Covered Software from * + * another source during runtime (such as by downloading it from the * + * Internet). * + * * + * o Links (statically or dynamically) to a library which does any of the * + * above. * + * * + * o Executes a helper program, module, or script to do any of the above. * + * * + * This list is not exclusive, but is meant to clarify our interpretation * + * of derived works with some common examples. Other people may interpret * + * the plain GPL differently, so we consider this a special exception to * + * the GPL that we apply to Covered Software. Works which meet any of * + * these conditions must conform to all of the terms of this license, * + * particularly including the GPL Section 3 requirements of providing * + * source code and allowing free redistribution of the work as a whole. * + * * + * As another special exception to the GPL terms, Aperture Labs Ltd. grants* + * permission to link the code of this program with any version of the * + * OpenSSL library which is distributed under a license identical to that * + * listed in the included docs/licenses/OpenSSL.txt file, and distribute * + * linked combinations including the two. * + * * + * Any redistribution of Covered Software, including any derived works, * + * must obey and carry forward all of the terms of this license, including * + * obeying all GPL rules and restrictions. For example, source code of * + * the whole work must be provided and free redistribution must be * + * allowed. All GPL references to "this License", are to be treated as * + * including the terms and conditions of this license text as well. * + * * + * Because this license imposes special exceptions to the GPL, Covered * + * Work may not be combined (even as part of a larger work) with plain GPL * + * software. The terms, conditions, and exceptions of this license must * + * be included as well. This license is incompatible with some other open * + * source licenses as well. In some cases we can relicense portions of * + * RFIDler or grant special permissions to use it in other open source * + * software. Please contact sales@aperturelabs.com with any such requests.* + * Similarly, we don't incorporate incompatible open source software into * + * Covered Software without special permission from the copyright holders. * + * * + * If you have any questions about the licensing restrictions on using * + * RFIDler in other works, are happy to help. As mentioned above, we also * + * offer alternative license to integrate RFIDler into proprietary * + * applications and appliances. These contracts have been sold to dozens * + * of software vendors, and generally include a perpetual license as well * + * as providing for priority support and updates. They also fund the * + * continued development of RFIDler. Please email sales@aperturelabs.com * + * for further information. * + * If you have received a written license agreement or contract for * + * Covered Software stating terms other than these, you may choose to use * + * and redistribute Covered Software under those terms instead of these. * + * * + * Source is provided to this software because we believe users have a * + * right to know exactly what a program is going to do before they run it. * + * This also allows you to audit the software for security holes (none * + * have been found so far). * + * * + * Source code also allows you to port RFIDler to new platforms, fix bugs, * + * and add new features. You are highly encouraged to send your changes * + * to the RFIDler mailing list for possible incorporation into the * + * main distribution. By sending these changes to Aperture Labs Ltd. or * + * one of the Aperture Labs Ltd. development mailing lists, or checking * + * them into the RFIDler source code repository, it is understood (unless * + * you specify otherwise) that you are offering the RFIDler Project * + * (Aperture Labs Ltd.) the unlimited, non-exclusive right to reuse, * + * modify, and relicense the code. RFIDler will always be available Open * + * Source, but this is important because the inability to relicense code * + * has caused devastating problems for other Free Software projects (such * + * as KDE and NASM). We also occasionally relicense the code to third * + * parties as discussed above. If you wish to specify special license * + * conditions of your contributions, just say so when you send them. * + * * + * This program 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 RFIDler * + * license file for more details (it's in a COPYING file included with * + * RFIDler, and also available from * + * https://github.com/ApertureLabsLtd/RFIDler/COPYING * + * * + ***************************************************************************/ + +// Author: Adam Laurie + + +#include +#include +#include "HardwareProfile.h" +#include "util.h" +#include "rfidler.h" +//#include "comms.h" + +// rtc +rtccTime RTC_time; // time structure +rtccDate RTC_date; // date structure + +// convert byte-reversed 8 digit hex to unsigned long +unsigned long hexreversetoulong(BYTE *hex) { + unsigned long ret = 0L; + unsigned int x; + BYTE i; + + if (strlen(hex) != 8) + return 0L; + + for (i = 0 ; i < 4 ; ++i) { + if (sscanf(hex, "%2X", &x) != 1) + return 0L; + ret += ((unsigned long) x) << i * 8; + hex += 2; + } + return ret; +} + +// convert byte-reversed 12 digit hex to unsigned long +unsigned long long hexreversetoulonglong(BYTE *hex) { + unsigned long long ret = 0LL; + BYTE tmp[9]; + + // this may seem an odd way to do it, but weird compiler issues were + // breaking direct conversion! + + tmp[8] = '\0'; + memset(tmp + 4, '0', 4); + memcpy(tmp, hex + 8, 4); + ret = hexreversetoulong(tmp); + ret <<= 32; + memcpy(tmp, hex, 8); + ret += hexreversetoulong(tmp); + return ret; +} + + diff --git a/tools/hitag2crack/crack5gpu/HardwareProfile.h b/tools/hitag2crack/crack5gpu/HardwareProfile.h new file mode 100644 index 000000000..bce139042 --- /dev/null +++ b/tools/hitag2crack/crack5gpu/HardwareProfile.h @@ -0,0 +1,524 @@ +/*************************************************************************** + * A copy of the GNU GPL is appended to this file. * + * * + * This licence is based on the nmap licence, and we express our gratitude * + * for the work that went into producing it. There is no other connection * + * between RFIDler and nmap either expressed or implied. * + * * + ********************** IMPORTANT RFIDler LICENSE TERMS ******************** + * * + * * + * All references to RFIDler herein imply all it's derivatives, namely: * + * * + * o RFIDler-LF Standard * + * o RFIDler-LF Lite * + * o RFIDler-LF Nekkid * + * * + * * + * RFIDler is (C) 2013-2014 Aperture Labs Ltd. * + * * + * This program is free software; you may redistribute and/or modify it * + * under the terms of the GNU General Public License as published by the * + * Free Software Foundation; Version 2 ("GPL"), BUT ONLY WITH ALL OF THE * + * CLARIFICATIONS AND EXCEPTIONS DESCRIBED HEREIN. This guarantees your * + * right to use, modify, and redistribute this software under certain * + * conditions. If you wish to embed RFIDler technology into proprietary * + * software or hardware, we sell alternative licenses * + * (contact sales@aperturelabs.com). * + * * + * Note that the GPL places important restrictions on "derivative works", * + * yet it does not provide a detailed definition of that term. To avoid * + * misunderstandings, we interpret that term as broadly as copyright law * + * allows. For example, we consider an application to constitute a * + * derivative work for the purpose of this license if it does any of the * + * following with any software or content covered by this license * + * ("Covered Software"): * + * * + * o Integrates source code from Covered Software. * + * * + * o Is designed specifically to execute Covered Software and parse the * + * results (as opposed to typical shell or execution-menu apps, which will * + * execute anything you tell them to). * + * * + * o Includes Covered Software in a proprietary executable installer. The * + * installers produced by InstallShield are an example of this. Including * + * RFIDler with other software in compressed or archival form does not * + * trigger this provision, provided appropriate open source decompression * + * or de-archiving software is widely available for no charge. For the * + * purposes of this license, an installer is considered to include Covered * + * Software even if it actually retrieves a copy of Covered Software from * + * another source during runtime (such as by downloading it from the * + * Internet). * + * * + * o Links (statically or dynamically) to a library which does any of the * + * above. * + * * + * o Executes a helper program, module, or script to do any of the above. * + * * + * This list is not exclusive, but is meant to clarify our interpretation * + * of derived works with some common examples. Other people may interpret * + * the plain GPL differently, so we consider this a special exception to * + * the GPL that we apply to Covered Software. Works which meet any of * + * these conditions must conform to all of the terms of this license, * + * particularly including the GPL Section 3 requirements of providing * + * source code and allowing free redistribution of the work as a whole. * + * * + * As another special exception to the GPL terms, Aperture Labs Ltd. grants* + * permission to link the code of this program with any version of the * + * OpenSSL library which is distributed under a license identical to that * + * listed in the included docs/licenses/OpenSSL.txt file, and distribute * + * linked combinations including the two. * + * * + * Any redistribution of Covered Software, including any derived works, * + * must obey and carry forward all of the terms of this license, including * + * obeying all GPL rules and restrictions. For example, source code of * + * the whole work must be provided and free redistribution must be * + * allowed. All GPL references to "this License", are to be treated as * + * including the terms and conditions of this license text as well. * + * * + * Because this license imposes special exceptions to the GPL, Covered * + * Work may not be combined (even as part of a larger work) with plain GPL * + * software. The terms, conditions, and exceptions of this license must * + * be included as well. This license is incompatible with some other open * + * source licenses as well. In some cases we can relicense portions of * + * RFIDler or grant special permissions to use it in other open source * + * software. Please contact sales@aperturelabs.com with any such requests.* + * Similarly, we don't incorporate incompatible open source software into * + * Covered Software without special permission from the copyright holders. * + * * + * If you have any questions about the licensing restrictions on using * + * RFIDler in other works, are happy to help. As mentioned above, we also * + * offer alternative license to integrate RFIDler into proprietary * + * applications and appliances. These contracts have been sold to dozens * + * of software vendors, and generally include a perpetual license as well * + * as providing for priority support and updates. They also fund the * + * continued development of RFIDler. Please email sales@aperturelabs.com * + * for further information. * + * If you have received a written license agreement or contract for * + * Covered Software stating terms other than these, you may choose to use * + * and redistribute Covered Software under those terms instead of these. * + * * + * Source is provided to this software because we believe users have a * + * right to know exactly what a program is going to do before they run it. * + * This also allows you to audit the software for security holes (none * + * have been found so far). * + * * + * Source code also allows you to port RFIDler to new platforms, fix bugs, * + * and add new features. You are highly encouraged to send your changes * + * to the RFIDler mailing list for possible incorporation into the * + * main distribution. By sending these changes to Aperture Labs Ltd. or * + * one of the Aperture Labs Ltd. development mailing lists, or checking * + * them into the RFIDler source code repository, it is understood (unless * + * you specify otherwise) that you are offering the RFIDler Project * + * (Aperture Labs Ltd.) the unlimited, non-exclusive right to reuse, * + * modify, and relicense the code. RFIDler will always be available Open * + * Source, but this is important because the inability to relicense code * + * has caused devastating problems for other Free Software projects (such * + * as KDE and NASM). We also occasionally relicense the code to third * + * parties as discussed above. If you wish to specify special license * + * conditions of your contributions, just say so when you send them. * + * * + * This program 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 RFIDler * + * license file for more details (it's in a COPYING file included with * + * RFIDler, and also available from * + * https://github.com/ApertureLabsLtd/RFIDler/COPYING * + * * + ***************************************************************************/ + +// Author: Adam Laurie + + + +#ifndef HARDWARE_PROFILE_UBW32_H +#define HARDWARE_PROFILE_UBW32_H + +//#include "plib.h" +typedef char BOOL; +typedef char BYTE; +typedef int rtccTime; +typedef int rtccDate; + + +#ifndef __PIC32MX__ +#define __PIC32MX__ +#endif + +#define GetSystemClock() (80000000ul) +#define GetPeripheralClock() (GetSystemClock()) +#define GetInstructionClock() (GetSystemClock()) + +//#define USE_SELF_POWER_SENSE_IO +#define tris_self_power TRISAbits.TRISA2 // Input +#define self_power 1 + +//#define USE_USB_BUS_SENSE_IO +#define tris_usb_bus_sense TRISBbits.TRISB5 // Input +#define USB_BUS_SENSE 1 + +// LEDs +#define mLED_1 LATEbits.LATE3 + +#define mLED_2 LATEbits.LATE2 +#define mLED_Comms mLED_2 + +#define mLED_3 LATEbits.LATE1 +#define mLED_Clock mLED_3 + +#define mLED_4 LATEbits.LATE0 +#define mLED_Emulate mLED_4 + +#define mLED_5 LATGbits.LATG6 +#define mLED_Read mLED_5 + +#define mLED_6 LATAbits.LATA15 +#define mLED_User mLED_6 + +#define mLED_7 LATDbits.LATD11 +#define mLED_Error mLED_7 + +// active low +#define mLED_ON 0 +#define mLED_OFF 1 + +#define mGetLED_1() mLED_1 +#define mGetLED_USB() mLED_1 +#define mGetLED_2() mLED_2 +#define mGetLED_Comms() mLED_2 +#define mGetLED_3() mLED_3 +#define mGetLED_Clock() mLED_3 +#define mGetLED_4() mLED_4 +#define mGetLED_Emulate() mLED_4 +#define mGetLED_5() mLED_5 +#define mGetLED_Read() mLED_5 +#define mGetLED_6() mLED_6 +#define mGetLED_User() mLED_6 +#define mGetLED_7() mLED_7 +#define mGetLED_Error() mLED_7 + +#define mLED_1_On() mLED_1 = mLED_ON +#define mLED_USB_On() mLED_1_On() +#define mLED_2_On() mLED_2 = mLED_ON +#define mLED_Comms_On() mLED_2_On() +#define mLED_3_On() mLED_3 = mLED_ON +#define mLED_Clock_On() mLED_3_On() +#define mLED_4_On() mLED_4 = mLED_ON +#define mLED_Emulate_On() mLED_4_On() +#define mLED_5_On() mLED_5 = mLED_ON +#define mLED_Read_On() mLED_5_On() +#define mLED_6_On() mLED_6 = mLED_ON +#define mLED_User_On() mLED_6_On() +#define mLED_7_On() mLED_7 = mLED_ON +#define mLED_Error_On() mLED_7_On() + +#define mLED_1_Off() mLED_1 = mLED_OFF +#define mLED_USB_Off() mLED_1_Off() +#define mLED_2_Off() mLED_2 = mLED_OFF +#define mLED_Comms_Off() mLED_2_Off() +#define mLED_3_Off() mLED_3 = mLED_OFF +#define mLED_Clock_Off() mLED_3_Off() +#define mLED_4_Off() mLED_4 = mLED_OFF +#define mLED_Emulate_Off() mLED_4_Off() +#define mLED_5_Off() mLED_5 = mLED_OFF +#define mLED_Read_Off() mLED_5_Off() +#define mLED_6_Off() mLED_6 = mLED_OFF +#define mLED_User_Off() mLED_6_Off() +#define mLED_7_Off() mLED_7 = mLED_OFF +#define mLED_Error_Off() mLED_7_Off() + +#define mLED_1_Toggle() mLED_1 = !mLED_1 +#define mLED_USB_Toggle() mLED_1_Toggle() +#define mLED_2_Toggle() mLED_2 = !mLED_2 +#define mLED_Comms_Toggle() mLED_2_Toggle() +#define mLED_3_Toggle() mLED_3 = !mLED_3 +#define mLED_Clock_Toggle() mLED_3_Toggle() +#define mLED_4_Toggle() mLED_4 = !mLED_4 +#define mLED_Emulate_Toggle() mLED_4_Toggle() +#define mLED_5_Toggle() mLED_5 = !mLED_5 +#define mLED_Read_Toggle( ) mLED_5_Toggle() +#define mLED_6_Toggle() mLED_6 = !mLED_6 +#define mLED_User_Toggle() mLED_6_Toggle() +#define mLED_7_Toggle() mLED_7 = !mLED_7 +#define mLED_Error_Toggle() mLED_7_Toggle() + +#define mLED_All_On() { mLED_1_On(); mLED_2_On(); mLED_3_On(); mLED_4_On(); mLED_5_On(); mLED_6_On(); mLED_7_On(); } +#define mLED_All_Off() { mLED_1_Off(); mLED_2_Off(); mLED_3_Off(); mLED_4_Off(); mLED_5_Off(); mLED_6_Off(); mLED_7_Off(); } + +// usb status lights +#define mLED_Both_Off() {mLED_USB_Off();mLED_Comms_Off();} +#define mLED_Both_On() {mLED_USB_On();mLED_Comms_On();} +#define mLED_Only_USB_On() {mLED_USB_On();mLED_Comms_Off();} +#define mLED_Only_Comms_On() {mLED_USB_Off();mLED_Comms_On();} + +/** SWITCH *********************************************************/ +#define swBootloader PORTEbits.RE7 +#define swUser PORTEbits.RE6 + +/** I/O pin definitions ********************************************/ +#define INPUT_PIN 1 +#define OUTPUT_PIN 0 + +#define TRUE 1 +#define FALSE 0 + +#define ENABLE 1 +#define DISABE 0 + +#define EVEN 0 +#define ODD 1 + +#define LOW FALSE +#define HIGH TRUE + +#define CLOCK_ON LOW +#define CLOCK_OFF HIGH + +// output coil control - select between reader/emulator circuits +#define COIL_MODE LATBbits.LATB4 +#define COIL_MODE_READER() COIL_MODE= LOW +#define COIL_MODE_EMULATOR() COIL_MODE= HIGH + +// coil for emulation +#define COIL_OUT LATGbits.LATG9 +#define COIL_OUT_HIGH() COIL_OUT=HIGH +#define COIL_OUT_LOW() COIL_OUT=LOW + +// door relay (active low) +#define DOOR_RELAY LATAbits.LATA14 +#define DOOR_RELAY_OPEN() DOOR_RELAY= HIGH +#define DOOR_RELAY_CLOSE() DOOR_RELAY= LOW + +// inductance/capacitance freq +#define IC_FREQUENCY PORTAbits.RA2 + +#define SNIFFER_COIL PORTDbits.RD12 // external reader clock detect +#define READER_ANALOGUE PORTBbits.RB11 // reader coil analogue +#define DIV_LOW_ANALOGUE PORTBbits.RB12 // voltage divider LOW analogue +#define DIV_HIGH_ANALOGUE PORTBbits.RB13 // voltage divider HIGH analogue + +// clock coil (normally controlled by OC Module, but defined here so we can force it high or low) +#define CLOCK_COIL PORTDbits.RD4 +#define CLOCK_COIL_MOVED PORTDbits.RD0 // temporary for greenwire + +// digital output after analogue reader circuit +#define READER_DATA PORTDbits.RD8 + +// trace / debug +#define DEBUG_PIN_1 LATCbits.LATC1 +#define DEBUG_PIN_1_TOGGLE() DEBUG_PIN_1= !DEBUG_PIN_1 +#define DEBUG_PIN_2 LATCbits.LATC2 +#define DEBUG_PIN_2_TOGGLE() DEBUG_PIN_2= !DEBUG_PIN_2 +#define DEBUG_PIN_3 LATCbits.LATC3 +#define DEBUG_PIN_3_TOGGLE() DEBUG_PIN_3= !DEBUG_PIN_3 +#define DEBUG_PIN_4 LATEbits.LATE5 +#define DEBUG_PIN_4_TOGGLE() DEBUG_PIN_4= !DEBUG_PIN_4 + +// spi (sdi1) for sd card (not directly referenced) +//#define SD_CARD_RX LATCbits.LATC4 +//#define SD_CARD_TX LATDbits.LATD0 +//#define SD_CARD_CLK LATDbits.LATD10 +//#define SD_CARD_SS LATDbits.LATD9 +// spi for SD card +#define SD_CARD_DET LATFbits.LATF0 +#define SD_CARD_WE LATFbits.LATF1 // write enable - unused for microsd but allocated anyway as library checks it +// (held LOW by default - cut solder bridge to GND to free pin if required) +#define SPI_SD SPI_CHANNEL1 +#define SPI_SD_BUFF SPI1BUF +#define SPI_SD_STAT SPI1STATbits +// see section below for more defines! + +// iso 7816 smartcard +// microchip SC module defines pins so we don't need to, but +// they are listed here to help avoid conflicts +#define ISO_7816_RX LATBbits.LATF2 // RX +#define ISO_7816_TX LATBbits.LATF8 // TX +#define ISO_7816_VCC LATBbits.LATB9 // Power +#define ISO_7816_CLK LATCbits.LATD1 // Clock +#define ISO_7816_RST LATEbits.LATE8 // Reset + +// user LED +#define USER_LED LATDbits.LATD7 +#define USER_LED_ON() LATDbits.LATD7=1 +#define USER_LED_OFF() LATDbits.LATD7=0 + +// LCR +#define LCR_CALIBRATE LATBbits.LATB5 + +// wiegand / clock & data +#define WIEGAND_IN_0 PORTDbits.RD5 +#define WIEGAND_IN_0_PULLUP CNPUEbits.CNPUE14 +#define WIEGAND_IN_0_PULLDOWN CNPDbits.CNPD14 +#define WIEGAND_IN_1 PORTDbits.RD6 +#define WIEGAND_IN_1_PULLUP CNPUEbits.CNPUE15 +#define WIEGAND_IN_1_PULLDOWN CNPDbits.CNPD15 +#define CAND_IN_DATA WIEGAND_IN_0 +#define CAND_IN_CLOCK WIEGAND_IN_1 + +#define WIEGAND_OUT_0 LATDbits.LATD3 +#define WIEGAND_OUT_1 LATDbits.LATD2 +#define WIEGAND_OUT_0_TRIS TRISDbits.TRISD3 +#define WIEGAND_OUT_1_TRIS TRISDbits.TRISD2 +#define CAND_OUT_DATA WIEGAND_OUT_0 +#define CAND_OUT_CLOCK WIEGAND_OUT_1 + +// connect/disconnect reader clock from coil - used to send RWD signals by creating gaps in carrier +#define READER_CLOCK_ENABLE LATEbits.LATE9 +#define READER_CLOCK_ENABLE_ON() READER_CLOCK_ENABLE=CLOCK_ON +#define READER_CLOCK_ENABLE_OFF(x) {READER_CLOCK_ENABLE=CLOCK_OFF; COIL_OUT=x;} + +// these input pins must NEVER bet set to output or they will cause short circuits! +// they can be used to see data from reader before it goes into or gate +#define OR_IN_A PORTAbits.RA4 +#define OR_IN_B PORTAbits.RA5 + + +// CNCON and CNEN are set to allow wiegand input pin weak pullups to be switched on +#define Init_GPIO() { \ + CNCONbits.ON= TRUE; \ + CNENbits.CNEN14= TRUE; \ + CNENbits.CNEN15= TRUE; \ + TRISAbits.TRISA2= INPUT_PIN; \ + TRISAbits.TRISA4= INPUT_PIN; \ + TRISAbits.TRISA5= INPUT_PIN; \ + TRISAbits.TRISA14= OUTPUT_PIN; \ + TRISAbits.TRISA15= OUTPUT_PIN; \ + TRISBbits.TRISB4= OUTPUT_PIN; \ + TRISBbits.TRISB5= OUTPUT_PIN; \ + TRISBbits.TRISB9= OUTPUT_PIN; \ + TRISBbits.TRISB11= INPUT_PIN; \ + TRISBbits.TRISB12= INPUT_PIN; \ + TRISBbits.TRISB13= INPUT_PIN; \ + TRISCbits.TRISC1= OUTPUT_PIN; \ + TRISCbits.TRISC2= OUTPUT_PIN; \ + TRISCbits.TRISC3= OUTPUT_PIN; \ + TRISCbits.TRISC4= INPUT_PIN; \ + TRISDbits.TRISD0= INPUT_PIN; \ + TRISDbits.TRISD1= OUTPUT_PIN; \ + TRISDbits.TRISD2= OUTPUT_PIN; \ + TRISDbits.TRISD3= OUTPUT_PIN; \ + TRISDbits.TRISD4= OUTPUT_PIN; \ + TRISDbits.TRISD5= INPUT_PIN; \ + TRISDbits.TRISD6= INPUT_PIN; \ + TRISDbits.TRISD7= OUTPUT_PIN; \ + TRISDbits.TRISD8= INPUT_PIN; \ + TRISDbits.TRISD11= OUTPUT_PIN; \ + TRISDbits.TRISD12= INPUT_PIN; \ + TRISEbits.TRISE0= OUTPUT_PIN; \ + TRISEbits.TRISE1= OUTPUT_PIN; \ + TRISEbits.TRISE2= OUTPUT_PIN; \ + TRISEbits.TRISE3= OUTPUT_PIN; \ + TRISEbits.TRISE5= OUTPUT_PIN; \ + TRISEbits.TRISE6= INPUT_PIN; \ + TRISEbits.TRISE7= INPUT_PIN; \ + TRISEbits.TRISE8= OUTPUT_PIN; \ + TRISEbits.TRISE9= OUTPUT_PIN; \ + TRISFbits.TRISF0= INPUT_PIN; \ + TRISFbits.TRISF1= INPUT_PIN; \ + TRISFbits.TRISF2= INPUT_PIN; \ + TRISFbits.TRISF8= OUTPUT_PIN; \ + TRISGbits.TRISG6= OUTPUT_PIN; \ + TRISGbits.TRISG12= INPUT_PIN; \ + TRISGbits.TRISG13= INPUT_PIN; \ + TRISGbits.TRISG9= OUTPUT_PIN; \ + LATBbits.LATB9= LOW; \ + LATCbits.LATC1= LOW; \ + LATCbits.LATC2= LOW; \ + LATCbits.LATC3= LOW; \ + LATDbits.LATD2= WIEGAND_IN_1; \ + LATDbits.LATD3= WIEGAND_IN_0; \ + LATEbits.LATE5= LOW; \ + LATEbits.LATE9= HIGH; \ + } + +// uart3 (CLI/API) speed +#define BAUDRATE3 115200UL +#define BRG_DIV3 4 +#define BRGH3 1 + +// spi for potentiometer +#define SPI_POT SPI_CHANNEL4 +#define SPI_POT_BUFF SPI4BUF +#define SPI_POT_STAT SPI4STATbits + +// spi for sd card - defines required for Microchip SD-SPI libs +// define interface type +#define USE_SD_INTERFACE_WITH_SPI + +#define MDD_USE_SPI_1 +#define SPI_START_CFG_1 (PRI_PRESCAL_64_1 | SEC_PRESCAL_8_1 | MASTER_ENABLE_ON | SPI_CKE_ON | SPI_SMP_ON) +#define SPI_START_CFG_2 (SPI_ENABLE) +// Define the SPI frequency +#define SPI_FREQUENCY (20000000) +// Description: SD-SPI Card Detect Input bit +#define SD_CD PORTFbits.RF0 +// Description: SD-SPI Card Detect TRIS bit +#define SD_CD_TRIS TRISFbits.TRISF0 +// Description: SD-SPI Write Protect Check Input bit +#define SD_WE PORTFbits.RF1 +// Description: SD-SPI Write Protect Check TRIS bit +#define SD_WE_TRIS TRISFbits.TRISF1 +// Description: The main SPI control register +#define SPICON1 SPI1CON +// Description: The SPI status register +#define SPISTAT SPI1STAT +// Description: The SPI Buffer +#define SPIBUF SPI1BUF +// Description: The receive buffer full bit in the SPI status register +#define SPISTAT_RBF SPI1STATbits.SPIRBF +// Description: The bitwise define for the SPI control register (i.e. _____bits) +#define SPICON1bits SPI1CONbits +// Description: The bitwise define for the SPI status register (i.e. _____bits) +#define SPISTATbits SPI1STATbits +// Description: The enable bit for the SPI module +#define SPIENABLE SPICON1bits.ON +// Description: The definition for the SPI baud rate generator register (PIC32) +#define SPIBRG SPI1BRG +// Description: The TRIS bit for the SCK pin +#define SPICLOCK TRISDbits.TRISD10 +// Description: The TRIS bit for the SDI pin +#define SPIIN TRISCbits.TRISC4 +// Description: The TRIS bit for the SDO pin +#define SPIOUT TRISDbits.TRISD0 +#define SD_CS LATDbits.LATD9 +// Description: SD-SPI Chip Select TRIS bit +#define SD_CS_TRIS TRISDbits.TRISD9 +//SPI library functions +#define putcSPI putcSPI1 +#define getcSPI getcSPI1 +#define OpenSPI(config1, config2) OpenSPI1(config1, config2) + +// Define setup parameters for OpenADC10 function +// Turn module on | Ouput in integer format | Trigger mode auto | Enable autosample +#define ADC_CONFIG1 (ADC_FORMAT_INTG | ADC_CLK_AUTO | ADC_AUTO_SAMPLING_ON) +// ADC ref external | Disable offset test | Disable scan mode | Perform 2 samples | Use dual buffers | Use alternate mode +#define ADC_CONFIG2 (ADC_VREF_AVDD_AVSS | ADC_OFFSET_CAL_DISABLE | ADC_SCAN_OFF | ADC_SAMPLES_PER_INT_1 | ADC_ALT_BUF_ON | ADC_ALT_INPUT_ON) + +// Use ADC internal clock | Set sample time +#define ADC_CONFIG3 (ADC_CONV_CLK_INTERNAL_RC | ADC_SAMPLE_TIME_0) + +// slow sample rate for tuning coils +#define ADC_CONFIG2_SLOW (ADC_VREF_AVDD_AVSS | ADC_OFFSET_CAL_DISABLE | ADC_SCAN_OFF | ADC_SAMPLES_PER_INT_16 | ADC_ALT_BUF_ON | ADC_ALT_INPUT_ON) +#define ADC_CONFIG3_SLOW (ADC_CONV_CLK_INTERNAL_RC | ADC_SAMPLE_TIME_31) + +// use AN11 +#define ADC_CONFIGPORT ENABLE_AN11_ANA +// Do not assign channels to scan +#define ADC_CONFIGSCAN SKIP_SCAN_ALL + +#define ADC_TO_VOLTS 0.003208F + + +// flash memory - int myvar = *(int*)(myflashmemoryaddress); + +// memory is 0x9D005000 to 0x9D07FFFF + +#define NVM_MEMORY_END 0x9D07FFFF +#define NVM_PAGE_SIZE 4096 +#define NVM_PAGES 2 // config & VTAG +#define RFIDLER_NVM_ADDRESS (NVM_MEMORY_END - (NVM_PAGE_SIZE * NVM_PAGES)) + +// UART timeout in us +#define SERIAL_TIMEOUT 100 + +#endif diff --git a/tools/hitag2crack/crack5gpu/Makefile b/tools/hitag2crack/crack5gpu/Makefile new file mode 100644 index 000000000..adcf61113 --- /dev/null +++ b/tools/hitag2crack/crack5gpu/Makefile @@ -0,0 +1,24 @@ +CFLAGS?=-Wall +#INCLUDE=-I/usr/local/cuda-7.5/include +INCLUDE=-I/opt/nvidia/cuda/include +#Linux +#LIBS=-L/usr/local/cuda-7.5/lib64 -lOpenCL +LIBS=-L/opt/nvidia/cuda/lib64 -lOpenCL +#Mac +#LIBS=-framework OpenCL + +all: ht2crack5.c utilpart.o ht2crack2utils.o hitagcrypto.o + $(CC) $(CFLAGS) ht2crack5.c -o ht2crack5gpu utilpart.o ht2crack2utils.o hitagcrypto.o $(LIBS) -lpthread + +utilpart.o: util.h utilpart.c + $(CC) $(CFLAGS) $(INCLUDE) -c utilpart.c + +hitagcrypto.o: hitagcrypto.h hitagcrypto.c + $(CC) $(CFLAGS) $(INCLUDE) -c hitagcrypto.c + +ht2crack2utils.o: ht2crack2utils.h ht2crack2utils.c + $(CC) $(CFLAGS) $(INCLUDE) -c ht2crack2utils.c + +clean: + rm -f *.o ht2crack5gpu +fresh: clean all diff --git a/tools/hitag2crack/crack5gpu/README.md b/tools/hitag2crack/crack5gpu/README.md new file mode 100644 index 000000000..292da2fc0 --- /dev/null +++ b/tools/hitag2crack/crack5gpu/README.md @@ -0,0 +1,27 @@ +ht2crack5gpu + + + +Build +----- + +It requires an OpenCL framework. + +If required, edit Makefile and adjust INCLUDE and LIBS directives to your setup. + +``` +make clean +make +``` + +Run +--- + +You'll need just two nR aR pairs. These are the +encrypted nonces and challenge response values. They should be in hex. + +``` +./ht2crack5gpu +``` + +UID is the UID of the tag that you used to gather the nR aR values. diff --git a/tools/hitag2crack/crack5gpu/hitagcrypto.c b/tools/hitag2crack/crack5gpu/hitagcrypto.c new file mode 100644 index 000000000..47449c3e3 --- /dev/null +++ b/tools/hitag2crack/crack5gpu/hitagcrypto.c @@ -0,0 +1,373 @@ +/*************************************************************************** + * A copy of the GNU GPL is appended to this file. * + * * + * This licence is based on the nmap licence, and we express our gratitude * + * for the work that went into producing it. There is no other connection * + * between RFIDler and nmap either expressed or implied. * + * * + ********************** IMPORTANT RFIDler LICENSE TERMS ******************** + * * + * * + * All references to RFIDler herein imply all it's derivatives, namely: * + * * + * o RFIDler-LF Standard * + * o RFIDler-LF Lite * + * o RFIDler-LF Nekkid * + * * + * * + * RFIDler is (C) 2013-2015 Aperture Labs Ltd. * + * * + * This program is free software; you may redistribute and/or modify it * + * under the terms of the GNU General Public License as published by the * + * Free Software Foundation; Version 2 ("GPL"), BUT ONLY WITH ALL OF THE * + * CLARIFICATIONS AND EXCEPTIONS DESCRIBED HEREIN. This guarantees your * + * right to use, modify, and redistribute this software under certain * + * conditions. If you wish to embed RFIDler technology into proprietary * + * software or hardware, we sell alternative licenses * + * (contact sales@aperturelabs.com). * + * * + * Note that the GPL places important restrictions on "derivative works", * + * yet it does not provide a detailed definition of that term. To avoid * + * misunderstandings, we interpret that term as broadly as copyright law * + * allows. For example, we consider an application to constitute a * + * derivative work for the purpose of this license if it does any of the * + * following with any software or content covered by this license * + * ("Covered Software"): * + * * + * o Integrates source code from Covered Software. * + * * + * o Is designed specifically to execute Covered Software and parse the * + * results (as opposed to typical shell or execution-menu apps, which will * + * execute anything you tell them to). * + * * + * o Includes Covered Software in a proprietary executable installer. The * + * installers produced by InstallShield are an example of this. Including * + * RFIDler with other software in compressed or archival form does not * + * trigger this provision, provided appropriate open source decompression * + * or de-archiving software is widely available for no charge. For the * + * purposes of this license, an installer is considered to include Covered * + * Software even if it actually retrieves a copy of Covered Software from * + * another source during runtime (such as by downloading it from the * + * Internet). * + * * + * o Links (statically or dynamically) to a library which does any of the * + * above. * + * * + * o Executes a helper program, module, or script to do any of the above. * + * * + * This list is not exclusive, but is meant to clarify our interpretation * + * of derived works with some common examples. Other people may interpret * + * the plain GPL differently, so we consider this a special exception to * + * the GPL that we apply to Covered Software. Works which meet any of * + * these conditions must conform to all of the terms of this license, * + * particularly including the GPL Section 3 requirements of providing * + * source code and allowing free redistribution of the work as a whole. * + * * + * As another special exception to the GPL terms, Aperture Labs Ltd. grants* + * permission to link the code of this program with any version of the * + * OpenSSL library which is distributed under a license identical to that * + * listed in the included docs/licenses/OpenSSL.txt file, and distribute * + * linked combinations including the two. * + * * + * Any redistribution of Covered Software, including any derived works, * + * must obey and carry forward all of the terms of this license, including * + * obeying all GPL rules and restrictions. For example, source code of * + * the whole work must be provided and free redistribution must be * + * allowed. All GPL references to "this License", are to be treated as * + * including the terms and conditions of this license text as well. * + * * + * Because this license imposes special exceptions to the GPL, Covered * + * Work may not be combined (even as part of a larger work) with plain GPL * + * software. The terms, conditions, and exceptions of this license must * + * be included as well. This license is incompatible with some other open * + * source licenses as well. In some cases we can relicense portions of * + * RFIDler or grant special permissions to use it in other open source * + * software. Please contact sales@aperturelabs.com with any such requests.* + * Similarly, we don't incorporate incompatible open source software into * + * Covered Software without special permission from the copyright holders. * + * * + * If you have any questions about the licensing restrictions on using * + * RFIDler in other works, are happy to help. As mentioned above, we also * + * offer alternative license to integrate RFIDler into proprietary * + * applications and appliances. These contracts have been sold to dozens * + * of software vendors, and generally include a perpetual license as well * + * as providing for priority support and updates. They also fund the * + * continued development of RFIDler. Please email sales@aperturelabs.com * + * for further information. * + * If you have received a written license agreement or contract for * + * Covered Software stating terms other than these, you may choose to use * + * and redistribute Covered Software under those terms instead of these. * + * * + * Source is provided to this software because we believe users have a * + * right to know exactly what a program is going to do before they run it. * + * This also allows you to audit the software for security holes (none * + * have been found so far). * + * * + * Source code also allows you to port RFIDler to new platforms, fix bugs, * + * and add new features. You are highly encouraged to send your changes * + * to the RFIDler mailing list for possible incorporation into the * + * main distribution. By sending these changes to Aperture Labs Ltd. or * + * one of the Aperture Labs Ltd. development mailing lists, or checking * + * them into the RFIDler source code repository, it is understood (unless * + * you specify otherwise) that you are offering the RFIDler Project * + * (Aperture Labs Ltd.) the unlimited, non-exclusive right to reuse, * + * modify, and relicense the code. RFIDler will always be available Open * + * Source, but this is important because the inability to relicense code * + * has caused devastating problems for other Free Software projects (such * + * as KDE and NASM). We also occasionally relicense the code to third * + * parties as discussed above. If you wish to specify special license * + * conditions of your contributions, just say so when you send them. * + * * + * This program 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 RFIDler * + * license file for more details (it's in a COPYING file included with * + * RFIDler, and also available from * + * https://github.com/ApertureLabsLtd/RFIDler/COPYING * + * * + ***************************************************************************/ + +// Author: unknown. +// Modifications for RFIDler: Tony Naggs , Adam Laurie + +// uncomment this to build file as a standalone crypto test program +// #define UNIT_TEST +// also uncomment to include verbose debug prints +// #define TEST_DEBUG + +//#include +#include "HardwareProfile.h" +#include "rfidler.h" +#include "hitagcrypto.h" +#include "util.h" + +#ifdef UNIT_TEST +#include +#endif + +#if defined(UNIT_TEST) && defined(TEST_DEBUG) +// Note that printf format %I64x prints 64 bit ints in MS Visual C/C++. +// This may need changing for other compilers/platforms. +#define DEBUG_PRINTF(...) printf(__VA_ARGS__) +#else +#define DEBUG_PRINTF(...) +#endif + + +/* Brief info about NXP Hitag 1, Hitag 2, Hitag S and Hitag u (mu) + + Hitag 125kHz RFID was created by a company called Mikron (Mikron Gesellschaft + fur Integrierte Mikroelektronik Mbh), of Austria, for micropayment applications. + At about the same time, late 1980s to early 1990s, Mikron developed the + similarly featured Mifare micropayment card for 13.56MHz RFID. + (Mikron's European Patent EP 0473569 A2 was filed 23 August 1991, with a + priority date of 23 Aug 1990.) + Mikron was subsequently acquired by Philips Semiconductors in 1995. + Philips Semiconductors divsion subsequently became NXP. + + + Modulation read/write device -> transponder: 100 % ASK and binary pulse + length coding + + Modulation transponder -> read/write device: Strong ASK modulation, + selectable Manchester or Biphase coding + + Hitag S, Hitag u; anti-collision procedure + + Fast anti-collision protocol + + Hitag u; optional Cyclic Redundancy Check (CRC) + + Reader Talks First mode + + Hitag 2 & later; Transponder Talks First (TTF) mode + + Temporary switch from Transponder Talks First into Reader Talks First + (RTF) Mode + + Data rate read/write device to transponder: 5.2 kbit/s + + Data rates transponder to read/write device: 2 kbit/s, 4 kbit/s, 8 kbit/s + + 32-bit password feature + + Hitag 2, S = 32-bit Unique Identifier + + Hitag u = 48-bit Unique Identifier + + Selectable password modes for reader / tag mutual authentication + (Hitag 1 has 2 pairs of keys, later versions have 1 pair) + + Hitag 2 & Hitag S; Selectable encrypted mode, 48 bit key + + Known tag types: + + HITAG 1 2048 bits total memory + + HITAG 2 256 Bit total memory Read/Write + 8 pages of 32 bits, inc UID (32), + secret key (64), password (24), config (8) + + HITAG S 32 32 bits Unique Identifier Read Only + HITAG S 256 256 bits total memory Read/Write + HITAG S 2048 2048 bits total memory Read/Write + + HITAG u RO64 64 bits total memory Read Only + HITAG u 128 bits total memory Read/Write + HITAG u Advanced 512 bits total memory Read/Write + HITAG u Advanced+ 1760 bits total memory Read/Write + + Default 48-bit key for Hitag 2, S encryption: + "MIKRON" = O N M I K R + Key = 4F 4E 4D 49 4B 52 + +*/ + + +// We want the crypto functions to be as fast as possible, so optimize! +// The best compiler optimization in Microchip's free XC32 edition is -O1 +#pragma GCC optimize("O1") + +// private, nonlinear function to generate 1 crypto bit +static uint32_t hitag2_crypt(uint64_t x); + + +// macros to pick out 4 bits in various patterns of 1s & 2s & make a new number +#define pickbits2_2(S, A, B) ( ((S >> A) & 3) | ((S >> (B - 2)) & 0xC) ) +#define pickbits1x4(S, A, B, C, D) ( ((S >> A) & 1) | ((S >> (B - 1)) & 2) | \ + ((S >> (C - 2)) & 4) | ((S >> (D - 3)) & 8) ) +#define pickbits1_1_2(S, A, B, C) ( ((S >> A) & 1) | ((S >> (B - 1)) & 2) | \ + ((S >> (C - 2)) & 0xC) ) +#define pickbits2_1_1(S, A, B, C) ( ((S >> A) & 3) | ((S >> (B - 2)) & 4) | \ + ((S >> (C - 3)) & 8) ) +#define pickbits1_2_1(S, A, B, C) ( ((S >> A) & 1) | ((S >> (B - 1)) & 6) | \ + ((S >> (C - 3)) & 8) ) + + +static uint32_t hitag2_crypt(uint64_t x) { + const uint32_t ht2_function4a = 0x2C79; // 0010 1100 0111 1001 + const uint32_t ht2_function4b = 0x6671; // 0110 0110 0111 0001 + const uint32_t ht2_function5c = 0x7907287B; // 0111 1001 0000 0111 0010 1000 0111 1011 + uint32_t bitindex; + + bitindex = (ht2_function4a >> pickbits2_2(x, 1, 4)) & 1; + bitindex |= ((ht2_function4b << 1) >> pickbits1_1_2(x, 7, 11, 13)) & 0x02; + bitindex |= ((ht2_function4b << 2) >> pickbits1x4(x, 16, 20, 22, 25)) & 0x04; + bitindex |= ((ht2_function4b << 3) >> pickbits2_1_1(x, 27, 30, 32)) & 0x08; + bitindex |= ((ht2_function4a << 4) >> pickbits1_2_1(x, 33, 42, 45)) & 0x10; + + DEBUG_PRINTF("hitag2_crypt bitindex = %02x\n", bitindex); + return (ht2_function5c >> bitindex) & 1; +} + +/* + * Parameters: + * Hitag_State* pstate - output, internal state after initialisation + * uint64_t sharedkey - 48 bit key shared between reader & tag + * uint32_t serialnum - 32 bit tag serial number + * uint32_t initvector - 32 bit random IV from reader, part of tag authentication + */ +void hitag2_init(Hitag_State *pstate, uint64_t sharedkey, uint32_t serialnum, uint32_t initvector) { + // init state, from serial number and lowest 16 bits of shared key + uint64_t state = ((sharedkey & 0xFFFF) << 32) | serialnum; + + // mix the initialisation vector and highest 32 bits of the shared key + initvector ^= (uint32_t)(sharedkey >> 16); + + // move 16 bits from (IV xor Shared Key) to top of uint64_t state + // these will be XORed in turn with output of the crypto function + state |= (uint64_t) initvector << 48; + initvector >>= 16; + + // unrolled loop is faster on PIC32 (MIPS), do 32 times + // shift register, then calc new bit + state >>= 1; + state = (state >> 1) ^ (uint64_t) hitag2_crypt(state) << 46; + state = (state >> 1) ^ (uint64_t) hitag2_crypt(state) << 46; + state = (state >> 1) ^ (uint64_t) hitag2_crypt(state) << 46; + state = (state >> 1) ^ (uint64_t) hitag2_crypt(state) << 46; + state = (state >> 1) ^ (uint64_t) hitag2_crypt(state) << 46; + state = (state >> 1) ^ (uint64_t) hitag2_crypt(state) << 46; + state = (state >> 1) ^ (uint64_t) hitag2_crypt(state) << 46; + state = (state >> 1) ^ (uint64_t) hitag2_crypt(state) << 46; + + state = (state >> 1) ^ (uint64_t) hitag2_crypt(state) << 46; + state = (state >> 1) ^ (uint64_t) hitag2_crypt(state) << 46; + state = (state >> 1) ^ (uint64_t) hitag2_crypt(state) << 46; + state = (state >> 1) ^ (uint64_t) hitag2_crypt(state) << 46; + state = (state >> 1) ^ (uint64_t) hitag2_crypt(state) << 46; + state = (state >> 1) ^ (uint64_t) hitag2_crypt(state) << 46; + state = (state >> 1) ^ (uint64_t) hitag2_crypt(state) << 46; + state = (state >> 1) ^ (uint64_t) hitag2_crypt(state) << 46; + + // highest 16 bits of IV XOR Shared Key + state |= (uint64_t) initvector << 47; + + state = (state >> 1) ^ (uint64_t) hitag2_crypt(state) << 46; + state = (state >> 1) ^ (uint64_t) hitag2_crypt(state) << 46; + state = (state >> 1) ^ (uint64_t) hitag2_crypt(state) << 46; + state = (state >> 1) ^ (uint64_t) hitag2_crypt(state) << 46; + state = (state >> 1) ^ (uint64_t) hitag2_crypt(state) << 46; + state = (state >> 1) ^ (uint64_t) hitag2_crypt(state) << 46; + state = (state >> 1) ^ (uint64_t) hitag2_crypt(state) << 46; + state = (state >> 1) ^ (uint64_t) hitag2_crypt(state) << 46; + + state = (state >> 1) ^ (uint64_t) hitag2_crypt(state) << 46; + state = (state >> 1) ^ (uint64_t) hitag2_crypt(state) << 46; + state = (state >> 1) ^ (uint64_t) hitag2_crypt(state) << 46; + state = (state >> 1) ^ (uint64_t) hitag2_crypt(state) << 46; + state = (state >> 1) ^ (uint64_t) hitag2_crypt(state) << 46; + state = (state >> 1) ^ (uint64_t) hitag2_crypt(state) << 46; + state = (state >> 1) ^ (uint64_t) hitag2_crypt(state) << 46; + state ^= (uint64_t) hitag2_crypt(state) << 47; + + DEBUG_PRINTF("hitag2_init result = %012I64x\n", state); + pstate->shiftreg = state; + /* naive version for reference, LFSR has 16 taps + pstate->lfsr = state ^ (state >> 2) ^ (state >> 3) ^ (state >> 6) + ^ (state >> 7) ^ (state >> 8) ^ (state >> 16) ^ (state >> 22) + ^ (state >> 23) ^ (state >> 26) ^ (state >> 30) ^ (state >> 41) + ^ (state >> 42) ^ (state >> 43) ^ (state >> 46) ^ (state >> 47); + */ + { + // optimise with one 64-bit intermediate + uint64_t temp = state ^ (state >> 1); + pstate->lfsr = state ^ (state >> 6) ^ (state >> 16) + ^ (state >> 26) ^ (state >> 30) ^ (state >> 41) + ^ (temp >> 2) ^ (temp >> 7) ^ (temp >> 22) + ^ (temp >> 42) ^ (temp >> 46); + } +} + + +/* + * Return up to 32 crypto bits. + * Last bit is in least significant bit, earlier bits are shifted left. + * Note that the Hitag transmission protocol is least significant bit, + * so we may want to change this, or add a function, that returns the + * crypto output bits in the other order. + * + * Parameters: + * Hitag_State* pstate - in/out, internal cipher state after initialisation + * uint32_t steps - number of bits requested, (capped at 32) + */ +uint32_t hitag2_nstep(Hitag_State *pstate, uint32_t steps) { + uint64_t state = pstate->shiftreg; + uint32_t result = 0; + uint64_t lfsr = pstate->lfsr; + + if (steps == 0) + return 0; + +// if (steps > 32) +// steps = 32; + + do { + // update shift registers + if (lfsr & 1) { + state = (state >> 1) | 0x800000000000; + lfsr = (lfsr >> 1) ^ 0xB38083220073; + + // accumulate next bit of crypto + result = (result << 1) | hitag2_crypt(state); + } else { + state >>= 1; + lfsr >>= 1; + + result = (result << 1) | hitag2_crypt(state); + } + } while (--steps); + + DEBUG_PRINTF("hitag2_nstep state = %012I64x, result %02x\n", state, result); + pstate->shiftreg = state; + pstate->lfsr = lfsr; + return result; +} + +// end of crypto core, revert to default optimization level +#pragma GCC reset_options diff --git a/tools/hitag2crack/crack5gpu/hitagcrypto.h b/tools/hitag2crack/crack5gpu/hitagcrypto.h new file mode 100644 index 000000000..274d3d82c --- /dev/null +++ b/tools/hitag2crack/crack5gpu/hitagcrypto.h @@ -0,0 +1,167 @@ +/*************************************************************************** + * A copy of the GNU GPL is appended to this file. * + * * + * This licence is based on the nmap licence, and we express our gratitude * + * for the work that went into producing it. There is no other connection * + * between RFIDler and nmap either expressed or implied. * + * * + ********************** IMPORTANT RFIDler LICENSE TERMS ******************** + * * + * * + * All references to RFIDler herein imply all it's derivatives, namely: * + * * + * o RFIDler-LF Standard * + * o RFIDler-LF Lite * + * o RFIDler-LF Nekkid * + * * + * * + * RFIDler is (C) 2013-2014 Aperture Labs Ltd. * + * * + * This program is free software; you may redistribute and/or modify it * + * under the terms of the GNU General Public License as published by the * + * Free Software Foundation; Version 2 ("GPL"), BUT ONLY WITH ALL OF THE * + * CLARIFICATIONS AND EXCEPTIONS DESCRIBED HEREIN. This guarantees your * + * right to use, modify, and redistribute this software under certain * + * conditions. If you wish to embed RFIDler technology into proprietary * + * software or hardware, we sell alternative licenses * + * (contact sales@aperturelabs.com). * + * * + * Note that the GPL places important restrictions on "derivative works", * + * yet it does not provide a detailed definition of that term. To avoid * + * misunderstandings, we interpret that term as broadly as copyright law * + * allows. For example, we consider an application to constitute a * + * derivative work for the purpose of this license if it does any of the * + * following with any software or content covered by this license * + * ("Covered Software"): * + * * + * o Integrates source code from Covered Software. * + * * + * o Is designed specifically to execute Covered Software and parse the * + * results (as opposed to typical shell or execution-menu apps, which will * + * execute anything you tell them to). * + * * + * o Includes Covered Software in a proprietary executable installer. The * + * installers produced by InstallShield are an example of this. Including * + * RFIDler with other software in compressed or archival form does not * + * trigger this provision, provided appropriate open source decompression * + * or de-archiving software is widely available for no charge. For the * + * purposes of this license, an installer is considered to include Covered * + * Software even if it actually retrieves a copy of Covered Software from * + * another source during runtime (such as by downloading it from the * + * Internet). * + * * + * o Links (statically or dynamically) to a library which does any of the * + * above. * + * * + * o Executes a helper program, module, or script to do any of the above. * + * * + * This list is not exclusive, but is meant to clarify our interpretation * + * of derived works with some common examples. Other people may interpret * + * the plain GPL differently, so we consider this a special exception to * + * the GPL that we apply to Covered Software. Works which meet any of * + * these conditions must conform to all of the terms of this license, * + * particularly including the GPL Section 3 requirements of providing * + * source code and allowing free redistribution of the work as a whole. * + * * + * As another special exception to the GPL terms, Aperture Labs Ltd. grants* + * permission to link the code of this program with any version of the * + * OpenSSL library which is distributed under a license identical to that * + * listed in the included docs/licenses/OpenSSL.txt file, and distribute * + * linked combinations including the two. * + * * + * Any redistribution of Covered Software, including any derived works, * + * must obey and carry forward all of the terms of this license, including * + * obeying all GPL rules and restrictions. For example, source code of * + * the whole work must be provided and free redistribution must be * + * allowed. All GPL references to "this License", are to be treated as * + * including the terms and conditions of this license text as well. * + * * + * Because this license imposes special exceptions to the GPL, Covered * + * Work may not be combined (even as part of a larger work) with plain GPL * + * software. The terms, conditions, and exceptions of this license must * + * be included as well. This license is incompatible with some other open * + * source licenses as well. In some cases we can relicense portions of * + * RFIDler or grant special permissions to use it in other open source * + * software. Please contact sales@aperturelabs.com with any such requests.* + * Similarly, we don't incorporate incompatible open source software into * + * Covered Software without special permission from the copyright holders. * + * * + * If you have any questions about the licensing restrictions on using * + * RFIDler in other works, are happy to help. As mentioned above, we also * + * offer alternative license to integrate RFIDler into proprietary * + * applications and appliances. These contracts have been sold to dozens * + * of software vendors, and generally include a perpetual license as well * + * as providing for priority support and updates. They also fund the * + * continued development of RFIDler. Please email sales@aperturelabs.com * + * for further information. * + * If you have received a written license agreement or contract for * + * Covered Software stating terms other than these, you may choose to use * + * and redistribute Covered Software under those terms instead of these. * + * * + * Source is provided to this software because we believe users have a * + * right to know exactly what a program is going to do before they run it. * + * This also allows you to audit the software for security holes (none * + * have been found so far). * + * * + * Source code also allows you to port RFIDler to new platforms, fix bugs, * + * and add new features. You are highly encouraged to send your changes * + * to the RFIDler mailing list for possible incorporation into the * + * main distribution. By sending these changes to Aperture Labs Ltd. or * + * one of the Aperture Labs Ltd. development mailing lists, or checking * + * them into the RFIDler source code repository, it is understood (unless * + * you specify otherwise) that you are offering the RFIDler Project * + * (Aperture Labs Ltd.) the unlimited, non-exclusive right to reuse, * + * modify, and relicense the code. RFIDler will always be available Open * + * Source, but this is important because the inability to relicense code * + * has caused devastating problems for other Free Software projects (such * + * as KDE and NASM). We also occasionally relicense the code to third * + * parties as discussed above. If you wish to specify special license * + * conditions of your contributions, just say so when you send them. * + * * + * This program 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 RFIDler * + * license file for more details (it's in a COPYING file included with * + * RFIDler, and also available from * + * https://github.com/ApertureLabsLtd/RFIDler/COPYING * + * * + ***************************************************************************/ + +// Author: unknown. +// Modifications for RFIDler: Tony Naggs , Adam Laurie + + +#ifndef HITAGCRYPTO_H +#define HITAGCRYPTO_H + +#include + +/* + Our model of Hitag 2 crypto uses 2 parallel shift registers: + a. 48 bit Feedback Shift Register, required for inputs to the nonlinear function. + b. 48 bit Linear Feedback Shift Register (LFSR). + A transform of initial register (a) value, which is then run in parallel. + Enables much faster calculation of the feedback values. + + API: + void hitag2_init(Hitag_State* pstate, uint64_t sharedkey, uint32_t serialnum, + uint32_t initvector); + Initialise state from 48 bit shared (secret) reader/tag key, + 32 bit tag serial number and 32 bit initialisation vector from reader. + + uint32_t hitag2_nstep(Hitag_State* pstate, uint32_t steps); + update shift register state and generate N cipher bits (N should be <= 32) + */ + + +typedef struct { + uint64_t shiftreg; // naive shift register, required for nonlinear fn input + uint64_t lfsr; // fast lfsr, used to make software faster +} Hitag_State; + +void hitag2_init(Hitag_State *pstate, uint64_t sharedkey, uint32_t serialnum, uint32_t initvector); + +uint32_t hitag2_nstep(Hitag_State *pstate, uint32_t steps); + +#endif /* HITAGCRYPTO_H */ + diff --git a/tools/hitag2crack/crack5gpu/ht2crack2utils.c b/tools/hitag2crack/crack5gpu/ht2crack2utils.c new file mode 100644 index 000000000..75d4c27a9 --- /dev/null +++ b/tools/hitag2crack/crack5gpu/ht2crack2utils.c @@ -0,0 +1,172 @@ +#include "ht2crack2utils.h" + +// writes a value into a buffer as a series of bytes +void writebuf(unsigned char *buf, uint64_t val, unsigned int len) { + int i; + char c; + + for (i = len - 1; i >= 0; i--) { + c = val & 0xff; + buf[i] = c; + val = val >> 8; + } + +} + + +/* simple hexdump for testing purposes */ +void shexdump(unsigned char *data, int data_len) { + int i; + + if (!data || (data_len <= 0)) { + printf("shexdump: invalid parameters\n"); + return; + } + + printf("Hexdump from %p:\n", data); + + for (i = 0; i < data_len; i++) { + if ((i % HEX_PER_ROW) == 0) { + printf("\n0x%04x: ", i); + } + printf("%02x ", data[i]); + } + printf("\n\n"); +} + + + +void printbin(unsigned char *c) { + int i, j; + unsigned char x; + + if (!c) { + printf("printbin: invalid params\n"); + return; + } + + for (i = 0; i < 6; i++) { + x = c[i]; + for (j = 0; j < 8; j++) { + printf("%d", (x & 0x80) >> 7); + x = x << 1; + } + } + printf("\n"); +} + + +void printbin2(uint64_t val, unsigned int size) { + int i; + uint64_t mask = 1; + + mask = mask << (size - 1); + + for (i = 0; i < size; i++) { + if (val & mask) { + printf("1"); + } else { + printf("0"); + } + val = val << 1; + } +} + + +void printstate(Hitag_State *hstate) { + printf("shiftreg =\t"); + printbin2(hstate->shiftreg, 48); + printf("\n"); +} + + + + +// convert hex char to binary +unsigned char hex2bin(unsigned char c) { + if ((c >= '0') && (c <= '9')) { + return (c - '0'); + } else if ((c >= 'a') && (c <= 'f')) { + return (c - 'a' + 10); + } else if ((c >= 'A') && (c <= 'F')) { + return (c - 'A' + 10); + } else { + return 0; + } +} + +// return a single bit from a value +int bitn(uint64_t x, int bit) { + uint64_t bitmask = 1; + + bitmask = bitmask << bit; + + if (x & bitmask) { + return 1; + } else { + return 0; + } +} + + +// the sub-function R that rollback depends upon +int fnR(uint64_t x) { + // renumbered bits because my state is 0-47, not 1-48 + return (bitn(x, 1) ^ bitn(x, 2) ^ bitn(x, 5) ^ bitn(x, 6) ^ bitn(x, 7) ^ + bitn(x, 15) ^ bitn(x, 21) ^ bitn(x, 22) ^ bitn(x, 25) ^ bitn(x, 29) ^ bitn(x, 40) ^ + bitn(x, 41) ^ bitn(x, 42) ^ bitn(x, 45) ^ bitn(x, 46) ^ bitn(x, 47)); +} + +// the rollback function that lets us go backwards in time +void rollback(Hitag_State *hstate, unsigned int steps) { + int i; + + for (i = 0; i < steps; i++) { + hstate->shiftreg = ((hstate->shiftreg << 1) & 0xffffffffffff) | fnR(hstate->shiftreg); + } + +} + + +// the three filter sub-functions that feed fnf +int fa(unsigned int i) { + return bitn(0x2C79, i); +} + +int fb(unsigned int i) { + return bitn(0x6671, i); +} + +int fc(unsigned int i) { + return bitn(0x7907287B, i); +} + +// the filter function that generates a bit of output from the prng state +int fnf(uint64_t s) { + unsigned int x1, x2, x3, x4, x5, x6; + + x1 = (bitn(s, 2) << 0) | (bitn(s, 3) << 1) | (bitn(s, 5) << 2) | (bitn(s, 6) << 3); + x2 = (bitn(s, 8) << 0) | (bitn(s, 12) << 1) | (bitn(s, 14) << 2) | (bitn(s, 15) << 3); + x3 = (bitn(s, 17) << 0) | (bitn(s, 21) << 1) | (bitn(s, 23) << 2) | (bitn(s, 26) << 3); + x4 = (bitn(s, 28) << 0) | (bitn(s, 29) << 1) | (bitn(s, 31) << 2) | (bitn(s, 33) << 3); + x5 = (bitn(s, 34) << 0) | (bitn(s, 43) << 1) | (bitn(s, 44) << 2) | (bitn(s, 46) << 3); + + x6 = (fa(x1) << 0) | (fb(x2) << 1) | (fb(x3) << 2) | (fb(x4) << 3) | (fa(x5) << 4); + + return fc(x6); +} + +// builds the lfsr for the prng (quick calcs for hitag2_nstep()) +void buildlfsr(Hitag_State *hstate) { + uint64_t state = hstate->shiftreg; + uint64_t temp; + + temp = state ^ (state >> 1); + hstate->lfsr = state ^ (state >> 6) ^ (state >> 16) + ^ (state >> 26) ^ (state >> 30) ^ (state >> 41) + ^ (temp >> 2) ^ (temp >> 7) ^ (temp >> 22) + ^ (temp >> 42) ^ (temp >> 46); +} + + + diff --git a/tools/hitag2crack/crack5gpu/ht2crack2utils.h b/tools/hitag2crack/crack5gpu/ht2crack2utils.h new file mode 100644 index 000000000..33e0e3036 --- /dev/null +++ b/tools/hitag2crack/crack5gpu/ht2crack2utils.h @@ -0,0 +1,35 @@ +#include +#include +#include +#include +#include +#include +#include +#include +#include +#include + +#include "HardwareProfile.h" +#include "rfidler.h" +#include "util.h" + +#include "hitagcrypto.h" + +#define HEX_PER_ROW 16 + + + +void writebuf(unsigned char *buf, uint64_t val, unsigned int len); +void shexdump(unsigned char *data, int data_len); +void printbin(unsigned char *c); +void printbin2(uint64_t val, unsigned int size); +void printstate(Hitag_State *hstate); +unsigned char hex2bin(unsigned char c); +int bitn(uint64_t x, int bit); +int fnR(uint64_t x); +void rollback(Hitag_State *hstate, unsigned int steps); +int fa(unsigned int i); +int fb(unsigned int i); +int fc(unsigned int i); +int fnf(uint64_t s); +void buildlfsr(Hitag_State *hstate); diff --git a/tools/hitag2crack/crack5gpu/ht2crack5.c b/tools/hitag2crack/crack5gpu/ht2crack5.c new file mode 100644 index 000000000..82a80c3ea --- /dev/null +++ b/tools/hitag2crack/crack5gpu/ht2crack5.c @@ -0,0 +1,421 @@ +/* ht2crack5.c + * + * This code is heavily based on the HiTag2 Hell CPU implementation + * from https://github.com/factoritbv/hitag2hell by FactorIT B.V., + * with the following changes: + * * Main takes a UID and 2 {nR},{aR} pairs as arguments + * and searches for states producing the first aR sample, + * reconstructs the corresponding key candidates + * and tests them against the second nR,aR pair; + * * Reduce max_bitslices and some type sizes to fit OpenCL + * * Reuses the Hitag helping functions of the other attacks. + */ + +#include +#include +#include +#include +#include +#include +#include +#include +#include +#include +#include +#ifdef __APPLE__ +#include +#else +#define CL_TARGET_OPENCL_VERSION 220 +#define CL_USE_DEPRECATED_OPENCL_1_2_APIS +#include +#endif +#include "ht2crack2utils.h" + +const uint8_t bits[9] = {20, 14, 4, 3, 1, 1, 1, 1, 1}; +#define lfsr_inv(state) (((state)<<1) | (__builtin_parityll((state) & ((0xce0044c101cd>>1)|(1ull<<(47)))))) +#define i4(x,a,b,c,d) ((uint32_t)((((x)>>(a))&1)<<3)|(((x)>>(b))&1)<<2|(((x)>>(c))&1)<<1|(((x)>>(d))&1)) +#define f(state) ((0xdd3929b >> ( (((0x3c65 >> i4(state, 2, 3, 5, 6) ) & 1) <<4) \ + | ((( 0xee5 >> i4(state, 8,12,14,15) ) & 1) <<3) \ + | ((( 0xee5 >> i4(state,17,21,23,26) ) & 1) <<2) \ + | ((( 0xee5 >> i4(state,28,29,31,33) ) & 1) <<1) \ + | (((0x3c65 >> i4(state,34,43,44,46) ) & 1) ))) & 1) + +#define MAX_BITSLICES 32 +#define VECTOR_SIZE (MAX_BITSLICES/8) +#define KERNELFILENAME "ht2crack5kernel.cl" + +typedef unsigned int __attribute__((aligned(VECTOR_SIZE))) __attribute__((vector_size(VECTOR_SIZE))) bitslice_value_t; +typedef union { + bitslice_value_t value; + uint8_t bytes[MAX_BITSLICES / 8]; +} bitslice_t; + +// we never actually set or use the lowest 2 bits the initial state, so we can save 2 bitslices everywhere +__thread bitslice_t state[-2 + 32 + 48]; + +bitslice_t keystream[32]; +bitslice_t bs_zeroes, bs_ones; + +#define f_a_bs(a,b,c,d) (~(((a|b)&c)^(a|d)^b)) // 6 ops +#define f_b_bs(a,b,c,d) (~(((d|c)&(a^b))^(d|a|b))) // 7 ops +#define f_c_bs(a,b,c,d,e) (~((((((c^e)|d)&a)^b)&(c^b))^(((d^e)|a)&((d^b)|c)))) // 13 ops +#define lfsr_bs(i) (state[-2+i+ 0].value ^ state[-2+i+ 2].value ^ state[-2+i+ 3].value ^ state[-2+i+ 6].value ^ \ + state[-2+i+ 7].value ^ state[-2+i+ 8].value ^ state[-2+i+16].value ^ state[-2+i+22].value ^ \ + state[-2+i+23].value ^ state[-2+i+26].value ^ state[-2+i+30].value ^ state[-2+i+41].value ^ \ + state[-2+i+42].value ^ state[-2+i+43].value ^ state[-2+i+46].value ^ state[-2+i+47].value); +#define get_bit(n, word) ((word >> (n)) & 1) + +const uint64_t expand(uint64_t mask, uint64_t value) { + uint64_t fill = 0; + for (uint64_t bit_index = 0; bit_index < 48; bit_index++) { + if (mask & 1) { + fill |= (value & 1) << bit_index; + value >>= 1; + } + mask >>= 1; + } + return fill; +} + +void bitslice(const uint64_t value, bitslice_t *restrict bitsliced_value, const size_t bit_len, bool reverse) { + size_t bit_idx; + for (bit_idx = 0; bit_idx < bit_len; bit_idx++) { + bool bit; + if (reverse) { + bit = get_bit(bit_len - 1 - bit_idx, value); + } else { + bit = get_bit(bit_idx, value); + } + if (bit) { + bitsliced_value[bit_idx].value = bs_ones.value; + } else { + bitsliced_value[bit_idx].value = bs_zeroes.value; + } + } +} + +uint32_t uid, nR1, aR1, nR2, aR2; + +// Reduce type size of candidates array to fit OpenCL +uint16_t candidates[(1 << 20) * 3]; +bitslice_t initial_bitslices[48]; +size_t filter_pos[20] = {4, 7, 9, 13, 16, 18, 22, 24, 27, 30, 32, 35, 45, 47 }; +size_t thread_count = 8; +size_t layer_0_found; + +static void try_state(uint64_t s); + +struct context { + char *kernelSource; // source for kernel + + cl_platform_id platform_id; // compute platform id + cl_device_id device_id; // compute device id + cl_context context; // compute context + cl_command_queue commands; // compute command queue + cl_program program; // compute program + cl_kernel kernel; // compute kernel + +// cl_mem cand_base; // device memory used for the candidate base + cl_mem keystream; // device memory used for the keystream array + cl_mem candidates; // device memory used for the candidates array + cl_mem matches; // device memory used for the matches array + cl_mem matches_found; // device memory used for the matches_found array +}; + + +void runKernel(struct context *ctx, uint32_t cand_base, uint64_t *matches, uint32_t *matches_found) { + int err; + size_t global[2]; + + // Write our data set into the input array in device memory + err = clEnqueueWriteBuffer(ctx->commands, ctx->matches_found, CL_TRUE, 0, sizeof(uint32_t), matches_found, 0, NULL, NULL); + // Set the arguments to our compute kernel + err = clSetKernelArg(ctx->kernel, 0, sizeof(uint32_t), &cand_base); + err |= clSetKernelArg(ctx->kernel, 4, sizeof(cl_mem), &ctx->matches_found); + if (err != CL_SUCCESS) { + printf("Error: Failed to set kernel arguments in runKernel! %d\n", err); + exit(1); + } + + // Execute the kernel over the entire range of our 2d input data set using 8K * 1K threads + global[0] = 8192; + global[1] = 1024; + err = clEnqueueNDRangeKernel(ctx->commands, ctx->kernel, 2, NULL, global, NULL, 0, NULL, NULL); + if (err) { + printf("Error: Failed to execute kernel!\n"); + exit(1); + } + + // Wait for the command commands to get serviced before reading back results + err = clFinish(ctx->commands); + if (err) { + printf("Error: Failed to execute kernel! clFinish = %d\n", err); + exit(1); + } + + // Read back the results from the device to verify the output + err = clEnqueueReadBuffer(ctx->commands, ctx->matches, CL_TRUE, 0, sizeof(uint64_t) * 8192, matches, 0, NULL, NULL); + if (err != CL_SUCCESS) { + printf("Error: Failed to read matches array! %d\n", err); + exit(1); + } + + err = clEnqueueReadBuffer(ctx->commands, ctx->matches_found, CL_TRUE, 0, sizeof(uint32_t), matches_found, 0, NULL, NULL); + if (err != CL_SUCCESS) { + printf("Error: Failed to read matches_found! %d\n", err); + exit(1); + } +} + +int main(int argc, char *argv[]) { + memset(candidates, 0, sizeof(candidates)); + struct context ctx; + uint64_t matches[8192]; + uint32_t matches_found[1]; + + // set constants + memset(bs_ones.bytes, 0xff, VECTOR_SIZE); + memset(bs_zeroes.bytes, 0x00, VECTOR_SIZE); + + uint32_t target = 0; + + if (argc < 6) { + printf("%s UID {nR1} {aR1} {nR2} {aR2}\n", argv[0]); + exit(1); + } + + if (!strncmp(argv[1], "0x", 2) || !strncmp(argv[1], "0X", 2)) { + uid = rev32(hexreversetoulong(argv[1] + 2)); + } else { + uid = rev32(hexreversetoulong(argv[1])); + } + + if (!strncmp(argv[2], "0x", 2) || !strncmp(argv[2], "0X", 2)) { + nR1 = rev32(hexreversetoulong(argv[2] + 2)); + } else { + nR1 = rev32(hexreversetoulong(argv[2])); + } + + aR1 = strtol(argv[3], NULL, 16); + + if (!strncmp(argv[4], "0x", 2) || !strncmp(argv[4], "0X", 2)) { + nR2 = rev32(hexreversetoulong(argv[4] + 2)); + } else { + nR2 = rev32(hexreversetoulong(argv[4])); + } + + aR2 = strtol(argv[5], NULL, 16); + + target = ~aR1; + // bitslice inverse target bits + bitslice(~target, keystream, 32, true); + + // bitslice all possible 256 values in the lowest 8 bits + memset(initial_bitslices[0].bytes, 0xaa, VECTOR_SIZE); + memset(initial_bitslices[1].bytes, 0xcc, VECTOR_SIZE); + memset(initial_bitslices[2].bytes, 0xf0, VECTOR_SIZE); + size_t interval = 1; + for (size_t bit = 3; bit < 8; bit++) { + for (size_t byte = 0; byte < VECTOR_SIZE;) { + for (size_t length = 0; length < interval; length++) { + initial_bitslices[bit].bytes[byte++] = 0x00; + } + for (size_t length = 0; length < interval; length++) { + initial_bitslices[bit].bytes[byte++] = 0xff; + } + } + interval <<= 1; + } + + // compute layer 0 output + for (size_t i0 = 0; i0 < 1 << 20; i0++) { + uint64_t state0 = expand(0x5806b4a2d16c, i0); + + if (f(state0) == target >> 31) { + // cf kernel, state is now split in 3 shorts >> 2 + candidates[(layer_0_found * 3) + 0] = (uint16_t)((state0 >> (32 + 2)) & 0xffff); + candidates[(layer_0_found * 3) + 1] = (uint16_t)((state0 >> (16 + 2)) & 0xffff); + candidates[(layer_0_found * 3) + 2] = (uint16_t)((state0 >> (0 + 2)) & 0xffff); + layer_0_found++; + } + } + + // load OpenCL kernel source + //////////////////////////// + struct stat filestat; + int fd; + + fd = open(KERNELFILENAME, O_RDONLY); + if (fd <= 0) { + printf("Cannot open %s\n", KERNELFILENAME); + exit(1); + } + + if (fstat(fd, &filestat)) { + printf("Cannot stat %s\n", KERNELFILENAME); + exit(1); + } + + ctx.kernelSource = (char *)malloc(filestat.st_size); + if (!ctx.kernelSource) { + printf("Cannot malloc kernelSource\n"); + exit(1); + } + + if (read(fd, ctx.kernelSource, filestat.st_size) < filestat.st_size) { + printf("Cannot read %s\n", KERNELFILENAME); + exit(1); + } + + close(fd); + + // discover and set up compute device + ///////////////////////////////////// + int err; + + // Connect to a compute device + err = clGetPlatformIDs(1, &(ctx.platform_id), NULL); + if (err != CL_SUCCESS) { + printf("Error: Failed to get platform id: %d\n", err); + exit(1); + } + + int gpu = 1; + err = clGetDeviceIDs(ctx.platform_id, gpu ? CL_DEVICE_TYPE_GPU : CL_DEVICE_TYPE_CPU, 1, &(ctx.device_id), NULL); + if (err != CL_SUCCESS) { + printf("Error: Failed to create a device group!: %d\n", err); + exit(1); + } + + // Create a compute context + ctx.context = clCreateContext(0, 1, &(ctx.device_id), NULL, NULL, &err); + if (!ctx.context) { + printf("Error: Failed to create a compute context!\n"); + exit(1); + } + + // Create a command commands + ctx.commands = clCreateCommandQueue(ctx.context, ctx.device_id, 0, &err); + if (!ctx.commands) { + printf("Error: Failed to create a command commands!\n"); + exit(1); + } + + // Create the compute program from the source buffer + ctx.program = clCreateProgramWithSource(ctx.context, 1, (const char **) & (ctx.kernelSource), NULL, &err); + if (!ctx.program) { + printf("Error: Failed to create compute program!\n"); + exit(1); + } + + // Build the program executable + err = clBuildProgram(ctx.program, 0, NULL, "-Werror", NULL, NULL); + + if (err != CL_SUCCESS) { + size_t len; + char buffer[1024 * 1024]; + + printf("Error: Failed to build program executable!\n"); + err = clGetProgramBuildInfo(ctx.program, ctx.device_id, CL_PROGRAM_BUILD_LOG, sizeof(buffer), buffer, &len); + if (err != CL_SUCCESS) { + printf("clGetProgramBuildInfo failed: %d\n", err); + exit(1); + } else { + printf("%s\n", buffer); + exit(1); + } + } + + // Create the compute kernel in the program we wish to run + ctx.kernel = clCreateKernel(ctx.program, "find_state", &err); + if (!ctx.kernel || err != CL_SUCCESS) { + printf("Error: Failed to create compute kernel!\n"); + exit(1); + } + + ctx.candidates = clCreateBuffer(ctx.context, CL_MEM_READ_ONLY, sizeof(uint16_t) * ((1 << 20) * 3), NULL, NULL); + ctx.keystream = clCreateBuffer(ctx.context, CL_MEM_READ_ONLY, VECTOR_SIZE * 32, NULL, NULL); + + ctx.matches = clCreateBuffer(ctx.context, CL_MEM_WRITE_ONLY, sizeof(uint64_t) * 8192, NULL, NULL); + ctx.matches_found = clCreateBuffer(ctx.context, CL_MEM_READ_WRITE, sizeof(uint32_t), NULL, NULL); + + if (!ctx.candidates || !ctx.keystream || !ctx.matches || !ctx.matches_found) { + printf("Error: Failed to allocate device memory!\n"); + exit(1); + } + + // set up constant vars + /////////////////////// + + // Write our data set into the input array in device memory + err = clEnqueueWriteBuffer(ctx.commands, ctx.keystream, CL_TRUE, 0, VECTOR_SIZE * 32, keystream, 0, NULL, NULL); + if (err != CL_SUCCESS) { + printf("Error: Failed to write to keystream array!\n"); + exit(1); + } + + err = clEnqueueWriteBuffer(ctx.commands, ctx.candidates, CL_TRUE, 0, sizeof(uint16_t) * ((1 << 20) * 3), candidates, 0, NULL, NULL); + if (err != CL_SUCCESS) { + printf("Error: Failed to write to candidates array!\n"); + exit(1); + } + + // Set the arguments to our compute kernel + err = clSetKernelArg(ctx.kernel, 1, sizeof(cl_mem), &ctx.candidates); + err |= clSetKernelArg(ctx.kernel, 2, sizeof(cl_mem), &ctx.keystream); + err |= clSetKernelArg(ctx.kernel, 3, sizeof(cl_mem), &ctx.matches); + if (err != CL_SUCCESS) { + printf("Error: Failed to set kernel arguments! %d\n", err); + exit(1); + } + + // run kernel + ///////////// + for (uint32_t step = 0; step < 64; step++) { + printf("slice %3u/64: ", step + 1); + fflush(stdout); + matches_found[0] = 0; + runKernel(&ctx, step << 13, matches, matches_found); + + printf("%5u candidates\n", matches_found[0]); + for (uint32_t match = 0; match < matches_found[0]; match++) { + try_state(matches[match]); + } + } + + printf("Key not found\n"); + exit(1); +} + +static void try_state(uint64_t s) { + Hitag_State hstate; + uint64_t keyrev, key, nR1xk; + uint32_t b = 0; + + hstate.shiftreg = s; + rollback(&hstate, 2); + + // recover key + keyrev = hstate.shiftreg & 0xffff; + nR1xk = (hstate.shiftreg >> 16) & 0xffffffff; + for (int i = 0; i < 32; i++) { + hstate.shiftreg = ((hstate.shiftreg) << 1) | ((uid >> (31 - i)) & 0x1); + b = (b << 1) | fnf(hstate.shiftreg); + } + keyrev |= (nR1xk ^ nR1 ^ b) << 16; + + // test key + hitag2_init(&hstate, keyrev, uid, nR2); + if ((aR2 ^ hitag2_nstep(&hstate, 32)) == 0xffffffff) { + + key = rev64(keyrev); + + printf("Key: "); + for (int i = 0; i < 6; i++) { + printf("%02X", (uint8_t)(key & 0xff)); + key = key >> 8; + } + printf("\n"); + exit(0); + } +} diff --git a/tools/hitag2crack/crack5gpu/ht2crack5kernel.cl b/tools/hitag2crack/crack5gpu/ht2crack5kernel.cl new file mode 100644 index 000000000..3f1803ee8 --- /dev/null +++ b/tools/hitag2crack/crack5gpu/ht2crack5kernel.cl @@ -0,0 +1,429 @@ +/* ht2crack5kernel.cl + * + * This code is heavily based on the HiTag2 Hell CPU implementation + * from https://github.com/factoritbv/hitag2hell by FactorIT B.V. + * This file is the file openocl.cl with the following change: + * * promote keystream from constant to argument. + */ + +#define MAX_BITSLICES 32 +#define KEYSTREAM_LENGTH 32 +typedef uint bitslice_t __attribute__((aligned(MAX_BITSLICES / 8))); + +inline uint lut3(uint a, uint b, uint c, uint imm) { + uint r; + asm("lop3.b32 %0, %1, %2, %3, %4;" + : "=r"(r) + : "r"(a), "r"(b), "r"(c), "i"(imm)); + return r; +} +#define f_a_bs_lut_1 (((0xf0|0xcc)&0xaa)^0xcc) +#define f_a_bs_lut_2 (~((0xf0|0xcc)^0xaa)) +#define f_a_bs(a,b,c,d) ((lut3(a,d,lut3(a,b,c,f_a_bs_lut_1),f_a_bs_lut_2))) // 2 luts + +#define f_b_bs_lut_1 (((0xf0|0xcc)&0xaa)) +#define f_b_bs_lut_2 (~((0xf0|0xcc|0xaa))) +#define f_b_bs(a,b,c,d) ((lut3(d,c,a^b,f_b_bs_lut_1)^lut3(d,a,b, f_b_bs_lut_2))) // 2 luts, 2 xors + +#define f_c_bs_lut_1 (((0xf0^0xcc)|0xaa)) +#define f_c_bs_lut_2 (~((0xf0^0xcc)&(0xaa^0xcc))) + +// 4 luts, 2 ands, 1 xor +#define f_c_bs(a,b,c,d,e) (((lut3((lut3(c,e,d, f_c_bs_lut_1) & a), b, c, f_c_bs_lut_2)) ^ (lut3(d,e,a, f_c_bs_lut_1) & lut3(d,b,c,f_c_bs_lut_1)))) + +// non-lut version of F: 20 lookups + 6*2 + 7*3 + 13 + = 66 ops +// lut version: 20 lookups + 2*2 + 4*3 + 7 + = 43 ops + +#define lfsr_lut (0xf0^0xaa^0xcc) +// 7 luts, 1 xor +#define lfsr_bs(i) ( lut3(lut3(lut3(state[-2+i+ 0], state[-2+i+ 2], state[-2+i+ 3], lfsr_lut), \ + lut3(state[-2+i+ 6], state[-2+i+ 7], state[-2+i+ 8], lfsr_lut), \ + lut3(state[-2+i+16], state[-2+i+22], state[-2+i+23], lfsr_lut), \ + lfsr_lut), \ + lut3(state[-2+i+26], state[-2+i+30], state[-2+i+41], lfsr_lut), \ + lut3(state[-2+i+42], state[-2+i+43], state[-2+i+46], lfsr_lut), lfsr_lut) ^ state[-2+i+47]) + +// 46 iterations * 4 ops +inline void bitslice(bitslice_t *restrict b, ulong x, const uchar n) { + for (uchar i = 0; i < n; ++i) { + b[i] = -(x & 1); + x >>= 1; + } +} + +// don't care about the complexity of this function +inline ulong unbitslice(const bitslice_t *restrict b, const uchar s, const uchar n) { + const bitslice_t mask = ((bitslice_t) 1) << s; + ulong result = 0; + for (char i = n - 1; i >= 0; --i) { + result <<= 1; + result |= (bool)(b[i] & mask); + } + return result; +} + +// format this array with 32 bitsliced vectors of ones and zeroes representing the inverted keystream + +__kernel +__attribute__((vec_type_hint(bitslice_t))) +void find_state(const uint candidate_index_base, + __global const ushort *restrict candidates, + __global const bitslice_t *restrict keystream, + __global ulong *restrict matches, + __global uint *restrict matches_found) { + // we never actually set or use the lowest 2 bits the initial state, so we can save 2 bitslices everywhere + bitslice_t state[-2 + 48 + KEYSTREAM_LENGTH]; + // set bits 0+2, 0+3, 0+5, 0+6, 0+8, 0+12, 0+14, 0+15, 0+17, 0+21, 0+23, 0+26, 0+28, 0+29, 0+31, 0+33, 0+34, 0+43, 0+44, 0+46 + // get the 48-bit cipher states as 3 16-bit words from the host memory queue (to save 25% throughput) + const uint index = 3 * (candidate_index_base + get_global_id(0)); // dimension 0 should at least keep the execution units saturated - 8k is fine + const ulong candidate = ((ulong) candidates[index] << 32) | ((ulong) candidates[index + 1] << 16) | candidates[index + 2]; + // set all 48 state bits except the lowest 2 + bitslice(&state[-2 + 2], candidate, 46); + // set bits 3, 6, 8, 12, 15 + state[-2 + 1 + 3] = 0xaaaaaaaa; + state[-2 + 1 + 6] = 0xcccccccc; + state[-2 + 1 + 8] = 0xf0f0f0f0; + state[-2 + 1 + 12] = 0xff00ff00; + state[-2 + 1 + 15] = 0xffff0000; + ushort i1 = get_global_id(1); // dimension 1 should be 1024 + state[-2 + 18] = -((bool)(i1 & 0x1)); + state[-2 + 22] = -((bool)(i1 & 0x2)); + state[-2 + 24] = -((bool)(i1 & 0x4)); + state[-2 + 27] = -((bool)(i1 & 0x8)); + state[-2 + 30] = -((bool)(i1 & 0x10)); + state[-2 + 32] = -((bool)(i1 & 0x20)); + state[-2 + 35] = -((bool)(i1 & 0x40)); + state[-2 + 45] = -((bool)(i1 & 0x80)); + state[-2 + 47] = -((bool)(i1 & 0x100)); + state[-2 + 48] = -((bool)(i1 & 0x200)); // guess lfsr output 0 + // 0xfc07fef3f9fe + const bitslice_t filter1_0 = f_a_bs(state[-2 + 3], state[-2 + 4], state[-2 + 6], state[-2 + 7]); + const bitslice_t filter1_1 = f_b_bs(state[-2 + 9], state[-2 + 13], state[-2 + 15], state[-2 + 16]); + const bitslice_t filter1_2 = f_b_bs(state[-2 + 18], state[-2 + 22], state[-2 + 24], state[-2 + 27]); + const bitslice_t filter1_3 = f_b_bs(state[-2 + 29], state[-2 + 30], state[-2 + 32], state[-2 + 34]); + const bitslice_t filter1_4 = f_a_bs(state[-2 + 35], state[-2 + 44], state[-2 + 45], state[-2 + 47]); + const bitslice_t filter1 = f_c_bs(filter1_0, filter1_1, filter1_2, filter1_3, filter1_4); + const bitslice_t results1 = filter1 ^ keystream[1]; + if (!results1) return; + const bitslice_t filter2_0 = f_a_bs(state[-2 + 4], state[-2 + 5], state[-2 + 7], state[-2 + 8]); + const bitslice_t filter2_3 = f_b_bs(state[-2 + 30], state[-2 + 31], state[-2 + 33], state[-2 + 35]); + const bitslice_t filter3_0 = f_a_bs(state[-2 + 5], state[-2 + 6], state[-2 + 8], state[-2 + 9]); + const bitslice_t filter5_2 = f_b_bs(state[-2 + 22], state[-2 + 26], state[-2 + 28], state[-2 + 31]); + const bitslice_t filter6_2 = f_b_bs(state[-2 + 23], state[-2 + 27], state[-2 + 29], state[-2 + 32]); + const bitslice_t filter7_2 = f_b_bs(state[-2 + 24], state[-2 + 28], state[-2 + 30], state[-2 + 33]); + const bitslice_t filter9_1 = f_b_bs(state[-2 + 17], state[-2 + 21], state[-2 + 23], state[-2 + 24]); + const bitslice_t filter9_2 = f_b_bs(state[-2 + 26], state[-2 + 30], state[-2 + 32], state[-2 + 35]); + const bitslice_t filter10_0 = f_a_bs(state[-2 + 12], state[-2 + 13], state[-2 + 15], state[-2 + 16]); + const bitslice_t filter11_0 = f_a_bs(state[-2 + 13], state[-2 + 14], state[-2 + 16], state[-2 + 17]); + const bitslice_t filter12_0 = f_a_bs(state[-2 + 14], state[-2 + 15], state[-2 + 17], state[-2 + 18]); + const bitslice_t filter14_1 = f_b_bs(state[-2 + 22], state[-2 + 26], state[-2 + 28], state[-2 + 29]); + const bitslice_t filter15_1 = f_b_bs(state[-2 + 23], state[-2 + 27], state[-2 + 29], state[-2 + 30]); + const bitslice_t filter15_3 = f_b_bs(state[-2 + 43], state[-2 + 44], state[-2 + 46], state[-2 + 48]); + const bitslice_t filter16_1 = f_b_bs(state[-2 + 24], state[-2 + 28], state[-2 + 30], state[-2 + 31]); + for (uchar i2 = 0; i2 < (1 << 5);) { + state[-2 + 10] = -((bool)(i2 & 0x1)); + state[-2 + 19] = -((bool)(i2 & 0x2)); + state[-2 + 25] = -((bool)(i2 & 0x4)); + state[-2 + 36] = -((bool)(i2 & 0x8)); + state[-2 + 49] = -((bool)(i2 & 0x10)); // guess lfsr output 1 + i2++; + // 0xfe07fffbfdff + const bitslice_t filter2_1 = f_b_bs(state[-2 + 10], state[-2 + 14], state[-2 + 16], state[-2 + 17]); + const bitslice_t filter2_2 = f_b_bs(state[-2 + 19], state[-2 + 23], state[-2 + 25], state[-2 + 28]); + const bitslice_t filter2_4 = f_a_bs(state[-2 + 36], state[-2 + 45], state[-2 + 46], state[-2 + 48]); + const bitslice_t filter2 = f_c_bs(filter2_0, filter2_1, filter2_2, filter2_3, filter2_4); + const bitslice_t results2 = results1 & (filter2 ^ keystream[2]); + if (!results2) continue; + state[-2 + 50] = lfsr_bs(2); + const bitslice_t filter3_3 = f_b_bs(state[-2 + 31], state[-2 + 32], state[-2 + 34], state[-2 + 36]); + const bitslice_t filter4_0 = f_a_bs(state[-2 + 6], state[-2 + 7], state[-2 + 9], state[-2 + 10]); + const bitslice_t filter4_1 = f_b_bs(state[-2 + 12], state[-2 + 16], state[-2 + 18], state[-2 + 19]); + const bitslice_t filter4_2 = f_b_bs(state[-2 + 21], state[-2 + 25], state[-2 + 27], state[-2 + 30]); + const bitslice_t filter7_0 = f_a_bs(state[-2 + 9], state[-2 + 10], state[-2 + 12], state[-2 + 13]); + const bitslice_t filter7_1 = f_b_bs(state[-2 + 15], state[-2 + 19], state[-2 + 21], state[-2 + 22]); + const bitslice_t filter8_2 = f_b_bs(state[-2 + 25], state[-2 + 29], state[-2 + 31], state[-2 + 34]); + const bitslice_t filter10_1 = f_b_bs(state[-2 + 18], state[-2 + 22], state[-2 + 24], state[-2 + 25]); + const bitslice_t filter10_2 = f_b_bs(state[-2 + 27], state[-2 + 31], state[-2 + 33], state[-2 + 36]); + const bitslice_t filter11_1 = f_b_bs(state[-2 + 19], state[-2 + 23], state[-2 + 25], state[-2 + 26]); + const bitslice_t filter13_0 = f_a_bs(state[-2 + 15], state[-2 + 16], state[-2 + 18], state[-2 + 19]); + const bitslice_t filter13_1 = f_b_bs(state[-2 + 21], state[-2 + 25], state[-2 + 27], state[-2 + 28]); + const bitslice_t filter16_0 = f_a_bs(state[-2 + 18], state[-2 + 19], state[-2 + 21], state[-2 + 22]); + const bitslice_t filter16_3 = f_b_bs(state[-2 + 44], state[-2 + 45], state[-2 + 47], state[-2 + 49]); + const bitslice_t filter17_1 = f_b_bs(state[-2 + 25], state[-2 + 29], state[-2 + 31], state[-2 + 32]); + const bitslice_t filter17_3 = f_b_bs(state[-2 + 45], state[-2 + 46], state[-2 + 48], state[-2 + 50]); + for (uchar i3 = 0; i3 < (1 << 3);) { + state[-2 + 11] = -((bool)(i3 & 0x1)); + state[-2 + 20] = -((bool)(i3 & 0x2)); + state[-2 + 37] = -((bool)(i3 & 0x4)); + i3++; + // 0xff07ffffffff + const bitslice_t filter3_1 = f_b_bs(state[-2 + 11], state[-2 + 15], state[-2 + 17], state[-2 + 18]); + const bitslice_t filter3_2 = f_b_bs(state[-2 + 20], state[-2 + 24], state[-2 + 26], state[-2 + 29]); + const bitslice_t filter3_4 = f_a_bs(state[-2 + 37], state[-2 + 46], state[-2 + 47], state[-2 + 49]); + const bitslice_t filter3 = f_c_bs(filter3_0, filter3_1, filter3_2, filter3_3, filter3_4); + const bitslice_t results3 = results2 & (filter3 ^ keystream[3]); + if (!results3) continue; + state[-2 + 51] = lfsr_bs(3); + state[-2 + 52] = lfsr_bs(4); + state[-2 + 53] = lfsr_bs(5); + state[-2 + 54] = lfsr_bs(6); + state[-2 + 55] = lfsr_bs(7); + const bitslice_t filter4_3 = f_b_bs(state[-2 + 32], state[-2 + 33], state[-2 + 35], state[-2 + 37]); + const bitslice_t filter5_0 = f_a_bs(state[-2 + 7], state[-2 + 8], state[-2 + 10], state[-2 + 11]); + const bitslice_t filter5_1 = f_b_bs(state[-2 + 13], state[-2 + 17], state[-2 + 19], state[-2 + 20]); + const bitslice_t filter6_0 = f_a_bs(state[-2 + 8], state[-2 + 9], state[-2 + 11], state[-2 + 12]); + const bitslice_t filter6_1 = f_b_bs(state[-2 + 14], state[-2 + 18], state[-2 + 20], state[-2 + 21]); + const bitslice_t filter8_0 = f_a_bs(state[-2 + 10], state[-2 + 11], state[-2 + 13], state[-2 + 14]); + const bitslice_t filter8_1 = f_b_bs(state[-2 + 16], state[-2 + 20], state[-2 + 22], state[-2 + 23]); + const bitslice_t filter9_0 = f_a_bs(state[-2 + 11], state[-2 + 12], state[-2 + 14], state[-2 + 15]); + const bitslice_t filter9_4 = f_a_bs(state[-2 + 43], state[-2 + 52], state[-2 + 53], state[-2 + 55]); + const bitslice_t filter11_2 = f_b_bs(state[-2 + 28], state[-2 + 32], state[-2 + 34], state[-2 + 37]); + const bitslice_t filter12_1 = f_b_bs(state[-2 + 20], state[-2 + 24], state[-2 + 26], state[-2 + 27]); + const bitslice_t filter14_0 = f_a_bs(state[-2 + 16], state[-2 + 17], state[-2 + 19], state[-2 + 20]); + const bitslice_t filter15_0 = f_a_bs(state[-2 + 17], state[-2 + 18], state[-2 + 20], state[-2 + 21]); + const bitslice_t filter17_0 = f_a_bs(state[-2 + 19], state[-2 + 20], state[-2 + 22], state[-2 + 23]); + for (uchar i4 = 0; i4 < (1 << 1);) { + state[-2 + 38] = -i4; + i4++; + // 0xff87ffffffff + const bitslice_t filter4_4 = f_a_bs(state[-2 + 38], state[-2 + 47], state[-2 + 48], state[-2 + 50]); + const bitslice_t filter4 = f_c_bs(filter4_0, filter4_1, filter4_2, filter4_3, filter4_4); + const bitslice_t results4 = results3 & (filter4 ^ keystream[4]); + if (!results4) continue; + state[-2 + 56] = lfsr_bs(8); + const bitslice_t filter5_3 = f_b_bs(state[-2 + 33], state[-2 + 34], state[-2 + 36], state[-2 + 38]); + const bitslice_t filter10_4 = f_a_bs(state[-2 + 44], state[-2 + 53], state[-2 + 54], state[-2 + 56]); + const bitslice_t filter12_2 = f_b_bs(state[-2 + 29], state[-2 + 33], state[-2 + 35], state[-2 + 38]); + for (uchar i5 = 0; i5 < (1 << 1);) { + state[-2 + 39] = -i5; + i5++; + // 0xffc7ffffffff + const bitslice_t filter5_4 = f_a_bs(state[-2 + 39], state[-2 + 48], state[-2 + 49], state[-2 + 51]); + const bitslice_t filter5 = f_c_bs(filter5_0, filter5_1, filter5_2, filter5_3, filter5_4); + const bitslice_t results5 = results4 & (filter5 ^ keystream[5]); + if (!results5) continue; + state[-2 + 57] = lfsr_bs(9); + const bitslice_t filter6_3 = f_b_bs(state[-2 + 34], state[-2 + 35], state[-2 + 37], state[-2 + 39]); + const bitslice_t filter11_4 = f_a_bs(state[-2 + 45], state[-2 + 54], state[-2 + 55], state[-2 + 57]); + const bitslice_t filter13_2 = f_b_bs(state[-2 + 30], state[-2 + 34], state[-2 + 36], state[-2 + 39]); + for (uchar i6 = 0; i6 < (1 << 1);) { + state[-2 + 40] = -i6; + i6++; + // 0xffe7ffffffff + const bitslice_t filter6_4 = f_a_bs(state[-2 + 40], state[-2 + 49], state[-2 + 50], state[-2 + 52]); + const bitslice_t filter6 = f_c_bs(filter6_0, filter6_1, filter6_2, filter6_3, filter6_4); + const bitslice_t results6 = results5 & (filter6 ^ keystream[6]); + if (!results6) continue; + state[-2 + 58] = lfsr_bs(10); + const bitslice_t filter7_3 = f_b_bs(state[-2 + 35], state[-2 + 36], state[-2 + 38], state[-2 + 40]); + const bitslice_t filter12_4 = f_a_bs(state[-2 + 46], state[-2 + 55], state[-2 + 56], state[-2 + 58]); + const bitslice_t filter14_2 = f_b_bs(state[-2 + 31], state[-2 + 35], state[-2 + 37], state[-2 + 40]); + const bitslice_t filter17_2 = f_b_bs(state[-2 + 34], state[-2 + 38], state[-2 + 40], state[-2 + 43]); +#pragma unroll + for (uchar i7 = 0; i7 < (1 << 1);) { + state[-2 + 41] = -i7; + i7++; + // 0xfff7ffffffff + const bitslice_t filter7_4 = f_a_bs(state[-2 + 41], state[-2 + 50], state[-2 + 51], state[-2 + 53]); + const bitslice_t filter7 = f_c_bs(filter7_0, filter7_1, filter7_2, filter7_3, filter7_4); + const bitslice_t results7 = results6 & (filter7 ^ keystream[7]); + if (!results7) continue; + state[-2 + 59] = lfsr_bs(11); + const bitslice_t filter8_3 = f_b_bs(state[-2 + 36], state[-2 + 37], state[-2 + 39], state[-2 + 41]); + const bitslice_t filter10_3 = f_b_bs(state[-2 + 38], state[-2 + 39], state[-2 + 41], state[-2 + 43]); + const bitslice_t filter10 = f_c_bs(filter10_0, filter10_1, filter10_2, filter10_3, filter10_4); + const bitslice_t filter12_3 = f_b_bs(state[-2 + 40], state[-2 + 41], state[-2 + 43], state[-2 + 45]); + const bitslice_t filter12 = f_c_bs(filter12_0, filter12_1, filter12_2, filter12_3, filter12_4); + const bitslice_t filter13_4 = f_a_bs(state[-2 + 47], state[-2 + 56], state[-2 + 57], state[-2 + 59]); + const bitslice_t filter15_2 = f_b_bs(state[-2 + 32], state[-2 + 36], state[-2 + 38], state[-2 + 41]); +#pragma unroll + for (uchar i8 = 0; i8 < (1 << 1);) { + state[-2 + 42] = -i8; + i8++; + // 0xffffffffffff + const bitslice_t filter8_4 = f_a_bs(state[-2 + 42], state[-2 + 51], state[-2 + 52], state[-2 + 54]); + const bitslice_t filter8 = f_c_bs(filter8_0, filter8_1, filter8_2, filter8_3, filter8_4); + bitslice_t results8 = results7 & (filter8 ^ keystream[8]); + if (!results8) continue; + const bitslice_t filter9_3 = f_b_bs(state[-2 + 37], state[-2 + 38], state[-2 + 40], state[-2 + 42]); + const bitslice_t filter9 = f_c_bs(filter9_0, filter9_1, filter9_2, filter9_3, filter9_4); + results8 &= (filter9 ^ keystream[9]); + if (!results8) continue; + results8 &= (filter10 ^ keystream[10]); + if (!results8) continue; + const bitslice_t filter11_3 = f_b_bs(state[-2 + 39], state[-2 + 40], state[-2 + 42], state[-2 + 44]); + const bitslice_t filter11 = f_c_bs(filter11_0, filter11_1, filter11_2, filter11_3, filter11_4); + results8 &= (filter11 ^ keystream[11]); + if (!results8) continue; + results8 &= (filter12 ^ keystream[12]); + if (!results8) continue; + const bitslice_t filter13_3 = f_b_bs(state[-2 + 41], state[-2 + 42], state[-2 + 44], state[-2 + 46]); + const bitslice_t filter13 = f_c_bs(filter13_0, filter13_1, filter13_2, filter13_3, filter13_4); + results8 &= (filter13 ^ keystream[13]); + if (!results8) continue; + state[-2 + 60] = lfsr_bs(12); + const bitslice_t filter14_3 = f_b_bs(state[-2 + 42], state[-2 + 43], state[-2 + 45], state[-2 + 47]); + const bitslice_t filter14_4 = f_a_bs(state[-2 + 48], state[-2 + 57], state[-2 + 58], state[-2 + 60]); + const bitslice_t filter14 = f_c_bs(filter14_0, filter14_1, filter14_2, filter14_3, filter14_4); + results8 &= (filter14 ^ keystream[14]); + if (!results8) continue; + state[-2 + 61] = lfsr_bs(13); + const bitslice_t filter15_4 = f_a_bs(state[-2 + 49], state[-2 + 58], state[-2 + 59], state[-2 + 61]); + const bitslice_t filter15 = f_c_bs(filter15_0, filter15_1, filter15_2, filter15_3, filter15_4); + results8 &= (filter15 ^ keystream[15]); + if (!results8) continue; + state[-2 + 62] = lfsr_bs(14); + const bitslice_t filter16_2 = f_b_bs(state[-2 + 33], state[-2 + 37], state[-2 + 39], state[-2 + 42]); + const bitslice_t filter16_4 = f_a_bs(state[-2 + 50], state[-2 + 59], state[-2 + 60], state[-2 + 62]); + const bitslice_t filter16 = f_c_bs(filter16_0, filter16_1, filter16_2, filter16_3, filter16_4); + results8 &= (filter16 ^ keystream[16]); + if (!results8) continue; + state[-2 + 63] = lfsr_bs(15); + const bitslice_t filter17_4 = f_a_bs(state[-2 + 51], state[-2 + 60], state[-2 + 61], state[-2 + 63]); + const bitslice_t filter17 = f_c_bs(filter17_0, filter17_1, filter17_2, filter17_3, filter17_4); + results8 &= (filter17 ^ keystream[17]); + if (!results8) continue; + state[-2 + 64] = lfsr_bs(16); + const bitslice_t filter18_0 = f_a_bs(state[-2 + 20], state[-2 + 21], state[-2 + 23], state[-2 + 24]); + const bitslice_t filter18_1 = f_b_bs(state[-2 + 26], state[-2 + 30], state[-2 + 32], state[-2 + 33]); + const bitslice_t filter18_2 = f_b_bs(state[-2 + 35], state[-2 + 39], state[-2 + 41], state[-2 + 44]); + const bitslice_t filter18_3 = f_b_bs(state[-2 + 46], state[-2 + 47], state[-2 + 49], state[-2 + 51]); + const bitslice_t filter18_4 = f_a_bs(state[-2 + 52], state[-2 + 61], state[-2 + 62], state[-2 + 64]); + const bitslice_t filter18 = f_c_bs(filter18_0, filter18_1, filter18_2, filter18_3, filter18_4); + results8 &= (filter18 ^ keystream[18]); + if (!results8) continue; + state[-2 + 65] = lfsr_bs(17); + const bitslice_t filter19_0 = f_a_bs(state[-2 + 21], state[-2 + 22], state[-2 + 24], state[-2 + 25]); + const bitslice_t filter19_1 = f_b_bs(state[-2 + 27], state[-2 + 31], state[-2 + 33], state[-2 + 34]); + const bitslice_t filter19_2 = f_b_bs(state[-2 + 36], state[-2 + 40], state[-2 + 42], state[-2 + 45]); + const bitslice_t filter19_3 = f_b_bs(state[-2 + 47], state[-2 + 48], state[-2 + 50], state[-2 + 52]); + const bitslice_t filter19_4 = f_a_bs(state[-2 + 53], state[-2 + 62], state[-2 + 63], state[-2 + 65]); + const bitslice_t filter19 = f_c_bs(filter19_0, filter19_1, filter19_2, filter19_3, filter19_4); + results8 &= (filter19 ^ keystream[19]); + if (!results8) continue; + state[-2 + 66] = lfsr_bs(18); + const bitslice_t filter20_0 = f_a_bs(state[-2 + 22], state[-2 + 23], state[-2 + 25], state[-2 + 26]); + const bitslice_t filter20_1 = f_b_bs(state[-2 + 28], state[-2 + 32], state[-2 + 34], state[-2 + 35]); + const bitslice_t filter20_2 = f_b_bs(state[-2 + 37], state[-2 + 41], state[-2 + 43], state[-2 + 46]); + const bitslice_t filter20_3 = f_b_bs(state[-2 + 48], state[-2 + 49], state[-2 + 51], state[-2 + 53]); + const bitslice_t filter20_4 = f_a_bs(state[-2 + 54], state[-2 + 63], state[-2 + 64], state[-2 + 66]); + const bitslice_t filter20 = f_c_bs(filter20_0, filter20_1, filter20_2, filter20_3, filter20_4); + results8 &= (filter20 ^ keystream[20]); + if (!results8) continue; + state[-2 + 67] = lfsr_bs(19); + const bitslice_t filter21_0 = f_a_bs(state[-2 + 23], state[-2 + 24], state[-2 + 26], state[-2 + 27]); + const bitslice_t filter21_1 = f_b_bs(state[-2 + 29], state[-2 + 33], state[-2 + 35], state[-2 + 36]); + const bitslice_t filter21_2 = f_b_bs(state[-2 + 38], state[-2 + 42], state[-2 + 44], state[-2 + 47]); + const bitslice_t filter21_3 = f_b_bs(state[-2 + 49], state[-2 + 50], state[-2 + 52], state[-2 + 54]); + const bitslice_t filter21_4 = f_a_bs(state[-2 + 55], state[-2 + 64], state[-2 + 65], state[-2 + 67]); + const bitslice_t filter21 = f_c_bs(filter21_0, filter21_1, filter21_2, filter21_3, filter21_4); + results8 &= (filter21 ^ keystream[21]); + if (!results8) continue; + state[-2 + 68] = lfsr_bs(20); + const bitslice_t filter22_0 = f_a_bs(state[-2 + 24], state[-2 + 25], state[-2 + 27], state[-2 + 28]); + const bitslice_t filter22_1 = f_b_bs(state[-2 + 30], state[-2 + 34], state[-2 + 36], state[-2 + 37]); + const bitslice_t filter22_2 = f_b_bs(state[-2 + 39], state[-2 + 43], state[-2 + 45], state[-2 + 48]); + const bitslice_t filter22_3 = f_b_bs(state[-2 + 50], state[-2 + 51], state[-2 + 53], state[-2 + 55]); + const bitslice_t filter22_4 = f_a_bs(state[-2 + 56], state[-2 + 65], state[-2 + 66], state[-2 + 68]); + const bitslice_t filter22 = f_c_bs(filter22_0, filter22_1, filter22_2, filter22_3, filter22_4); + results8 &= (filter22 ^ keystream[22]); + if (!results8) continue; + state[-2 + 69] = lfsr_bs(21); + const bitslice_t filter23_0 = f_a_bs(state[-2 + 25], state[-2 + 26], state[-2 + 28], state[-2 + 29]); + const bitslice_t filter23_1 = f_b_bs(state[-2 + 31], state[-2 + 35], state[-2 + 37], state[-2 + 38]); + const bitslice_t filter23_2 = f_b_bs(state[-2 + 40], state[-2 + 44], state[-2 + 46], state[-2 + 49]); + const bitslice_t filter23_3 = f_b_bs(state[-2 + 51], state[-2 + 52], state[-2 + 54], state[-2 + 56]); + const bitslice_t filter23_4 = f_a_bs(state[-2 + 57], state[-2 + 66], state[-2 + 67], state[-2 + 69]); + const bitslice_t filter23 = f_c_bs(filter23_0, filter23_1, filter23_2, filter23_3, filter23_4); + results8 &= (filter23 ^ keystream[23]); + if (!results8) continue; + state[-2 + 70] = lfsr_bs(22); + const bitslice_t filter24_0 = f_a_bs(state[-2 + 26], state[-2 + 27], state[-2 + 29], state[-2 + 30]); + const bitslice_t filter24_1 = f_b_bs(state[-2 + 32], state[-2 + 36], state[-2 + 38], state[-2 + 39]); + const bitslice_t filter24_2 = f_b_bs(state[-2 + 41], state[-2 + 45], state[-2 + 47], state[-2 + 50]); + const bitslice_t filter24_3 = f_b_bs(state[-2 + 52], state[-2 + 53], state[-2 + 55], state[-2 + 57]); + const bitslice_t filter24_4 = f_a_bs(state[-2 + 58], state[-2 + 67], state[-2 + 68], state[-2 + 70]); + const bitslice_t filter24 = f_c_bs(filter24_0, filter24_1, filter24_2, filter24_3, filter24_4); + results8 &= (filter24 ^ keystream[24]); + if (!results8) continue; + state[-2 + 71] = lfsr_bs(23); + const bitslice_t filter25_0 = f_a_bs(state[-2 + 27], state[-2 + 28], state[-2 + 30], state[-2 + 31]); + const bitslice_t filter25_1 = f_b_bs(state[-2 + 33], state[-2 + 37], state[-2 + 39], state[-2 + 40]); + const bitslice_t filter25_2 = f_b_bs(state[-2 + 42], state[-2 + 46], state[-2 + 48], state[-2 + 51]); + const bitslice_t filter25_3 = f_b_bs(state[-2 + 53], state[-2 + 54], state[-2 + 56], state[-2 + 58]); + const bitslice_t filter25_4 = f_a_bs(state[-2 + 59], state[-2 + 68], state[-2 + 69], state[-2 + 71]); + const bitslice_t filter25 = f_c_bs(filter25_0, filter25_1, filter25_2, filter25_3, filter25_4); + results8 &= (filter25 ^ keystream[25]); + if (!results8) continue; + state[-2 + 72] = lfsr_bs(24); + const bitslice_t filter26_0 = f_a_bs(state[-2 + 28], state[-2 + 29], state[-2 + 31], state[-2 + 32]); + const bitslice_t filter26_1 = f_b_bs(state[-2 + 34], state[-2 + 38], state[-2 + 40], state[-2 + 41]); + const bitslice_t filter26_2 = f_b_bs(state[-2 + 43], state[-2 + 47], state[-2 + 49], state[-2 + 52]); + const bitslice_t filter26_3 = f_b_bs(state[-2 + 54], state[-2 + 55], state[-2 + 57], state[-2 + 59]); + const bitslice_t filter26_4 = f_a_bs(state[-2 + 60], state[-2 + 69], state[-2 + 70], state[-2 + 72]); + const bitslice_t filter26 = f_c_bs(filter26_0, filter26_1, filter26_2, filter26_3, filter26_4); + results8 &= (filter26 ^ keystream[26]); + if (!results8) continue; + state[-2 + 73] = lfsr_bs(25); + const bitslice_t filter27_0 = f_a_bs(state[-2 + 29], state[-2 + 30], state[-2 + 32], state[-2 + 33]); + const bitslice_t filter27_1 = f_b_bs(state[-2 + 35], state[-2 + 39], state[-2 + 41], state[-2 + 42]); + const bitslice_t filter27_2 = f_b_bs(state[-2 + 44], state[-2 + 48], state[-2 + 50], state[-2 + 53]); + const bitslice_t filter27_3 = f_b_bs(state[-2 + 55], state[-2 + 56], state[-2 + 58], state[-2 + 60]); + const bitslice_t filter27_4 = f_a_bs(state[-2 + 61], state[-2 + 70], state[-2 + 71], state[-2 + 73]); + const bitslice_t filter27 = f_c_bs(filter27_0, filter27_1, filter27_2, filter27_3, filter27_4); + results8 &= (filter27 ^ keystream[27]); + if (!results8) continue; + state[-2 + 74] = lfsr_bs(26); + const bitslice_t filter28_0 = f_a_bs(state[-2 + 30], state[-2 + 31], state[-2 + 33], state[-2 + 34]); + const bitslice_t filter28_1 = f_b_bs(state[-2 + 36], state[-2 + 40], state[-2 + 42], state[-2 + 43]); + const bitslice_t filter28_2 = f_b_bs(state[-2 + 45], state[-2 + 49], state[-2 + 51], state[-2 + 54]); + const bitslice_t filter28_3 = f_b_bs(state[-2 + 56], state[-2 + 57], state[-2 + 59], state[-2 + 61]); + const bitslice_t filter28_4 = f_a_bs(state[-2 + 62], state[-2 + 71], state[-2 + 72], state[-2 + 74]); + const bitslice_t filter28 = f_c_bs(filter28_0, filter28_1, filter28_2, filter28_3, filter28_4); + results8 &= (filter28 ^ keystream[28]); + if (!results8) continue; + state[-2 + 75] = lfsr_bs(27); + const bitslice_t filter29_0 = f_a_bs(state[-2 + 31], state[-2 + 32], state[-2 + 34], state[-2 + 35]); + const bitslice_t filter29_1 = f_b_bs(state[-2 + 37], state[-2 + 41], state[-2 + 43], state[-2 + 44]); + const bitslice_t filter29_2 = f_b_bs(state[-2 + 46], state[-2 + 50], state[-2 + 52], state[-2 + 55]); + const bitslice_t filter29_3 = f_b_bs(state[-2 + 57], state[-2 + 58], state[-2 + 60], state[-2 + 62]); + const bitslice_t filter29_4 = f_a_bs(state[-2 + 63], state[-2 + 72], state[-2 + 73], state[-2 + 75]); + const bitslice_t filter29 = f_c_bs(filter29_0, filter29_1, filter29_2, filter29_3, filter29_4); + results8 &= (filter29 ^ keystream[29]); + if (!results8) continue; + state[-2 + 76] = lfsr_bs(28); + const bitslice_t filter30_0 = f_a_bs(state[-2 + 32], state[-2 + 33], state[-2 + 35], state[-2 + 36]); + const bitslice_t filter30_1 = f_b_bs(state[-2 + 38], state[-2 + 42], state[-2 + 44], state[-2 + 45]); + const bitslice_t filter30_2 = f_b_bs(state[-2 + 47], state[-2 + 51], state[-2 + 53], state[-2 + 56]); + const bitslice_t filter30_3 = f_b_bs(state[-2 + 58], state[-2 + 59], state[-2 + 61], state[-2 + 63]); + const bitslice_t filter30_4 = f_a_bs(state[-2 + 64], state[-2 + 73], state[-2 + 74], state[-2 + 76]); + const bitslice_t filter30 = f_c_bs(filter30_0, filter30_1, filter30_2, filter30_3, filter30_4); + results8 &= (filter30 ^ keystream[30]); + if (!results8) continue; + state[-2 + 77] = lfsr_bs(29); + const bitslice_t filter31_0 = f_a_bs(state[-2 + 33], state[-2 + 34], state[-2 + 36], state[-2 + 37]); + const bitslice_t filter31_1 = f_b_bs(state[-2 + 39], state[-2 + 43], state[-2 + 45], state[-2 + 46]); + const bitslice_t filter31_2 = f_b_bs(state[-2 + 48], state[-2 + 52], state[-2 + 54], state[-2 + 57]); + const bitslice_t filter31_3 = f_b_bs(state[-2 + 59], state[-2 + 60], state[-2 + 62], state[-2 + 64]); + const bitslice_t filter31_4 = f_a_bs(state[-2 + 65], state[-2 + 74], state[-2 + 75], state[-2 + 77]); + const bitslice_t filter31 = f_c_bs(filter31_0, filter31_1, filter31_2, filter31_3, filter31_4); + results8 &= (filter31 ^ keystream[31]); + if (!results8) continue; + uchar match_index = 0; + // Save results + while (results8 && (match_index < MAX_BITSLICES)) { + uchar shift = clz(results8) + 1; + match_index += shift; + // take the state from layer 2 so we can recover the lowest 2 bits on the host by inverting the LFSR + matches[atomic_inc(matches_found)] = unbitslice(&state[-2 + 2], MAX_BITSLICES - match_index, 48); + results8 <<= shift; + } + } // 8 + } // 7 + } // 6 + } // 5 + } // 4 + } // 3 + } // 2 +} // 1 + diff --git a/tools/hitag2crack/crack5gpu/rfidler.h b/tools/hitag2crack/crack5gpu/rfidler.h new file mode 100644 index 000000000..933547e6b --- /dev/null +++ b/tools/hitag2crack/crack5gpu/rfidler.h @@ -0,0 +1,412 @@ +/*************************************************************************** + * A copy of the GNU GPL is appended to this file. * + * * + * This licence is based on the nmap licence, and we express our gratitude * + * for the work that went into producing it. There is no other connection * + * between RFIDler and nmap either expressed or implied. * + * * + ********************** IMPORTANT RFIDler LICENSE TERMS ******************** + * * + * * + * All references to RFIDler herein imply all it's derivatives, namely: * + * * + * o RFIDler-LF Standard * + * o RFIDler-LF Lite * + * o RFIDler-LF Nekkid * + * * + * * + * RFIDler is (C) 2013-2015 Aperture Labs Ltd. * + * * + * This program is free software; you may redistribute and/or modify it * + * under the terms of the GNU General Public License as published by the * + * Free Software Foundation; Version 2 ("GPL"), BUT ONLY WITH ALL OF THE * + * CLARIFICATIONS AND EXCEPTIONS DESCRIBED HEREIN. This guarantees your * + * right to use, modify, and redistribute this software under certain * + * conditions. If you wish to embed RFIDler technology into proprietary * + * software or hardware, we sell alternative licenses * + * (contact sales@aperturelabs.com). * + * * + * Note that the GPL places important restrictions on "derivative works", * + * yet it does not provide a detailed definition of that term. To avoid * + * misunderstandings, we interpret that term as broadly as copyright law * + * allows. For example, we consider an application to constitute a * + * derivative work for the purpose of this license if it does any of the * + * following with any software or content covered by this license * + * ("Covered Software"): * + * * + * o Integrates source code from Covered Software. * + * * + * o Is designed specifically to execute Covered Software and parse the * + * results (as opposed to typical shell or execution-menu apps, which will * + * execute anything you tell them to). * + * * + * o Includes Covered Software in a proprietary executable installer. The * + * installers produced by InstallShield are an example of this. Including * + * RFIDler with other software in compressed or archival form does not * + * trigger this provision, provided appropriate open source decompression * + * or de-archiving software is widely available for no charge. For the * + * purposes of this license, an installer is considered to include Covered * + * Software even if it actually retrieves a copy of Covered Software from * + * another source during runtime (such as by downloading it from the * + * Internet). * + * * + * o Links (statically or dynamically) to a library which does any of the * + * above. * + * * + * o Executes a helper program, module, or script to do any of the above. * + * * + * This list is not exclusive, but is meant to clarify our interpretation * + * of derived works with some common examples. Other people may interpret * + * the plain GPL differently, so we consider this a special exception to * + * the GPL that we apply to Covered Software. Works which meet any of * + * these conditions must conform to all of the terms of this license, * + * particularly including the GPL Section 3 requirements of providing * + * source code and allowing free redistribution of the work as a whole. * + * * + * As another special exception to the GPL terms, Aperture Labs Ltd. grants* + * permission to link the code of this program with any version of the * + * OpenSSL library which is distributed under a license identical to that * + * listed in the included docs/licenses/OpenSSL.txt file, and distribute * + * linked combinations including the two. * + * * + * Any redistribution of Covered Software, including any derived works, * + * must obey and carry forward all of the terms of this license, including * + * obeying all GPL rules and restrictions. For example, source code of * + * the whole work must be provided and free redistribution must be * + * allowed. All GPL references to "this License", are to be treated as * + * including the terms and conditions of this license text as well. * + * * + * Because this license imposes special exceptions to the GPL, Covered * + * Work may not be combined (even as part of a larger work) with plain GPL * + * software. The terms, conditions, and exceptions of this license must * + * be included as well. This license is incompatible with some other open * + * source licenses as well. In some cases we can relicense portions of * + * RFIDler or grant special permissions to use it in other open source * + * software. Please contact sales@aperturelabs.com with any such requests.* + * Similarly, we don't incorporate incompatible open source software into * + * Covered Software without special permission from the copyright holders. * + * * + * If you have any questions about the licensing restrictions on using * + * RFIDler in other works, are happy to help. As mentioned above, we also * + * offer alternative license to integrate RFIDler into proprietary * + * applications and appliances. These contracts have been sold to dozens * + * of software vendors, and generally include a perpetual license as well * + * as providing for priority support and updates. They also fund the * + * continued development of RFIDler. Please email sales@aperturelabs.com * + * for further information. * + * If you have received a written license agreement or contract for * + * Covered Software stating terms other than these, you may choose to use * + * and redistribute Covered Software under those terms instead of these. * + * * + * Source is provided to this software because we believe users have a * + * right to know exactly what a program is going to do before they run it. * + * This also allows you to audit the software for security holes (none * + * have been found so far). * + * * + * Source code also allows you to port RFIDler to new platforms, fix bugs, * + * and add new features. You are highly encouraged to send your changes * + * to the RFIDler mailing list for possible incorporation into the * + * main distribution. By sending these changes to Aperture Labs Ltd. or * + * one of the Aperture Labs Ltd. development mailing lists, or checking * + * them into the RFIDler source code repository, it is understood (unless * + * you specify otherwise) that you are offering the RFIDler Project * + * (Aperture Labs Ltd.) the unlimited, non-exclusive right to reuse, * + * modify, and relicense the code. RFIDler will always be available Open * + * Source, but this is important because the inability to relicense code * + * has caused devastating problems for other Free Software projects (such * + * as KDE and NASM). We also occasionally relicense the code to third * + * parties as discussed above. If you wish to specify special license * + * conditions of your contributions, just say so when you send them. * + * * + * This program 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 RFIDler * + * license file for more details (it's in a COPYING file included with * + * RFIDler, and also available from * + * https://github.com/ApertureLabsLtd/RFIDler/COPYING * + * * + ***************************************************************************/ + +// Author: Adam Laurie + +#include +#include + +// BCD hardware revision for usb descriptor (usb_descriptors.c) +#define RFIDLER_HW_VERSION 0x020 + +// max sizes in BITS +#define MAXBLOCKSIZE 512 +#define MAXTAGSIZE 4096 +#define MAXUID 512 + +#define TMP_LARGE_BUFF_LEN 2048 +#define TMP_SMALL_BUFF_LEN 256 +#define ANALOGUE_BUFF_LEN 8192 + +#define COMMS_BUFFER_SIZE 128 + +#define DETECT_BUFFER_SIZE 512 + +#define SAMPLEMASK ~(BIT_1 | BIT_0) // mask to remove two bottom bits from analogue sample - we will then use those for reader & bit period + +// globals + +extern BOOL WiegandOutput; // Output wiegand data whenenver UID is read +extern BYTE *EMU_Reset_Data; // Pointer to full array of bits as bytes, stored as 0x00/0x01, '*' terminated +extern BYTE *EMU_Data; // Pointer to current location in EMU_Reset_Data +extern BYTE EMU_ThisBit; // The next data bit to transmit +extern BYTE EMU_SubCarrier_T0; // Number of Frame Clocks for sub-carrier '0' +extern BYTE EMU_SubCarrier_T1; // Number of Frame Clocks for sub-carrier '1' +extern unsigned int EMU_Repeat; // Number of times to transmit full data set +extern BOOL EMU_Background; // Emulate in the background until told to stop +extern unsigned int EMU_DataBitRate; // Number of Frame Clocks per bit +extern BYTE TmpBits[TMP_LARGE_BUFF_LEN]; // Shared scratchpad +extern BYTE ReaderPeriod; // Flag for sample display +extern unsigned char Comms_In_Buffer[COMMS_BUFFER_SIZE]; // USB/Serial buffer +extern BYTE Interface; // user interface - CLI or API +extern BYTE CommsChannel; // user comms channel - USB or UART +extern BOOL FakeRead; // flag for analogue sampler to signal it wants access to buffers during read +extern BOOL PWD_Mode; // is this tag password protected? +extern BYTE Password[9]; // 32 bits as HEX string set with LOGIN +extern unsigned int Led_Count; // LED status counter, also used for entropy +extern unsigned long Reader_Bit_Count; // Reader ISR bit counter +extern char Previous; // Reader ISR previous bit type + +// RWD (read/write device) coil state +extern BYTE RWD_State; // current state of RWD coil +extern unsigned int RWD_Fc; // field clock in uS +extern unsigned int RWD_Gap_Period; // length of command gaps in OC5 ticks +extern unsigned int RWD_Zero_Period; // length of '0' in OC5 ticks +extern unsigned int RWD_One_Period; // length of '1' in OC5 ticks +extern unsigned int RWD_Sleep_Period; // length of initial sleep to reset tag in OC5 ticks +extern unsigned int RWD_Wake_Period; // length required for tag to restart in OC5 ticks +extern unsigned int RWD_Wait_Switch_TX_RX; // length to wait when switching from TX to RX in OC5 ticks +extern unsigned int RWD_Wait_Switch_RX_TX; // length to wait when switching from RX to TX in OC5 ticks +extern unsigned int RWD_Post_Wait; // low level ISR wait period in OC5 ticks +extern unsigned int RWD_OC5_config; // Output Compare Module settings +extern unsigned int RWD_OC5_r; // Output Compare Module primary compare value +extern unsigned int RWD_OC5_rs; // Output Compare Module secondary compare value +extern BYTE RWD_Command_Buff[TMP_SMALL_BUFF_LEN]; // Command buffer, array of bits as bytes, stored as 0x00/0x01, '*' terminated +extern BYTE *RWD_Command_ThisBit; // Current command bit +extern BOOL Reader_ISR_State; // current state of reader ISR + +// NVM variables +// timings etc. that want to survive a reboot should go here +typedef struct { + BYTE Name[7]; // will be set to "RFIDler" so we can test for new device + BYTE AutoRun[128]; // optional command to run at startup + unsigned char TagType; + unsigned int PSK_Quality; + unsigned int Timeout; + unsigned int Wiegand_Pulse; + unsigned int Wiegand_Gap; + BOOL Wiegand_IdleState; + unsigned int FrameClock; + unsigned char Modulation; + unsigned int DataRate; + unsigned int DataRateSub0; + unsigned int DataRateSub1; + unsigned int DataBits; + unsigned int DataBlocks; + unsigned int BlockSize; + unsigned char SyncBits; + BYTE Sync[4]; + BOOL BiPhase; + BOOL Invert; + BOOL Manchester; + BOOL HalfDuplex; + unsigned int Repeat; + unsigned int PotLow; + unsigned int PotHigh; + unsigned int RWD_Gap_Period; + unsigned int RWD_Zero_Period; + unsigned int RWD_One_Period; + unsigned int RWD_Sleep_Period; + unsigned int RWD_Wake_Period; + unsigned int RWD_Wait_Switch_TX_RX; + unsigned int RWD_Wait_Switch_RX_TX; +} StoredConfig; + +// somewhere to store TAG data. this will be interpreted according to the TAG +// type. +typedef struct { + BYTE TagType; // raw tag type + BYTE EmulatedTagType; // tag type this tag is configured to emulate + BYTE UID[MAXUID + 1]; // Null-terminated HEX string + BYTE Data[MAXTAGSIZE]; // raw data + unsigned char DataBlocks; // number of blocks in Data field + unsigned int BlockSize; // blocksize in bits +} VirtualTag; + +extern StoredConfig RFIDlerConfig; +extern VirtualTag RFIDlerVTag; +extern BYTE TmpBuff[NVM_PAGE_SIZE]; +extern BYTE DataBuff[ANALOGUE_BUFF_LEN]; +extern unsigned int DataBuffCount; +extern const BYTE *ModulationSchemes[]; +extern const BYTE *OnOff[]; +extern const BYTE *HighLow[]; +extern const BYTE *TagTypes[]; + +// globals for ISRs +extern BYTE EmulationMode; +extern unsigned long HW_Bits; +extern BYTE HW_Skip_Bits; +extern unsigned int PSK_Min_Pulse; +extern BOOL PSK_Read_Error; +extern BOOL Manchester_Error; +extern BOOL SnifferMode; +extern unsigned int Clock_Tick_Counter; +extern BOOL Clock_Tick_Counter_Reset; + +// smart card lib +#define MAX_ATR_LEN (BYTE)33 +extern BYTE scCardATR[MAX_ATR_LEN]; +extern BYTE scATRLength; + +// RTC +extern rtccTime RTC_time; // time structure +extern rtccDate RTC_date; // date structure + +// digital pots +#define POTLOW_DEFAULT 100 +#define POTHIGH_DEFAULT 150 +#define DC_OFFSET 60 // analogue circuit DC offset (as close as we can get without using 2 LSB) +#define VOLTS_TO_POT 0.019607843F + +// RWD/clock states +#define RWD_STATE_INACTIVE 0 // RWD not in use +#define RWD_STATE_GO_TO_SLEEP 1 // RWD coil shutdown request +#define RWD_STATE_SLEEPING 2 // RWD coil shutdown for sleep period +#define RWD_STATE_WAKING 3 // RWD active for pre-determined period after reset +#define RWD_STATE_START_SEND 4 // RWD starting send of data +#define RWD_STATE_SENDING_GAP 5 // RWD sending a gap +#define RWD_STATE_SENDING_BIT 6 // RWD sending a data bit +#define RWD_STATE_POST_WAIT 7 // RWD finished sending data, now in forced wait period +#define RWD_STATE_ACTIVE 8 // RWD finished, now just clocking a carrier + +// reader ISR states +#define READER_STOPPED 0 // reader not in use +#define READER_IDLING 1 // reader ISR running to preserve timing, but not reading +#define READER_RUNNING 2 // reader reading bits + + +// user interface types +#define INTERFACE_API 0 +#define INTERFACE_CLI 1 + +// comms channel +#define COMMS_NONE 0 +#define COMMS_USB 1 +#define COMMS_UART 2 + +#define MAX_HISTORY 2 // disable most of history for now - memory issue + +// tag write retries +#define TAG_WRITE_RETRY 5 + +// modulation modes - uppdate ModulationSchemes[] in tags.c if you change this +#define MOD_MODE_NONE 0 +#define MOD_MODE_ASK_OOK 1 +#define MOD_MODE_FSK1 2 +#define MOD_MODE_FSK2 3 +#define MOD_MODE_PSK1 4 +#define MOD_MODE_PSK2 5 +#define MOD_MODE_PSK3 6 + +// TAG types - update TagTypes[] in tags.c if you add to this list +#define TAG_TYPE_NONE 0 +#define TAG_TYPE_ASK_RAW 1 +#define TAG_TYPE_FSK1_RAW 2 +#define TAG_TYPE_FSK2_RAW 3 +#define TAG_TYPE_PSK1_RAW 4 +#define TAG_TYPE_PSK2_RAW 5 +#define TAG_TYPE_PSK3_RAW 6 +#define TAG_TYPE_HITAG1 7 +#define TAG_TYPE_HITAG2 8 +#define TAG_TYPE_EM4X02 9 +#define TAG_TYPE_Q5 10 +#define TAG_TYPE_HID_26 11 +#define TAG_TYPE_INDALA_64 12 +#define TAG_TYPE_INDALA_224 13 +#define TAG_TYPE_UNIQUE 14 +#define TAG_TYPE_FDXB 15 +#define TAG_TYPE_T55X7 16 // same as Q5 but different timings and no modulation-defeat +#define TAG_TYPE_AWID_26 17 +#define TAG_TYPE_EM4X05 18 +#define TAG_TYPE_TAMAGOTCHI 19 +#define TAG_TYPE_HDX 20 // same underlying data as FDX-B, but different modulation & telegram + +// various + +#define BINARY 0 +#define HEX 1 + +#define NO_ADDRESS -1 + +#define ACK TRUE +#define NO_ACK FALSE + +#define BLOCK TRUE +#define NO_BLOCK FALSE + +#define DATA TRUE +#define NO_DATA FALSE + +#define DEBUG_PIN_ON HIGH +#define DEBUG_PIN_OFF LOW + +#define FAST FALSE +#define SLOW TRUE + +#define NO_TRIGGER 0 + +#define LOCK TRUE +#define NO_LOCK FALSE + +#define NFC_MODE TRUE +#define NO_NFC_MODE FALSE + +#define ONESHOT_READ TRUE +#define NO_ONESHOT_READ FALSE + +#define RESET TRUE +#define NO_RESET FALSE + +#define SHUTDOWN_CLOCK TRUE +#define NO_SHUTDOWN_CLOCK FALSE + +#define SYNC TRUE +#define NO_SYNC FALSE + +#define VERIFY TRUE +#define NO_VERIFY FALSE + +#define VOLATILE FALSE +#define NON_VOLATILE TRUE + +#define NEWLINE TRUE +#define NO_NEWLINE FALSE + +#define WAIT TRUE +#define NO_WAIT FALSE + +#define WIPER_HIGH 0 +#define WIPER_LOW 1 + +// conversion for time to ticks +#define US_TO_TICKS 1000000L +#define US_OVER_10_TO_TICKS 10000000L +#define US_OVER_100_TO_TICKS 100000000L +// we can't get down to this level on pic, but we want to standardise on timings, so for now we fudge it +#define CONVERT_TO_TICKS(x) ((x / 10) * (GetSystemClock() / US_OVER_10_TO_TICKS)) +#define CONVERT_TICKS_TO_US(x) (x / (GetSystemClock() / US_TO_TICKS)) +#define TIMER5_PRESCALER 16 +#define MAX_TIMER5_TICKS (65535 * TIMER5_PRESCALER) + +// other conversions + +// bits to hex digits +#define HEXDIGITS(x) (x / 4) +#define HEXTOBITS(x) (x * 4) diff --git a/tools/hitag2crack/crack5gpu/util.h b/tools/hitag2crack/crack5gpu/util.h new file mode 100644 index 000000000..c2399c37c --- /dev/null +++ b/tools/hitag2crack/crack5gpu/util.h @@ -0,0 +1,147 @@ +/*************************************************************************** + * A copy of the GNU GPL is appended to this file. * + * * + * This licence is based on the nmap licence, and we express our gratitude * + * for the work that went into producing it. There is no other connection * + * between RFIDler and nmap either expressed or implied. * + * * + ********************** IMPORTANT RFIDler LICENSE TERMS ******************** + * * + * * + * All references to RFIDler herein imply all it's derivatives, namely: * + * * + * o RFIDler-LF Standard * + * o RFIDler-LF Lite * + * o RFIDler-LF Nekkid * + * * + * * + * RFIDler is (C) 2013-2015 Aperture Labs Ltd. * + * * + * This program is free software; you may redistribute and/or modify it * + * under the terms of the GNU General Public License as published by the * + * Free Software Foundation; Version 2 ("GPL"), BUT ONLY WITH ALL OF THE * + * CLARIFICATIONS AND EXCEPTIONS DESCRIBED HEREIN. This guarantees your * + * right to use, modify, and redistribute this software under certain * + * conditions. If you wish to embed RFIDler technology into proprietary * + * software or hardware, we sell alternative licenses * + * (contact sales@aperturelabs.com). * + * * + * Note that the GPL places important restrictions on "derivative works", * + * yet it does not provide a detailed definition of that term. To avoid * + * misunderstandings, we interpret that term as broadly as copyright law * + * allows. For example, we consider an application to constitute a * + * derivative work for the purpose of this license if it does any of the * + * following with any software or content covered by this license * + * ("Covered Software"): * + * * + * o Integrates source code from Covered Software. * + * * + * o Is designed specifically to execute Covered Software and parse the * + * results (as opposed to typical shell or execution-menu apps, which will * + * execute anything you tell them to). * + * * + * o Includes Covered Software in a proprietary executable installer. The * + * installers produced by InstallShield are an example of this. Including * + * RFIDler with other software in compressed or archival form does not * + * trigger this provision, provided appropriate open source decompression * + * or de-archiving software is widely available for no charge. For the * + * purposes of this license, an installer is considered to include Covered * + * Software even if it actually retrieves a copy of Covered Software from * + * another source during runtime (such as by downloading it from the * + * Internet). * + * * + * o Links (statically or dynamically) to a library which does any of the * + * above. * + * * + * o Executes a helper program, module, or script to do any of the above. * + * * + * This list is not exclusive, but is meant to clarify our interpretation * + * of derived works with some common examples. Other people may interpret * + * the plain GPL differently, so we consider this a special exception to * + * the GPL that we apply to Covered Software. Works which meet any of * + * these conditions must conform to all of the terms of this license, * + * particularly including the GPL Section 3 requirements of providing * + * source code and allowing free redistribution of the work as a whole. * + * * + * As another special exception to the GPL terms, Aperture Labs Ltd. grants* + * permission to link the code of this program with any version of the * + * OpenSSL library which is distributed under a license identical to that * + * listed in the included docs/licenses/OpenSSL.txt file, and distribute * + * linked combinations including the two. * + * * + * Any redistribution of Covered Software, including any derived works, * + * must obey and carry forward all of the terms of this license, including * + * obeying all GPL rules and restrictions. For example, source code of * + * the whole work must be provided and free redistribution must be * + * allowed. All GPL references to "this License", are to be treated as * + * including the terms and conditions of this license text as well. * + * * + * Because this license imposes special exceptions to the GPL, Covered * + * Work may not be combined (even as part of a larger work) with plain GPL * + * software. The terms, conditions, and exceptions of this license must * + * be included as well. This license is incompatible with some other open * + * source licenses as well. In some cases we can relicense portions of * + * RFIDler or grant special permissions to use it in other open source * + * software. Please contact sales@aperturelabs.com with any such requests.* + * Similarly, we don't incorporate incompatible open source software into * + * Covered Software without special permission from the copyright holders. * + * * + * If you have any questions about the licensing restrictions on using * + * RFIDler in other works, are happy to help. As mentioned above, we also * + * offer alternative license to integrate RFIDler into proprietary * + * applications and appliances. These contracts have been sold to dozens * + * of software vendors, and generally include a perpetual license as well * + * as providing for priority support and updates. They also fund the * + * continued development of RFIDler. Please email sales@aperturelabs.com * + * for further information. * + * If you have received a written license agreement or contract for * + * Covered Software stating terms other than these, you may choose to use * + * and redistribute Covered Software under those terms instead of these. * + * * + * Source is provided to this software because we believe users have a * + * right to know exactly what a program is going to do before they run it. * + * This also allows you to audit the software for security holes (none * + * have been found so far). * + * * + * Source code also allows you to port RFIDler to new platforms, fix bugs, * + * and add new features. You are highly encouraged to send your changes * + * to the RFIDler mailing list for possible incorporation into the * + * main distribution. By sending these changes to Aperture Labs Ltd. or * + * one of the Aperture Labs Ltd. development mailing lists, or checking * + * them into the RFIDler source code repository, it is understood (unless * + * you specify otherwise) that you are offering the RFIDler Project * + * (Aperture Labs Ltd.) the unlimited, non-exclusive right to reuse, * + * modify, and relicense the code. RFIDler will always be available Open * + * Source, but this is important because the inability to relicense code * + * has caused devastating problems for other Free Software projects (such * + * as KDE and NASM). We also occasionally relicense the code to third * + * parties as discussed above. If you wish to specify special license * + * conditions of your contributions, just say so when you send them. * + * * + * This program 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 RFIDler * + * license file for more details (it's in a COPYING file included with * + * RFIDler, and also available from * + * https://github.com/ApertureLabsLtd/RFIDler/COPYING * + * * + ***************************************************************************/ + +// Author: Adam Laurie + +/* + * Hitag Crypto support macros + * These macros reverse the bit order in a byte, or *within* each byte of a + * 16 , 32 or 64 bit unsigned integer. (Not across the whole 16 etc bits.) + */ +#define rev8(X) ((((X) >> 7) &1) + (((X) >> 5) &2) + (((X) >> 3) &4) \ + + (((X) >> 1) &8) + (((X) << 1) &16) + (((X) << 3) &32) \ + + (((X) << 5) &64) + (((X) << 7) &128) ) +#define rev16(X) (rev8 (X) + (rev8 (X >> 8) << 8)) +#define rev32(X) (rev16(X) + (rev16(X >> 16) << 16)) +#define rev64(X) (rev32(X) + (rev32(X >> 32) << 32)) + + +unsigned long hexreversetoulong(BYTE *hex); +unsigned long long hexreversetoulonglong(BYTE *hex); + diff --git a/tools/hitag2crack/crack5gpu/utilpart.c b/tools/hitag2crack/crack5gpu/utilpart.c new file mode 100644 index 000000000..c46148491 --- /dev/null +++ b/tools/hitag2crack/crack5gpu/utilpart.c @@ -0,0 +1,180 @@ +/*************************************************************************** + * A copy of the GNU GPL is appended to this file. * + * * + * This licence is based on the nmap licence, and we express our gratitude * + * for the work that went into producing it. There is no other connection * + * between RFIDler and nmap either expressed or implied. * + * * + ********************** IMPORTANT RFIDler LICENSE TERMS ******************** + * * + * * + * All references to RFIDler herein imply all it's derivatives, namely: * + * * + * o RFIDler-LF Standard * + * o RFIDler-LF Lite * + * o RFIDler-LF Nekkid * + * * + * * + * RFIDler is (C) 2013-2014 Aperture Labs Ltd. * + * * + * This program is free software; you may redistribute and/or modify it * + * under the terms of the GNU General Public License as published by the * + * Free Software Foundation; Version 2 ("GPL"), BUT ONLY WITH ALL OF THE * + * CLARIFICATIONS AND EXCEPTIONS DESCRIBED HEREIN. This guarantees your * + * right to use, modify, and redistribute this software under certain * + * conditions. If you wish to embed RFIDler technology into proprietary * + * software or hardware, we sell alternative licenses * + * (contact sales@aperturelabs.com). * + * * + * Note that the GPL places important restrictions on "derivative works", * + * yet it does not provide a detailed definition of that term. To avoid * + * misunderstandings, we interpret that term as broadly as copyright law * + * allows. For example, we consider an application to constitute a * + * derivative work for the purpose of this license if it does any of the * + * following with any software or content covered by this license * + * ("Covered Software"): * + * * + * o Integrates source code from Covered Software. * + * * + * o Is designed specifically to execute Covered Software and parse the * + * results (as opposed to typical shell or execution-menu apps, which will * + * execute anything you tell them to). * + * * + * o Includes Covered Software in a proprietary executable installer. The * + * installers produced by InstallShield are an example of this. Including * + * RFIDler with other software in compressed or archival form does not * + * trigger this provision, provided appropriate open source decompression * + * or de-archiving software is widely available for no charge. For the * + * purposes of this license, an installer is considered to include Covered * + * Software even if it actually retrieves a copy of Covered Software from * + * another source during runtime (such as by downloading it from the * + * Internet). * + * * + * o Links (statically or dynamically) to a library which does any of the * + * above. * + * * + * o Executes a helper program, module, or script to do any of the above. * + * * + * This list is not exclusive, but is meant to clarify our interpretation * + * of derived works with some common examples. Other people may interpret * + * the plain GPL differently, so we consider this a special exception to * + * the GPL that we apply to Covered Software. Works which meet any of * + * these conditions must conform to all of the terms of this license, * + * particularly including the GPL Section 3 requirements of providing * + * source code and allowing free redistribution of the work as a whole. * + * * + * As another special exception to the GPL terms, Aperture Labs Ltd. grants* + * permission to link the code of this program with any version of the * + * OpenSSL library which is distributed under a license identical to that * + * listed in the included docs/licenses/OpenSSL.txt file, and distribute * + * linked combinations including the two. * + * * + * Any redistribution of Covered Software, including any derived works, * + * must obey and carry forward all of the terms of this license, including * + * obeying all GPL rules and restrictions. For example, source code of * + * the whole work must be provided and free redistribution must be * + * allowed. All GPL references to "this License", are to be treated as * + * including the terms and conditions of this license text as well. * + * * + * Because this license imposes special exceptions to the GPL, Covered * + * Work may not be combined (even as part of a larger work) with plain GPL * + * software. The terms, conditions, and exceptions of this license must * + * be included as well. This license is incompatible with some other open * + * source licenses as well. In some cases we can relicense portions of * + * RFIDler or grant special permissions to use it in other open source * + * software. Please contact sales@aperturelabs.com with any such requests.* + * Similarly, we don't incorporate incompatible open source software into * + * Covered Software without special permission from the copyright holders. * + * * + * If you have any questions about the licensing restrictions on using * + * RFIDler in other works, are happy to help. As mentioned above, we also * + * offer alternative license to integrate RFIDler into proprietary * + * applications and appliances. These contracts have been sold to dozens * + * of software vendors, and generally include a perpetual license as well * + * as providing for priority support and updates. They also fund the * + * continued development of RFIDler. Please email sales@aperturelabs.com * + * for further information. * + * If you have received a written license agreement or contract for * + * Covered Software stating terms other than these, you may choose to use * + * and redistribute Covered Software under those terms instead of these. * + * * + * Source is provided to this software because we believe users have a * + * right to know exactly what a program is going to do before they run it. * + * This also allows you to audit the software for security holes (none * + * have been found so far). * + * * + * Source code also allows you to port RFIDler to new platforms, fix bugs, * + * and add new features. You are highly encouraged to send your changes * + * to the RFIDler mailing list for possible incorporation into the * + * main distribution. By sending these changes to Aperture Labs Ltd. or * + * one of the Aperture Labs Ltd. development mailing lists, or checking * + * them into the RFIDler source code repository, it is understood (unless * + * you specify otherwise) that you are offering the RFIDler Project * + * (Aperture Labs Ltd.) the unlimited, non-exclusive right to reuse, * + * modify, and relicense the code. RFIDler will always be available Open * + * Source, but this is important because the inability to relicense code * + * has caused devastating problems for other Free Software projects (such * + * as KDE and NASM). We also occasionally relicense the code to third * + * parties as discussed above. If you wish to specify special license * + * conditions of your contributions, just say so when you send them. * + * * + * This program 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 RFIDler * + * license file for more details (it's in a COPYING file included with * + * RFIDler, and also available from * + * https://github.com/ApertureLabsLtd/RFIDler/COPYING * + * * + ***************************************************************************/ + +// Author: Adam Laurie + + +#include +#include +#include "HardwareProfile.h" +#include "util.h" +#include "rfidler.h" +//#include "comms.h" + +// rtc +rtccTime RTC_time; // time structure +rtccDate RTC_date; // date structure + +// convert byte-reversed 8 digit hex to unsigned long +unsigned long hexreversetoulong(BYTE *hex) { + unsigned long ret = 0L; + unsigned int x; + BYTE i; + + if (strlen(hex) != 8) + return 0L; + + for (i = 0 ; i < 4 ; ++i) { + if (sscanf(hex, "%2X", &x) != 1) + return 0L; + ret += ((unsigned long) x) << i * 8; + hex += 2; + } + return ret; +} + +// convert byte-reversed 12 digit hex to unsigned long +unsigned long long hexreversetoulonglong(BYTE *hex) { + unsigned long long ret = 0LL; + BYTE tmp[9]; + + // this may seem an odd way to do it, but weird compiler issues were + // breaking direct conversion! + + tmp[8] = '\0'; + memset(tmp + 4, '0', 4); + memcpy(tmp, hex + 8, 4); + ret = hexreversetoulong(tmp); + ret <<= 32; + memcpy(tmp, hex, 8); + ret += hexreversetoulong(tmp); + return ret; +} + +