d'un CUDA je besoin d'aide s'il vous plaît. J'ai commencé à programmer un forer/password guesser brut commun avec CUDA (2.3/3.0beta). J'ai essayé différentes manières de générer tous les "candidats" en texte brut possibles d'un jeu de caractères ASCII défini.Générer toutes les combinaisons d'un tableau de caractères à l'intérieur du noyau
Dans cet exemple de code que je veux générer tous les 74^4 combinaisons possibles (et la sortie juste le résultat à l'hôte/stdout).
$ ./combinations
Total number of combinations : 29986576
Maximum output length : 4
ASCII charset length : 74
ASCII charset : 0x30 - 0x7a
":;<=>[email protected][\]^_`abcdefghijklmnopqrstuvwxy"
code CUDA (compilé avec 2.3 et 3.0b - sm_10) - combinaions.cu:
#include <stdio.h>
#include <cuda.h>
__device__ uchar4 charset_global = {0x30, 0x30, 0x30, 0x30};
__shared__ __device__ uchar4 charset[128];
__global__ void combo_kernel(uchar4 * result_d, unsigned int N)
{
int totalThreads = blockDim.x * gridDim.x ;
int tasksPerThread = (N % totalThreads) == 0 ? N/totalThreads : N/totalThreads + 1;
int myThreadIdx = blockIdx.x * blockDim.x + threadIdx.x ;
int endIdx = myThreadIdx + totalThreads * tasksPerThread ;
if(endIdx > N) endIdx = N;
const unsigned int m = 74 + 0x30;
for(int idx = myThreadIdx ; idx < endIdx ; idx += totalThreads) {
charset[threadIdx.x].x = charset_global.x;
charset[threadIdx.x].y = charset_global.y;
charset[threadIdx.x].z = charset_global.z;
charset[threadIdx.x].w = charset_global.w;
__threadfence();
if(charset[threadIdx.x].x < m) {
charset[threadIdx.x].x++;
} else if(charset[threadIdx.x].y < m) {
charset[threadIdx.x].x = 0x30; // = 0
charset[threadIdx.x].y++;
} else if(charset[threadIdx.x].z < m) {
charset[threadIdx.x].y = 0x30; // = 0
charset[threadIdx.x].z++;
} else if(charset[threadIdx.x].w < m) {
charset[threadIdx.x].z = 0x30;
charset[threadIdx.x].w++;; // = 0
}
charset_global.x = charset[threadIdx.x].x;
charset_global.y = charset[threadIdx.x].y;
charset_global.z = charset[threadIdx.x].z;
charset_global.w = charset[threadIdx.x].w;
result_d[idx].x = charset_global.x;
result_d[idx].y = charset_global.y;
result_d[idx].z = charset_global.z;
result_d[idx].w = charset_global.w;
}
}
#define BLOCKS 65535
#define THREADS 128
int main(int argc, char **argv)
{
const int ascii_chars = 74;
const int max_len = 4;
const unsigned int N = pow((float)ascii_chars, max_len);
size_t size = N * sizeof(uchar4);
uchar4 *result_d, *result_h;
result_h = (uchar4 *)malloc(size);
cudaMalloc((void **)&result_d, size);
cudaMemset(result_d, 0, size);
printf("Total number of combinations\t: %d\n\n", N);
printf("Maximum output length\t: %d\n", max_len);
printf("ASCII charset length\t: %d\n\n", ascii_chars);
printf("ASCII charset\t: 0x30 - 0x%02x\n ", 0x30 + ascii_chars);
for(int i=0; i < ascii_chars; i++)
printf("%c",i + 0x30);
printf("\n\n");
combo_kernel <<< BLOCKS, THREADS >>> (result_d, N);
cudaThreadSynchronize();
printf("CUDA kernel done\n");
printf("hit key to continue...\n");
getchar();
cudaMemcpy(result_h, result_d, size, cudaMemcpyDeviceToHost);
for (unsigned int i=0; i<N; i++)
printf("result[%06u]\t%c%c%c%c\n",i, result_h[i].x, result_h[i].y, result_h[i].z, result_h[i].w);
free(result_h);
cudaFree(result_d);
}
Le code devrait compiler sans problème, mais la sortie est pas ce que je pensais.
En mode d'émulation:
CUDA kernel done hit
key to continue...
result[000000] 1000
...
result[000128] 5000
Sur le mode de sortie:
CUDA kernel done hit
key to continue...
result[000000] 1000
...
result[012288] 5000
J'ai aussi utilisé __threadfence() et ou __syncthreads() sur différentes lignes du code aussi sans succès ...
ps. Si possible, je veux générer tout à l'intérieur de la fonction du noyau. J'ai aussi essayé « pré » génération des candidats en texte clair possible à l'intérieur hôte fonction principale et memcpy à dispositif, cela ne fonctionne qu'avec une taille charset très limitée (en raison de la mémoire de l'appareil limité).
une idée sur la sortie, pourquoi le répéter (même avec __threadfence() ou __syncthreads())?
toute autre méthode pour générer le texte brut (candidats) à l'intérieur du noyau CUDA rapide :-) (~ 75^8)?
grâce un million
accueille janvier
Salut, merci pour votre réponse! Mon idée était que "__device__ uchar4 charset_global" est une sorte de matrice principale. chaque bloc de thread doit aller chercher "valeur actuelle de charset_global" dans le jeu de caractères partagé [128], faire la combinaison suivante (remplir un calcul avec le jeu de caractères ici) et enfin écrire la combinaison "déjà calculé par le fil" . (ainsi le thread suivant peut utiliser la "combinaison déjà faite" comme décalage). J'espère que vous m'avez bien compris;)) ps. "Vos instructions if ne réinitialisent pas correctement les itérateurs de boucle" - doivent être correctes sur userland - origin: combfunc aocp – sead
Je n'ai aucune idée de ce que signifie "travailler correctement dans userland", mais vous pouvez voir l'utilisation du code dans mon edit qu'il y a effectivement des problèmes avec l'itération de la boucle. – Eric
L'algorithme que vous décrivez (dans votre commentaire) est un algorithme série. En d'autres termes, aucun thread ne peut calculer un mot de passe unique avant d'avoir obtenu le résultat d'un thread précédent. Aucun thread ne peut fonctionner en parallèle car il commencerait par le même mot de passe initial et le permuterait de la même manière, produisant une sortie dupliquée. La façon de paralléliser ceci est de comprendre que vous générerez 74^N combinaisons possibles et chaque thread générera 74^N/M de ces combinaisons complètement indépendant de ce que n'importe quel autre thread fait. – Eric