From: miha-q <> Date: Mon, 4 Mar 2024 14:29:43 +0000 (-0500) Subject: Mon Mar 4 09:29:43 AM EST 2024 X-Git-Url: http://www.foleosoft.com/?a=commitdiff_plain;h=28eeb71ca61a81d74746140c1f03a04f09075c5e;p=QAnsel.git Mon Mar 4 09:29:43 AM EST 2024 --- diff --git a/src/QAnsel.c b/src/QAnsel.c index f90ee3a..79104d7 100644 --- a/src/QAnsel.c +++ b/src/QAnsel.c @@ -2,16 +2,16 @@ #include #include #include -#include "complex.c" -#include "gates.c" +#include "complex2.c" +#include "gates2.c" #include "display.c" #include "chacha20.c" #define QUBITS_MAX 14 -uint8_t HIDDEN_VARIABLE = 0; +unsigned char HIDDEN_VARIABLE = 0; FILE* RANDOM_FILE; -#define GPU_ENABLED -uint8_t USE_GPU = 0; +//#define GPU_ENABLED +unsigned char USE_GPU = 0; #ifdef GPU_ENABLED #define CL_USE_DEPRECATED_OPENCL_1_2_APIS #define CL_TARGET_OPENCL_VERSION 120 @@ -23,14 +23,14 @@ uint8_t USE_GPU = 0; typedef struct { char n[128]; - uint8_t q0, q1, q2; + unsigned char q0, q1, q2; float arg0, arg1, arg2; } QInstr; float qansel_rand_s(float s) { - uint32_t tmp; - memcpy(&tmp, &s, sizeof(uint32_t)); + unsigned int tmp; + memcpy(&tmp, &s, sizeof(unsigned int)); srand(tmp); } float qansel_rand_h() @@ -41,8 +41,8 @@ float qansel_rand_t() { if (RANDOM_FILE) { - uint32_t num = 0; - for (uint8_t i = 0; i < 4; i++) + unsigned int num = 0; + for (unsigned char i = 0; i < 4; i++) { num = (num << 8) | fgetc(RANDOM_FILE); } @@ -61,18 +61,18 @@ float qansel_rand() return HIDDEN_VARIABLE ? qansel_rand_h() : qansel_rand_t(); } -void qansel_cnot(cpx_mtx_t* stateVector, uint8_t qubitCount, uint8_t bitA, uint8_t bitB) +void qansel_cnot(cpx_mtx_t* stateVector, unsigned char qubitCount, unsigned char bitA, unsigned char bitB) { - uint32_t retLen = (uint32_t)pow(2, qubitCount); + unsigned int retLen = (unsigned int)pow(2, qubitCount); cpx_mtx_t ret; cpx_mtx_init(&ret, 1, retLen); cpx_t n; - for (uint32_t i = 0; i < retLen; i++) //asdfasdfsdfsdf + for (unsigned int i = 0; i < retLen; i++) //asdfasdfsdfsdf { - uint8_t bitAVal = (i >> bitA) & 1; - uint8_t bitBVal = (i >> bitB) & 1; - uint8_t bitBNew = bitAVal ? !bitBVal : bitBVal; - uint32_t j = (i & ~(1 << bitB)) | (bitBNew << bitB); + unsigned char bitAVal = (i >> bitA) & 1; + unsigned char bitBVal = (i >> bitB) & 1; + unsigned char bitBNew = bitAVal ? !bitBVal : bitBVal; + unsigned int j = (i & ~(1 << bitB)) | (bitBNew << bitB); cpx_mtx_get(stateVector, 0, i, &n); cpx_mtx_set(&ret, 0, j, &n); } @@ -82,19 +82,19 @@ void qansel_cnot(cpx_mtx_t* stateVector, uint8_t qubitCount, uint8_t bitA, uint8 stateVector->cols = ret.cols; } -void qansel_swap(cpx_mtx_t* stateVector, uint8_t qubitCount, uint8_t bitA, uint8_t bitB) +void qansel_swap(cpx_mtx_t* stateVector, unsigned char qubitCount, unsigned char bitA, unsigned char bitB) { - uint32_t retLen = (uint32_t)pow(2, qubitCount); + unsigned int retLen = (unsigned int)pow(2, qubitCount); cpx_mtx_t ret; cpx_mtx_init(&ret, 1, retLen); cpx_t n; - for (uint32_t i = 0; i < retLen; i++) + for (unsigned int i = 0; i < retLen; i++) { - uint8_t bitAVal = (i >> bitA) & 1; - uint8_t bitBVal = (i >> bitB) & 1; - uint8_t bitANew = bitBVal; - uint8_t bitBNew = bitAVal; - uint32_t j = (i & ~((1 << bitA) | (1 << bitB))) | ((bitANew << bitA) | (bitBNew << bitB)); + unsigned char bitAVal = (i >> bitA) & 1; + unsigned char bitBVal = (i >> bitB) & 1; + unsigned char bitANew = bitBVal; + unsigned char bitBNew = bitAVal; + unsigned int j = (i & ~((1 << bitA) | (1 << bitB))) | ((bitANew << bitA) | (bitBNew << bitB)); cpx_mtx_get(stateVector, 0, i, &n); cpx_mtx_set(&ret, 0, j, &n); } @@ -104,20 +104,20 @@ void qansel_swap(cpx_mtx_t* stateVector, uint8_t qubitCount, uint8_t bitA, uint8 stateVector->cols = ret.cols; } -void qansel_fredkin(cpx_mtx_t* stateVector, uint8_t qubitCount, uint8_t bitA, uint8_t bitB, uint8_t bitC) +void qansel_fredkin(cpx_mtx_t* stateVector, unsigned char qubitCount, unsigned char bitA, unsigned char bitB, unsigned char bitC) { - uint32_t retLen = (uint32_t)pow(2, qubitCount); + unsigned int retLen = (unsigned int)pow(2, qubitCount); cpx_mtx_t ret; cpx_mtx_init(&ret, 1, retLen); cpx_t n; - for (uint32_t i = 0; i < retLen; i++) + for (unsigned int i = 0; i < retLen; i++) { - uint8_t bitAVal = (i >> bitA) & 1; - uint8_t bitBVal = (i >> bitB) & 1; - uint8_t bitCVal = (i >> bitC) & 1; - uint8_t bitBNew = bitAVal ? bitCVal : bitBVal; - uint8_t bitCNew = bitAVal ? bitBVal : bitCVal; - uint32_t j = (i & ~((1 << bitB) | (1 << bitC))) | ((bitBNew << bitB) | (bitCNew << bitC)); + unsigned char bitAVal = (i >> bitA) & 1; + unsigned char bitBVal = (i >> bitB) & 1; + unsigned char bitCVal = (i >> bitC) & 1; + unsigned char bitBNew = bitAVal ? bitCVal : bitBVal; + unsigned char bitCNew = bitAVal ? bitBVal : bitCVal; + unsigned int j = (i & ~((1 << bitB) | (1 << bitC))) | ((bitBNew << bitB) | (bitCNew << bitC)); cpx_mtx_get(stateVector, 0, i, &n); cpx_mtx_set(&ret, 0, j, &n); } @@ -128,19 +128,19 @@ void qansel_fredkin(cpx_mtx_t* stateVector, uint8_t qubitCount, uint8_t bitA, ui } -void qansel_toffoli(cpx_mtx_t* stateVector, uint8_t qubitCount, uint8_t bitA, uint8_t bitB, uint8_t bitC) +void qansel_toffoli(cpx_mtx_t* stateVector, unsigned char qubitCount, unsigned char bitA, unsigned char bitB, unsigned char bitC) { - uint32_t retLen = (uint32_t)pow(2, qubitCount); + unsigned int retLen = (unsigned int)pow(2, qubitCount); cpx_mtx_t ret; cpx_mtx_init(&ret, 1, retLen); cpx_t n; - for (uint32_t i = 0; i < retLen; i++) + for (unsigned int i = 0; i < retLen; i++) { - uint8_t bitAVal = (i >> bitA) & 1; - uint8_t bitBVal = (i >> bitB) & 1; - uint8_t bitCVal = (i >> bitC) & 1; - uint8_t bitCNew = (bitAVal && bitBVal) ? !bitCVal : bitCVal; - uint32_t j = (i & ~(1 << bitC)) | (bitCNew << bitC); + unsigned char bitAVal = (i >> bitA) & 1; + unsigned char bitBVal = (i >> bitB) & 1; + unsigned char bitCVal = (i >> bitC) & 1; + unsigned char bitCNew = (bitAVal && bitBVal) ? !bitCVal : bitCVal; + unsigned int j = (i & ~(1 << bitC)) | (bitCNew << bitC); cpx_mtx_get(stateVector, 0, i, &n); cpx_mtx_set(&ret, 0, j, &n); } @@ -170,7 +170,7 @@ float* qansel_unitary(float theta, float phi, float lambda) return m.ptr; } -void qansel_instruction(cpx_mtx_t* stateVector, uint8_t qubitCount, QInstr* instr) +void qansel_instruction(cpx_mtx_t* stateVector, unsigned char qubitCount, QInstr* instr) { cpx_mtx_t tmp; cpx_mtx_t gate; @@ -194,17 +194,19 @@ void qansel_instruction(cpx_mtx_t* stateVector, uint8_t qubitCount, QInstr* inst cpx_t n; cpx_mtx_t filter; cpx_mtx_init(&filter, 2, 2); - uint8_t qubit = qubitCount - (instr->q0) - 1; + unsigned char qubit = qubitCount - (instr->q0) - 1; if (qubit == 0) { - memcpy(filter.ptr, gate_ptr, 16 * sizeof(float)); + //memcpy(filter.ptr, gate_ptr, 16 * sizeof(float)); + memcpy(filter.ptr, gate_ptr, 8 * sizeof(float)); } else { - memcpy(filter.ptr, Identity, 16 * sizeof(float)); + //memcpy(filter.ptr, Identity, 16 * sizeof(float)); + memcpy(filter.ptr, Identity, 8 * sizeof(float)); } - for (uint8_t i = 1; i < qubitCount; i++) + for (unsigned char i = 1; i < qubitCount; i++) { if (qubit == i) { @@ -217,7 +219,8 @@ void qansel_instruction(cpx_mtx_t* stateVector, uint8_t qubitCount, QInstr* inst tmp.rows = filter.rows * gate.rows; tmp.cols = filter.cols * gate.cols; - tmp.ptr = malloc((tmp.rows * 2) * (tmp.cols * 2) * sizeof(float)); + //tmp.ptr = malloc((tmp.rows * 2) * (tmp.cols * 2) * sizeof(float)); + tmp.ptr = malloc(tmp.rows * (tmp.cols * 2) * sizeof(float)); #ifdef GPU_ENABLED if (USE_GPU/* && (tmp.rows >= 512 || tmp.cols >= 512)*/) @@ -245,11 +248,18 @@ void qansel_instruction(cpx_mtx_t* stateVector, uint8_t qubitCount, QInstr* inst ); } #else - cpx_ncpx_knk_mt + //cpx_ncpx_knk_mt + //( + // tmp.ptr, tmp.rows, tmp.cols, + // filter.ptr, filter.rows, filter.cols, + // gate.ptr, gate.rows, gate.cols + //); + cpx_mtx_knk ( - tmp.ptr, tmp.rows, tmp.cols, - filter.ptr, filter.rows, filter.cols, - gate.ptr, gate.rows, gate.cols + tmp.ptr, filter.ptr, gate.ptr, + tmp.rows, tmp.cols, + filter.rows, filter.cols, + gate.rows, gate.cols ); #endif @@ -279,10 +289,15 @@ void qansel_instruction(cpx_mtx_t* stateVector, uint8_t qubitCount, QInstr* inst ); } #else - cpx_ncpx_mmul_mt + //cpx_ncpx_mmul_mt + //( + // tmp.ptr, stateVector->ptr, filter.ptr, + // stateVector->rows * 2, filter.cols * 2, stateVector->cols * 2 + //); + cpx_mtx_dot ( tmp.ptr, stateVector->ptr, filter.ptr, - stateVector->rows * 2, filter.cols * 2, stateVector->cols * 2 + stateVector->rows, filter.cols, stateVector->cols ); #endif free(stateVector->ptr); @@ -291,24 +306,24 @@ void qansel_instruction(cpx_mtx_t* stateVector, uint8_t qubitCount, QInstr* inst if (instr->n[0] == 'u') free(gate_ptr); } -uint8_t qansel_measure(cpx_mtx_t* stateVector, uint8_t qubitCount, uint8_t qubit) +unsigned char qansel_measure(cpx_mtx_t* stateVector, unsigned char qubitCount, unsigned char qubit) { - uint32_t qubitCountPow2 = (uint32_t)pow(2, qubitCount); + unsigned int qubitCountPow2 = (unsigned int)pow(2, qubitCount); cpx_t n; float prob0 = 0; - for (uint32_t i = 0; i < qubitCountPow2; i++) + for (unsigned int i = 0; i < qubitCountPow2; i++) { - uint8_t bit = (i >> qubit) & 1; + unsigned char bit = (i >> qubit) & 1; cpx_mtx_get(stateVector, 0, i, &n); if (bit == 0) prob0 += cpx_magsqr(&n); } float r = qansel_rand(); - uint8_t newBit = r < prob0 ? 0 : 1; + unsigned char newBit = r < prob0 ? 0 : 1; float probTot = 0; - for (uint32_t i = 0; i < qubitCountPow2; i++) + for (unsigned int i = 0; i < qubitCountPow2; i++) { - uint8_t bit = (i >> qubit) & 1; + unsigned char bit = (i >> qubit) & 1; cpx_mtx_get(stateVector, 0, i, &n); if (bit != newBit) { @@ -323,9 +338,9 @@ uint8_t qansel_measure(cpx_mtx_t* stateVector, uint8_t qubitCount, uint8_t qubit } float multiplier = sqrt(1 / probTot); - for (uint32_t i = 0; i < qubitCountPow2; i++) + for (unsigned int i = 0; i < qubitCountPow2; i++) { - uint8_t bit = (i >> qubit) & 1; + unsigned char bit = (i >> qubit) & 1; cpx_mtx_get(stateVector, 0, i, &n); if (bit == newBit) { @@ -338,20 +353,20 @@ uint8_t qansel_measure(cpx_mtx_t* stateVector, uint8_t qubitCount, uint8_t qubit return newBit; } -void qansel_run(uint8_t qubitCount, uint8_t bitCount, QInstr* instr, uint8_t* retBitVect, uint32_t instrLen, uint8_t gfx) +void qansel_run(unsigned char qubitCount, unsigned char bitCount, QInstr* instr, unsigned char* retBitVect, unsigned int instrLen, unsigned char gfx) { - uint32_t qubitCountPow2 = (uint32_t)pow(2, qubitCount); + unsigned int qubitCountPow2 = (unsigned int)pow(2, qubitCount); - uint8_t bitVector[bitCount]; + unsigned char bitVector[bitCount]; cpx_mtx_t stateVector; cpx_mtx_init(&stateVector, 1, qubitCountPow2); cpx_mtx_set2(&stateVector, 0, 0, 1, 0); if (gfx) display(&stateVector, qubitCount); - uint8_t flags; + unsigned char flags; - for (uint8_t i = 0; i < bitCount; i++) bitVector[i] = 0; + for (unsigned char i = 0; i < bitCount; i++) bitVector[i] = 0; - for (uint32_t i = 0; i < instrLen; i++) + for (unsigned int i = 0; i < instrLen; i++) { cpx_t qqq; cpx_mtx_get(&stateVector, 0, 0, &qqq); @@ -377,7 +392,7 @@ void qansel_run(uint8_t qubitCount, uint8_t bitCount, QInstr* instr, uint8_t* re } else if (strcmp(instr[i].n, "if_all") == 0) { - uint8_t val = 0; + unsigned char val = 0; for (int32_t j = bitCount - 1; j >= 0; j--) { val = (val << 1) | bitVector[j]; @@ -409,7 +424,7 @@ void qansel_run(uint8_t qubitCount, uint8_t bitCount, QInstr* instr, uint8_t* re { cpx_mtx_t tmp; cpx_mtx_init(&tmp, 1, 2); - for (uint32_t j = 0; j < qubitCountPow2; j++) + for (unsigned int j = 0; j < qubitCountPow2; j++) { if ((j >> instr[i].q0) & 1) { @@ -477,10 +492,10 @@ void qansel_run(uint8_t qubitCount, uint8_t bitCount, QInstr* instr, uint8_t* re } else if (strcmp(instr[i].n, "born_all") == 0) { - for (uint32_t j = 0; j < qubitCountPow2; j++) + for (unsigned int j = 0; j < qubitCountPow2; j++) { - uint32_t tmp = j; - for (uint8_t k = 0; k < qubitCount; k++) + unsigned int tmp = j; + for (unsigned char k = 0; k < qubitCount; k++) { putchar('0' + (tmp >> (qubitCount - 1) & 1)); tmp <<= 1; @@ -493,7 +508,7 @@ void qansel_run(uint8_t qubitCount, uint8_t bitCount, QInstr* instr, uint8_t* re else if (strcmp(instr[i].n, "born") == 0) { float prob = 0; - for (uint32_t j = 0; j < qubitCountPow2; j++) + for (unsigned int j = 0; j < qubitCountPow2; j++) { cpx_t n; cpx_mtx_get(&stateVector, 0, j, &n); @@ -509,8 +524,8 @@ void qansel_run(uint8_t qubitCount, uint8_t bitCount, QInstr* instr, uint8_t* re { HIDDEN_VARIABLE = 1; float tmp1 = (float)instr[i].arg0; - uint32_t tmp2; - memcpy(&tmp2, &tmp1, sizeof(uint32_t)); + unsigned int tmp2; + memcpy(&tmp2, &tmp1, sizeof(unsigned int)); srand(tmp2); } else if (strcmp(instr[i].n, "rand") == 0) @@ -520,18 +535,18 @@ void qansel_run(uint8_t qubitCount, uint8_t bitCount, QInstr* instr, uint8_t* re else if (strcmp(instr[i].n, "reset_all") == 0) { cpx_mtx_set2(&stateVector, 0, 0, 1, 0); - for (uint32_t j = 1; j < qubitCountPow2; j++) + for (unsigned int j = 1; j < qubitCountPow2; j++) { cpx_mtx_set2(&stateVector, 0, j, 0, 0); } - for (uint8_t j = 0; j < bitCount; j++) + for (unsigned char j = 0; j < bitCount; j++) { bitVector[j] = 0; } } else if (strcmp(instr[i].n, "resetq") == 0) { - uint8_t bit = qansel_measure(&stateVector, qubitCount, instr[i].q0); + unsigned char bit = qansel_measure(&stateVector, qubitCount, instr[i].q0); if (bit) { instr[i].n[0] = 'x'; @@ -556,7 +571,7 @@ void qansel_run(uint8_t qubitCount, uint8_t bitCount, QInstr* instr, uint8_t* re if (retBitVect != NULL) { - for (uint32_t i = 0; i < bitCount; i++) + for (unsigned int i = 0; i < bitCount; i++) { retBitVect[i] = bitVector[i]; } @@ -568,21 +583,21 @@ void process(int argc, char** argv) { struct timespec ts; clock_gettime(CLOCK_MONOTONIC, &ts); - float seed = (float)((uint64_t)ts.tv_sec * 1000000000LL + ts.tv_nsec); + float seed = (float)((unsigned long)ts.tv_sec * 1000000000LL + ts.tv_nsec); qansel_rand_s(seed); char** lines = malloc(0); - uint32_t* lineIDs = malloc(0); + unsigned int* lineIDs = malloc(0); char* text = malloc(0); - uint32_t textLen = 0; - uint32_t linesLen = 0; + unsigned int textLen = 0; + unsigned int linesLen = 0; int c; int pc = -1; - uint8_t comment = 0; - uint8_t commentM = 0; - uint32_t lineID = 1; - uint8_t skipSpaces = 1; - uint8_t inGate = 0; - uint8_t fullSample = 0; + unsigned char comment = 0; + unsigned char commentM = 0; + unsigned int lineID = 1; + unsigned char skipSpaces = 1; + unsigned char inGate = 0; + unsigned char fullSample = 0; while ( (c = getchar()) != EOF ) { if (c == '/' && commentM == 0 && comment == 0) @@ -628,7 +643,7 @@ void process(int argc, char** argv) } text = realloc(text, textLen + 1); text[textLen++] = 0; - lineIDs = realloc(lineIDs, (linesLen + 1) * sizeof(uint32_t)); + lineIDs = realloc(lineIDs, (linesLen + 1) * sizeof(unsigned int)); lineIDs[linesLen] = lineID; lines = realloc(lines, (linesLen + 1) * sizeof(char*)); lines[linesLen] = malloc(strlen(text) + 1); @@ -648,21 +663,21 @@ void process(int argc, char** argv) } free(text); - uint8_t qubitCount = 0xFF; - uint8_t bitCount = 0xFF; + unsigned char qubitCount = 0xFF; + unsigned char bitCount = 0xFF; QInstr* instr = malloc(0); - uint32_t instrLen = 0; - uint8_t doDisplay = 0; - uint8_t errFound = 0; - for (uint32_t i = 0; i < linesLen; i++) + unsigned int instrLen = 0; + unsigned char doDisplay = 0; + unsigned char errFound = 0; + for (unsigned int i = 0; i < linesLen; i++) { lineID = i; char g; int q0, q1, q2; float a0, a1, a2; - uint32_t matchedCount; - uint32_t lineLen = strlen(lines[i]); + unsigned int matchedCount; + unsigned int lineLen = strlen(lines[i]); if (lineLen > 1) { while (lineLen > 1 && (lines[i][lineLen - 1] == ' ' || lines[i][lineLen - 1] == '\t')) @@ -865,7 +880,7 @@ void process(int argc, char** argv) } instr = realloc(instr, (instrLen + 1) * sizeof(QInstr)); strcpy(instr[instrLen].n, "density"); - instr[instrLen++].q0 = (uint8_t)q0; + instr[instrLen++].q0 = (unsigned char)q0; } else if ( @@ -893,7 +908,7 @@ void process(int argc, char** argv) instr = realloc(instr, (instrLen + 1) * sizeof(QInstr)); instr[instrLen].n[0] = g; instr[instrLen].n[1] = 0; - instr[instrLen++].q0 = (uint8_t)q0; + instr[instrLen++].q0 = (unsigned char)q0; } else if (sscanf(lines[i], "cx q[%i], q[%i]%n", &q0, &q1, &matchedCount) == 2) { @@ -1287,7 +1302,7 @@ void process(int argc, char** argv) } - for (uint32_t i = 0; i < linesLen; i++) free(lines[i]); + for (unsigned int i = 0; i < linesLen; i++) free(lines[i]); free(lines); if (errFound) @@ -1337,36 +1352,36 @@ void process(int argc, char** argv) if (fullSample) { - uint16_t stats[65536]; - for (uint32_t i = 0; i < (1 << bitCount); i++) + unsigned short stats[65536]; + for (unsigned int i = 0; i < (1 << bitCount); i++) { stats[i] = 0; } - uint8_t bitVect[bitCount]; + unsigned char bitVect[bitCount]; for (int i = 0; i < bitCount; i++) bitVect[i] = 0; - uint8_t* dat = malloc(instrLen * sizeof(QInstr)); + unsigned char* dat = malloc(instrLen * sizeof(QInstr)); memcpy(dat, instr, instrLen * sizeof(QInstr)); - uint32_t shots = 1000; - for (uint32_t i = 0; i < shots; i++) + unsigned int shots = 1000; + for (unsigned int i = 0; i < shots; i++) { qansel_run(qubitCount, bitCount, instr, bitVect, instrLen, doDisplay); memcpy(instr, dat, instrLen * sizeof(QInstr)); - uint16_t stat = 0; - for (int8_t j = bitCount - 1; j >= 0; j--) + unsigned short stat = 0; + for (signed char j = bitCount - 1; j >= 0; j--) { stat = (stat << 1) | bitVect[j]; } stats[stat]++; } free(dat); - uint32_t count = 0; - for (uint32_t i = 0; i < (1 << bitCount); i++) + unsigned int count = 0; + for (unsigned int i = 0; i < (1 << bitCount); i++) { - uint32_t tmp = i; - for (uint8_t j = 0; j < bitCount; j++) + unsigned int tmp = i; + for (unsigned char j = 0; j < bitCount; j++) { - uint8_t bit = (tmp >> (bitCount - 1) & 1); + unsigned char bit = (tmp >> (bitCount - 1) & 1); if (j == (bitCount - (fullSample - 1) - 1) && bit) { count += stats[i]; diff --git a/src/complex.c b/src/complex.c index 453af0b..54548e3 100644 --- a/src/complex.c +++ b/src/complex.c @@ -377,6 +377,7 @@ void* cpx_ncpx_knk_mtc(void *context) ptrR[((colR * 2) + 1) + (rowR * 2) * (colsR * 2)] = -i1; ptrR[(colR * 2) + ((rowR * 2) + 1) * (colsR * 2)] = i1; ptrR[((colR * 2) + 1) + ((rowR * 2) + 1) * (colsR * 2)] = r1; + } } } diff --git a/src/complex2.c b/src/complex2.c new file mode 100644 index 0000000..0f8fc2c --- /dev/null +++ b/src/complex2.c @@ -0,0 +1,173 @@ +#ifndef __cpx__ +#define __cpx__ +#include +#include +#include +#include +#include +#include "cores.c" + +typedef struct +{ + float real, imaginary; +} cpx_t; + +typedef struct +{ + float *ptr; + int rows, cols; +} cpx_mtx_t; + +uint8_t* cpx_str(cpx_t* n) +{ + uint8_t* r; + int z; + + float rl = n->real; + float ig = n->imaginary >= 0 ? n->imaginary : -(n->imaginary); + if (ig == 0) + { + z = snprintf(NULL, 0, "%f", rl); + r = malloc(z + 1); + sprintf(r, "%f", rl); + + } + else + { + uint8_t op = n->imaginary >= 0 ? '+' : '-'; + z = snprintf(NULL, 0, "%f %c %fi", rl, op, ig); + r = malloc(z + 1); + sprintf(r, "%f %c %fi", rl, op, ig); + } + return r; +} + +cpx_t cpx_new(float r, float i) +{ + cpx_t n; + n.real = r; + n.imaginary = i; + return n; +} + +void cpx_mul(cpx_t* r, cpx_t* a, cpx_t* b) +{ + //FOIL + float first = a->real * b->real; //real + float outer = a->real * b->imaginary; //imaginary + float inner = a->imaginary * b->real; //imaginary + float last = -(a->imaginary * b->imaginary); //real + r->real = first + last; + r->imaginary = outer + inner; +} + +float cpx_magsqr(cpx_t* n) +{ + return (n->real * n->real) + (n->imaginary * n->imaginary); +} + +float cpx_mag(cpx_t* n) +{ + return sqrt((n->real * n->real) + (n->imaginary * n->imaginary)); +} + +void cpx_mtx_set(cpx_mtx_t* m, int row, int col, cpx_t* n) +{ + m->ptr[row * ((m->cols) * 2) + (col * 2)] = n->real; + m->ptr[row * ((m->cols) * 2) + (col * 2) + 1] = n->imaginary; +} + +void cpx_mtx_set2(cpx_mtx_t* m, int row, int col, float real, float imaginary) +{ + m->ptr[row * ((m->cols) * 2) + (col * 2)] = real; + m->ptr[row * ((m->cols) * 2) + (col * 2) + 1] = imaginary; +} + +void cpx_mtx_get(cpx_mtx_t* m, int row, int col, cpx_t* n) +{ + n->real = m->ptr[row * ((m->cols) * 2) + (col * 2)]; + n->imaginary = m->ptr[row * ((m->cols) * 2) + (col * 2) + 1]; +} + +void cpx_mtx_init(cpx_mtx_t* m, int rows, int cols) +{ + int z = rows * (cols * 2) * sizeof(float); + m->ptr = malloc(z); + m->rows = rows; + m->cols = cols; + memset(m->ptr, 0, z); +} + +void cpx_mtx_free(cpx_mtx_t* m) +{ + if (m->ptr != NULL) free(m->ptr); + m->rows = 0; + m->cols = 0; +} + +void cpx_mtx_print(cpx_mtx_t* m) +{ + for (int r = 0; r < m->rows; r++) + { + if (r > 0) printf("\n"); + for (int c = 0; c < m->cols; c++) + { + cpx_t n; + cpx_mtx_get(m, r, c, &n); + uint8_t* s = cpx_str(&n); + if (c > 0) printf(", "); + printf("%s", s); + free(s); + } + } +} + +//This is for testing GPU functions on the CPU +#define __kernel +#define __global +int GPU_GLOBAL_ID_0, GPU_GLOBAL_ID_1, GPU_GLOBAL_ID_2; +int get_global_id(int id) +{ + switch (id) + { + case 0: return GPU_GLOBAL_ID_0; + case 1: return GPU_GLOBAL_ID_1; + case 2: return GPU_GLOBAL_ID_2; + } +} +#include "kernel2.cl" + +void cpx_mtx_dot(float* ptrR, float* ptrA, float* ptrB, size_t rowsA, size_t colsB, size_t shared) +{ + for (int i = 0; i < rowsA; i++) + { + for (int j = 0; j < colsB; j++) + { + GPU_GLOBAL_ID_0 = i; + GPU_GLOBAL_ID_1 = j; + kernel_dot(ptrR, ptrA, ptrB, rowsA, colsB, shared); + } + } +} + +void cpx_mtx_knk(float* ptrR, float* ptrA, float* ptrB, int rowsR, int colsR, int rowsA, int colsA, int rowsB, int colsB) +{ + for (int i = 0; i < rowsR; i++) + { + for (int j = 0; j < colsR; j++) + { + GPU_GLOBAL_ID_0 = i; + GPU_GLOBAL_ID_1 = j; + kernel_knk(ptrR, ptrA, ptrB, rowsR, colsR, rowsA, colsA, rowsB, colsB); + } + } +} + +uint8_t mtx_init() +{ + return 1; +} + +void mtx_clean() {} + +#endif diff --git a/src/gates2.c b/src/gates2.c new file mode 100644 index 0000000..63bbf90 --- /dev/null +++ b/src/gates2.c @@ -0,0 +1,53 @@ + +float Identity[] = +{ + 1, 0, 0, 0, + 0, 0, 1, 0, +}; + +float PauliX[] = +{ + 0, 0, 1, 0, + 1, 0, 0, 0, +}; + +float PauliY[] = +{ + 0, 0, 0, 1, + 0,-1, 0, 0, +}; + +float PauliZ[] = +{ + 1, 0, 0, 0, + 0, 0, -1, 0, +}; + +// 1/sqrt(2) +#define R 0.7071067811865475 +float Hadamard[] = +{ + R, 0, R, 0, + R, 0, -R, 0, +}; + +float PhaseS[] = +{ + 1, 0, 0, 0, + 0, 0, 0,-1, +}; + +// 1/sqrt(2) + 1/sqrt(2)i +float PhaseT[] = +{ + 1, 0, 0, 0, + 0, 0, R,-R, +}; + +float ControlledNOT[] = +{ + 1, 0, 0, 0, 0, 0, 0, 0, + 0, 0, 1, 0, 0, 0, 0, 0, + 0, 0, 0, 0, 0, 0, 1, 0, + 0, 0, 0, 0, 1, 0, 0, 0, +}; \ No newline at end of file diff --git a/src/kernel2.cl b/src/kernel2.cl new file mode 100644 index 0000000..c19c7a5 --- /dev/null +++ b/src/kernel2.cl @@ -0,0 +1,83 @@ +__kernel void kernel_dot +( + __global float* ptrR, + __global float* ptrA, + __global float* ptrB, + const int rowsA, + const int colsB, + const int shared +) +{ + const int colsA = shared; + const int rowsB = shared; + const int rowsR = rowsA; + const int colsR = colsB; + const int rowR = get_global_id(0); + const int colR = get_global_id(1); + int posA, posB; + float rR = 0; + float iR = 0; + + const int posR = rowR * (colsR * 2) + (colR * 2); + + for (int i = 0; i < shared; i++) + { + int posA = rowR * (colsA * 2) + (i * 2); + int posB = i * (colsB * 2) + (colR * 2); + + float rA = ptrA[posA]; + float iA = ptrA[posA + 1]; + float rB = ptrB[posB]; + float iB = ptrB[posB + 1]; + + //(rA + iA)(rB + iB) + float first = rA * rB; + float outer = rA * iB; + float inner = iA * rB; + float lasts = iA * iB; + + rR += first + lasts; + iR += outer + inner; + } + ptrR[rowR * (colsR * 2) + (colR * 2)] = rR; + ptrR[rowR * (colsR * 2) + (colR * 2) + 1] = iR; +} + +__kernel void kernel_knk +( + __global float* ptrR, + __global float* ptrA, + __global float* ptrB, + const int rowsR, + const int colsR, + const int rowsA, + const int colsA, + const int rowsB, + const int colsB +) +{ + int rowR = get_global_id(0); + int colR = get_global_id(1); + + int rowA = rowR / rowsB; + int colA = colR / colsB; + int rowB = rowR % rowsB; + int colB = colR % colsB; + + int posA = rowA * (colsA * 2) + (colA * 2); + int posB = rowB * (colsB * 2) + (colB * 2); + + float rA = ptrA[posA]; + float iA = ptrA[posA + 1]; + float rB = ptrB[posB]; + float iB = ptrB[posB + 1]; + + //(rA + iA)(rB + iB) + float first = rA * rB; + float outer = rA * iB; + float inner = iA * rB; + float lasts = iA * iB; + ptrR[rowR * (colsR * 2) + (colR * 2)] = first + lasts; + ptrR[rowR * (colsR * 2) + (colR * 2) + 1] = outer + inner; +} +