Insomni’hack 2013 – Armory level3

This challenge was the last level on the ARM platform. It was a crackme with a stripped binary including a basic anti-debugging trick. Sadly, only one team managed to complete this challenge before the end of Insomni’hack and another wasn’t far from what we discussed later.

Running the binary alone we can learn that two things should be provided: a username and a serial number. Maybe the two are linked, maybe not…

The check for the username is easily spotted using IDA and searching for cross-references to strcmp(). Also, the serial is XOR’d with this username.


.text:00008AE8 E4 01 9F E5 LDR R0, =aJackknife ; "jackknife"
.text:00008AEC 14 10 1B E5 LDR R1, [R11,#s2] ; s2
.text:00008AF0 68 FE FF EB BL strcmp
.text:00008AF4 00 30 A0 E1 MOV R3, R0
.text:00008AF8 00 00 53 E3 CMP R3, #0
.text:00008AFC 08 00 00 1A BNE loc_8b24
...
.text:00008B70 1C 20 1B E5 LDR R2, [R11,#var_1C]
.text:00008B74 10 30 1B E5 LDR R3, [R11,#var_10]
.text:00008B78 03 30 82 E0 ADD R3, R2, R3
.text:00008B7C 14 10 1B E5 LDR R1, [R11,#s2]
.text:00008B80 10 20 1B E5 LDR R2, [R11,#var_10]
.text:00008B84 02 20 81 E0 ADD R2, R1, R2
.text:00008B88 00 10 D2 E5 LDRB R1, [R2]
.text:00008B8C 18 00 1B E5 LDR R0, [R11,#s]
.text:00008B90 10 20 1B E5 LDR R2, [R11,#var_10]
.text:00008B94 02 20 80 E0 ADD R2, R0, R2
.text:00008B98 00 20 D2 E5 LDRB R2, [R2]
.text:00008B9C 02 20 21 E0 EOR R2, R1, R2
.text:00008BA0 72 20 EF E6 UXTB R2, R2
.text:00008BA4 00 20 C3 E5 STRB R2, [R3]
.text:00008BA8 10 30 1B E5 LDR R3, [R11,#var_10]
.text:00008BAC 01 30 83 E2 ADD R3, R3, #1
.text:00008BB0 10 30 0B E5 STR R3, [R11,#var_10]

Next step: run the binary in GDB in order to see how the serial number is validated.


level3@sploitboard:~$ gdb ./level3
(gdb) r foo bar
Starting program: /home/level3/level3 foo bar
Ammo store
----------
Validating your credentials to download ammo for your weapons...

Program received signal SIGTRAP, Trace/breakpoint trap.
0xb6ec0bfc in raise () from /lib/arm-linux-gnueabihf/libc.so.6

Wait what?!? You didn’t put any breakpoint and still a SIGTRAP is raised. Looking more closely we can see that the raise(3) function is called multiple times in the application so it seems that this is our anti-debugging trick.


Direction Type Address Text
--------- ---- ------- ----
Up p sub_889C+20 BL raise
Up p sub_88C8+18 BL raise
Up p sub_88F0+1C BL raise
Up p sub_891C+1C BL raise
Up p sub_8948+18 BL raise
Up p sub_8988+C BL raise
p sub_8A48+9C BL raise
Down p sub_8A48+F0 BL raise
Down p sub_8A48+188 BL raise

We can also see that a signal handler is defined for SIGTRAP and SIGABRT and that this handler is incrementing a global variable:


.text:000089F4 LDR R3, =sig_handler
.text:000089F8 STR R3, [R11,#act]
.text:000089FC SUB R3, R11, #-act
.text:00008A00 MOV R0, #5 ; sig
.text:00008A04 MOV R1, R3 ; act
.text:00008A08 MOV R2, #0 ; oact
.text:00008A0C BL sigaction
.text:00008A10 SUB R3, R11, #-act
.text:00008A14 MOV R0, #6 ; sig
.text:00008A18 MOV R1, R3 ; act
.text:00008A1C MOV R2, #0 ; oact
.text:00008A20 BL
...
; sig_handler:
.text:000089BC 18 30 9F E5 LDR R3, =dword_10F7C
.text:000089C0 00 30 93 E5 LDR R3, [R3]
.text:000089C4 01 20 83 E2 ADD R2, R3, #1
.text:000089C8 0C 30 9F E5 LDR R3, =dword_10F7C
.text:000089CC 00 20 83 E5 STR R2, [R3]

The first check that is done on the XOR’d serial number is using this counter so we might have to run this code to know this value without reversing the complete application. Using a breakpoint on raise(3) and the call method in GDB we can emulate the SIGTRAP without actually triggering it.


.text:00008BD4 1C 30 1B E5 LDR R3, [R11,#XOR_serial]
.text:00008BD8 00 30 D3 E5 LDRB R3, [R3]
.text:00008BDC 03 20 A0 E1 MOV R2, R3
.text:00008BE0 DC 30 9F E5 LDR R3, =dword_10F7C
.text:00008BE4 00 30 93 E5 LDR R3, [R3]
.text:00008BE8 03 30 82 E0 ADD R3, R2, R3
.text:00008BEC 73 30 EF E6 UXTB R3, R3
.text:00008BF0 A3 00 53 E3 CMP R3, #0xA3

Breakpoint 1, 0xb6ec0bbc in raise () from /lib/arm-linux-gnueabihf/libc.so.6
(gdb) call (void)0x89a0(5)
$1 = 35232
(gdb) bt
#0 0xb6ec0bbc in raise () from /lib/arm-linux-gnueabihf/libc.so.6
#1 0x00008ae8 in ?? ()
#2 0x00008ae8 in ?? ()
Backtrace stopped: previous frame identical to this frame (corrupt stack?)
(gdb) set $pc=0x8ae8
(gdb) c
Breakpoint 4, 0x00008be0 in ?? ()
(gdb) x /5i $pc
=> 0x8be0: ldr r3, [pc, #220] ; 0x8cc4
0x8be4: ldr r3, [r3]
0x8be8: add r3, r2, r3
0x8bec: uxtb r3, r3
0x8bf0: cmp r3, #163 ; 0xa3
(gdb) p *0x10f7c
$22 = 2

This means that the XOR’d value added to 2 should be equal to 0xa3. Continuing this lenghty process, we can get all the characters of the serial number:


level3@sploitboard:~$ ./level3 jackknife `python -c 'print "xcb"+"bxeaxe2aaaaa"'`
Ammo store
----------
Validating your credentials to download ammo for your weapons...
36760a05c853e6a7444b6a3de2c100591e0ad9c0193c56ca562949f4f7342a808cead86e34f940fb88221403451345e36edddcce51934eb4ea64af7c27aa6650

Reverse engineering de binaires pour GPGPU

Introduction

L’utilisation des GPGPU (General Purpose Graphical Processing Unit) n’est pas un sujet nouveau dans le monde de la sécurité. De nombreux projets de récupération de mot-de-passe utilisant d’hors et déjà cette technologie dans le but d’améliorer leurs performances (pyrit, CUDA multihash bruteforcer, …). Lors de la conférence RUXCON 2008, Daniel Reynaud a fait une présentation sur l’utilisation des GPGPU par des codes malveillants. Ces derniers délocaliseraient l’exécution de certains algorithmes (par exemple celui générant les noms de domaines permettant de se connecter à un canal de contrôle) sur la carte graphique afin d’empêcher une analyse standard de code ia32.

Deux ans après cette présentation de tels codes malveillants se font rares notament du fait que les auteurs cherchent à infecter le plus de machines possibles, ce qui est plus compliqué quand le code est compilé pour un type de chipset graphique précis. Néanmoins, si l’on regarde les caractéristiques de la gamme Apple, plateforme devenant une cible grandissante pour les auteurs de malwares,  on s’aperçoit que la majorité des machines sont équipées d’une carte graphique NVIDIA et que le framework OpenCL est inclu dans l’installation de base. On pourrait imaginer que ces deux propriétés favoriseront l’utilisation des GPGPU lors d’attaques ciblant un système OSX.

La possible utilisation des GPGPU par des codes malveillants ou des solutions de protection logicielle sont des motifs poussant à se pencher sur l’analyse du code compilé pour ces plateformes (et également pour la beauté de l’Art comme diraient certains).

Les bases

Différents frameworks existent pour le développement de code pour GPGPU, comme par exemple CUDA (NVIDIA) et OpenCL (sur-couche permettant d’obtenir du code compilable sur différentes architectures). Le développement se fait dans une extension du C et le passage d’arguments est réalisé via des appels à des fonctions prédéfinies depuis le code exécuté sur le CPU (ex: CUDA!cudaMemcpy, openCL!clEnqueueReadBuffer, …). Dans le cas d’un code utilisant le GPGPU pour décoder des informations, la récupération du texte-clair lors d’une analyse serait simplifiée via la mise en place de breakpoints sur ces API. Ci-dessous, un exemple de kernel CUDA réalisant un parcours de chaine de caractères.

[sourcecode language=”cpp”]
// not so usefull function
__global__ void mystrlen(char* str, int* length)
{
int i = 0;

while (str[i])
++i;
*length = i;
}
[/sourcecode]

L’appel à cette fonction peut se faire de la manière suivante depuis le code C:

[sourcecode language=”cpp”]
cudaMalloc((void**)&devLen, sizeof (int));
cudaMalloc((void**)&devStr, MAX_SIZE * sizeof (char));
cudaMemcpy(devStr, argv[1], strlen(argv[1]) + 1, cudaMemcpyHostToDevice);
mystrlen <<< 1, 1 >>> (devStr, devLen);
cudaMemcpy(&len, devLen, sizeof (int), cudaMemcpyDeviceToHost);
[/sourcecode]

Dans le cas de CUDA, le code est compilé dans une forme intermédiaire, PTX, qui sera ensuite inclu dans le binaire. En spécifiant une architecture donnée, le PTX est compilé en CUBIN et optimisé pour cette dernière.

Extraction du code PTX

Dans le cas où l’application a été compilée sans spécifier d’architecture, le code PTX sera directement présent dans le binaire et compilé pour l’architecture virtuelle sm_10 par défaut. Une recherche sur des mots clefs comme .target permet de le localiser. Travis Goodspeed a développé un outil réalisant l’extraction du code PTX depuis ce type de binaire.

[sourcecode language=”bash”]
$ nvcc mystrlen.cu
$ cudaDump < a.out
.version 1.4
.target sm_10, map_f64_to_f32
// compiled with /usr/local/cuda/bin//../open64/lib//be
// nvopencc 3.2 built on 2010-11-11

.entry _Z8mystrlenPcPi (
.param .u32 __cudaparm__Z8mystrlenPcPi_str,
.param .u32 __cudaparm__Z8mystrlenPcPi_length)
{
.reg .u32 %r<10>;
.reg .pred %p<4>;
.loc 28 16 0
$LDWbegin__Z8mystrlenPcPi:
ld.param.u32 %r1, [__cudaparm__Z8mystrlenPcPi_str];
ld.global.s8 %r2, [%r1+0];
mov.u32 %r3, 0;
setp.eq.s32 %p1, %r2, %r3;
@%p1 bra $Lt_0_2306;
ld.param.u32 %r1, [__cudaparm__Z8mystrlenPcPi_str];
mov.s32 %r4, %r1;
mov.s32 %r5, 0;
$Lt_0_1794:
//<loop> Loop body line 16, nesting depth: 1, estimated iterations: unknown
.loc 28 21 0
add.s32 %r5, %r5, 1;
add.u32 %r4, %r4, 1;
ld.global.s8 %r6, [%r4+0];
mov.u32 %r7, 0;
setp.ne.s32 %p2, %r6, %r7;
@%p2 bra $Lt_0_1794;
bra.uni $Lt_0_1282;
$Lt_0_2306:
mov.s32 %r5, 0;
$Lt_0_1282:
.loc 28 22 0
ld.param.u32 %r8, [__cudaparm__Z8mystrlenPcPi_length];
st.global.s32 [%r8+0], %r5;
.loc 28 23 0 exit;
$LDWend__Z8mystrlenPcPi:
} // _Z8mystrlenPcPi
[/sourcecode]

Dans le cas d’un binaire compilé pour une architecture spécifique, comme dans l’exemple ci-dessous, l’outil decuda permet à partir du CUBIN d’obtenir le code PTX, excepté pour du code compilé à partir de CUDA 3.x du à un changement de format (CUBIN > ELF). La solution consiste à utiliser le script elfToCubin.py pour réaliser la conversion.

[sourcecode language=”bash”]
$ nvcc mystrlen.cu -arch=compute_10 -code=sm_10,sm_13
[/sourcecode]

Le binaire compilé à l’aide de la ligne de commande précédente encapsule les 2 binaires ELF correspondants à sm_10 et sm_13 dans le segment nommé __const.

ELF dans la section __const du binaire

Analyse du code

Une des premières choses que l’on peut remarquer en lisant le code PTX ci-dessus est le fait qu’une partie des symboles a été conservée. On retrouve mystrlen et ses deux arguments str et length.

Les instructions se décomposent de la manière suivante: opération.<espace>.type. La première instruction de la fonction mystrlen lit (ld)  la valeur de type entier non-signé (u32) contenue dans le paramètre (param) str et la stocke dans le registre r1.

A la ligne 19, l’opération de comparaison setp est utilisée afin de vérifier si les registres r2 et r3 sont égaux (eq) et stocke le résultat dans le registre de prédiction p1. La ligne suivante correspond à une opération de modification du flux d’exécution. Ici, l’opération à réalisée (bra) est précédée d’un “predicate guard” (@[!]<predicate register>). Dans ce cas, le saut au label $Lt_0_2386 sera exécuté si le registre p1 contient la valeur vrai (r2 == r3). On retrouve le corps de la boucle while entre les lignes 27 et 33 avec les registres r4 et r5 qui représentent respectivement un pointeur sur str et i.

La documentation fournie par NVIDIA permet d’en apprendre plus sur ce langage, mais le jeu d’instructions reste limité et assez proche de langages assembleurs courants pour qu’il soit facilement compréhensible.

En plus d’une analyse statique, il est possible de réaliser une analyse dynamique de kernel CUDA à l’aide de gdb-cuda.