commit 816523ac5ec43d452ef3e19b1d29fbdc9ff3ff5a
parent 6fb6e8a74ab8fa368234ab28da2089968fe3263d
Author: Cedric <cedric.zwahlen@students.bfh.ch>
Date: Tue, 21 Nov 2023 13:03:56 +0100
Minor improvements
Diffstat:
5 files changed, 172 insertions(+), 174 deletions(-)
diff --git a/source/rsa-test.c b/source/rsa-test.c
@@ -150,16 +150,12 @@ void generate_random_pairs(DIGIT_T *bases, unsigned long *b_len,
}
// returns how many public keys were read – either 1 or n
-
-int pairs_from_buffers(DIGIT_T *bases, unsigned long *b_len,
+int pairs_from_files(DIGIT_T *bases, unsigned long *b_len,
DIGIT_T *exponents, unsigned long *e_len,
DIGIT_T *moduli, unsigned long *m_len,
DIGIT_T *signatures, unsigned long *s_len,
unsigned int *n) {
-
-
-
FILE *pkfile;
FILE *msfile;
@@ -213,7 +209,7 @@ int pairs_from_buffers(DIGIT_T *bases, unsigned long *b_len,
while (1) {
- char m_buf[2048];
+ char m_buf[2048]; // temp storage, large enough
char s_buf[2048];
if (fscanf(msfile, "%s %s ", m_buf,s_buf) == -1)
@@ -225,7 +221,7 @@ int pairs_from_buffers(DIGIT_T *bases, unsigned long *b_len,
// printf("%s: %lu\n", m_buf, m_buf_len);
// printf("%s: %lu\n", s_buf, s_buf_len);
- DIGIT_T base [sz*2];
+ DIGIT_T base [sz*2]; // temp storage, large enough
DIGIT_T signature [sz*2];
mpSetZero(base, sz*2);
@@ -234,7 +230,7 @@ int pairs_from_buffers(DIGIT_T *bases, unsigned long *b_len,
mpConvFromHex(base, m_buf_len, m_buf);
mpConvFromHex(signature, s_buf_len, s_buf);
- unsigned long max_len = 64;
+ unsigned long max_len = 64; // the maximum of DIGIT_T types we need
b_len[j] = (j == 0 ? 0 : b_len[j - 1]) + max_len;
s_len[j] = (j == 0 ? 0 : s_len[j - 1]) + max_len;
@@ -257,12 +253,39 @@ int pairs_from_buffers(DIGIT_T *bases, unsigned long *b_len,
}
+int number_of_pairs(void) {
+
+ FILE *msfile;
+
+ msfile = fopen("lib-gpu-generate/msgsig.txt", "r");
+
+ if (msfile == NULL) {
+ printf("Auxiliary files not found.");
+ abort();
+ }
+
+ int i = 0;
+
+ while (1) {
+
+ if (fscanf(msfile, "%*s %*s ") == -1)
+ break;
+
+ i++;
+ }
+
+ fclose(msfile);
+
+ return i;
+}
+
int verify_pairs_with_opencl(DIGIT_T *bases, unsigned long *b_len,
DIGIT_T *exponents, unsigned long *e_len,
DIGIT_T *moduli, unsigned long *m_len,
DIGIT_T *signatures, unsigned long *s_len,
const unsigned int n,
- const unsigned int pks) {
+ const unsigned int pks,
+ unsigned long *result) {
int err; // error code returned from api calls
@@ -537,7 +560,7 @@ int verify_pairs_with_opencl(DIGIT_T *bases, unsigned long *b_len,
return EXIT_FAILURE;
}
- printf("KERNEL IS EXECUTING...");
+ printf("KERNEL IS EXECUTING...\n");
// Wait for the command commands to get serviced before reading back results
//
@@ -553,8 +576,7 @@ int verify_pairs_with_opencl(DIGIT_T *bases, unsigned long *b_len,
exit(1);
}
-
- printf("VERIFICATION RESULT: %lu\n",signature_is_valid);
+ *result = signature_is_valid;
// Shutdown and cleanup
//
@@ -577,35 +599,41 @@ int verify_pairs_with_opencl(DIGIT_T *bases, unsigned long *b_len,
}
-
-
int rsa_tests(void) {
setup_gcry();
+ unsigned int pairs = number_of_pairs();
-
- unsigned int gen_n_pairs = 16;
-
- // MARK: must be 0ed first
- DIGIT_T *q = malloc(516 * gen_n_pairs); // does not set memory to 0 on linux, keep that in mind
- DIGIT_T *r = malloc(516 * gen_n_pairs);
- DIGIT_T *s = malloc(516 * gen_n_pairs);
- DIGIT_T *t = malloc(516 * gen_n_pairs);
+ int digit_sz = 64 * pairs * sizeof(DIGIT_T);
+ int arr_sz = pairs * sizeof(unsigned long);
+
+ DIGIT_T *q = malloc(digit_sz);
+ DIGIT_T *r = malloc(digit_sz);
+ DIGIT_T *s = malloc(digit_sz);
+ DIGIT_T *t = malloc(digit_sz);
+ memset(q, 0, digit_sz);
+ memset(r, 0, digit_sz);
+ memset(s, 0, digit_sz);
+ memset(t, 0, digit_sz);
- size_t *u = malloc(gen_n_pairs * sizeof(unsigned long));
- size_t *v = malloc(gen_n_pairs * sizeof(unsigned long));
- size_t *w = malloc(gen_n_pairs * sizeof(unsigned long));
- size_t *x = malloc(gen_n_pairs * sizeof(unsigned long));
+ size_t *u = malloc(arr_sz);
+ size_t *v = malloc(arr_sz);
+ size_t *w = malloc(arr_sz);
+ size_t *x = malloc(arr_sz);
- unsigned int pks = gen_n_pairs;
+ memset(u, 0, arr_sz);
+ memset(v, 0, arr_sz);
+ memset(w, 0, arr_sz);
+ memset(x, 0, arr_sz);
+ unsigned int pks = pairs;
- pks = pairs_from_buffers(q, u,
+ pks = pairs_from_files(q, u,
r, v,
s, w,
- t, x, &gen_n_pairs);
+ t, x, &pairs);
printf("--");
@@ -615,24 +643,38 @@ int rsa_tests(void) {
s, w,
t, x, gen_n_pairs);
*/
+
+ unsigned long result = 0;
+
struct timespec t1, t2;
clock_gettime(CLOCK_REALTIME, &t1);
-
+
verify_pairs_with_opencl(q, u,
- r, v,
- s, w,
- t, x, gen_n_pairs, pks);
+ r, v,
+ s, w,
+ t, x,
+ pairs, pks, &result);
clock_gettime(CLOCK_REALTIME, &t2);
+ printf("VERIFICATION RESULT: %lu\n",result);
- printf("\nGPU verification: %ld ms\n", (t2.tv_nsec - t1.tv_nsec) / 1000000);
+ printf("\nGPU verification took %ld ms\n", (t2.tv_nsec - t1.tv_nsec) / 1000000);
+
+ free(q);
+ free(r);
+ free(s);
+ free(t);
+ free(u);
+ free(v);
+ free(w);
+ free(x);
return 0;
}
-
+/*
static void show_sexp(const char *prefix, gcry_sexp_t a) {
char *buf;
size_t size;
@@ -646,3 +688,4 @@ static void show_sexp(const char *prefix, gcry_sexp_t a) {
fprintf(stderr, "%.*s", (int) size, buf);
gcry_free(buf);
}
+*/
diff --git a/xcode/.DS_Store b/xcode/.DS_Store
Binary files differ.
diff --git a/xcode/lib-gpu-verify.xcodeproj/project.xcworkspace/xcuserdata/cedriczwahlen.xcuserdatad/UserInterfaceState.xcuserstate b/xcode/lib-gpu-verify.xcodeproj/project.xcworkspace/xcuserdata/cedriczwahlen.xcuserdatad/UserInterfaceState.xcuserstate
Binary files differ.
diff --git a/xcode/lib-gpu-verify.xcodeproj/xcuserdata/cedriczwahlen.xcuserdatad/xcdebugger/Breakpoints_v2.xcbkptlist b/xcode/lib-gpu-verify.xcodeproj/xcuserdata/cedriczwahlen.xcuserdatad/xcdebugger/Breakpoints_v2.xcbkptlist
@@ -1279,8 +1279,8 @@
filePath = "../source/rsa-test.c"
startingColumnNumber = "9223372036854775807"
endingColumnNumber = "9223372036854775807"
- startingLineNumber = "588"
- endingLineNumber = "588"
+ startingLineNumber = "606"
+ endingLineNumber = "606"
landmarkName = "rsa_tests()"
landmarkType = "9">
<Locations>
@@ -1462,9 +1462,9 @@
filePath = "../source/rsa-test.c"
startingColumnNumber = "9223372036854775807"
endingColumnNumber = "9223372036854775807"
- startingLineNumber = "571"
- endingLineNumber = "571"
- landmarkName = "verify_pairs_with_opencl(bases, b_len, exponents, e_len, moduli, m_len, signatures, s_len, n, pks)"
+ startingLineNumber = "593"
+ endingLineNumber = "593"
+ landmarkName = "verify_pairs_with_opencl(bases, b_len, exponents, e_len, moduli, m_len, signatures, s_len, n, pks, result)"
landmarkType = "9">
<Locations>
<Location
@@ -1758,9 +1758,9 @@
filePath = "../source/rsa-test.c"
startingColumnNumber = "9223372036854775807"
endingColumnNumber = "9223372036854775807"
- startingLineNumber = "256"
- endingLineNumber = "256"
- landmarkName = "pairs_from_buffers(bases, b_len, exponents, e_len, moduli, m_len, signatures, s_len, n)"
+ startingLineNumber = "252"
+ endingLineNumber = "252"
+ landmarkName = "pairs_from_files(bases, b_len, exponents, e_len, moduli, m_len, signatures, s_len, n)"
landmarkType = "9">
</BreakpointContent>
</BreakpointProxy>
@@ -1774,9 +1774,9 @@
filePath = "../source/rsa-test.c"
startingColumnNumber = "9223372036854775807"
endingColumnNumber = "9223372036854775807"
- startingLineNumber = "214"
- endingLineNumber = "214"
- landmarkName = "pairs_from_buffers(bases, b_len, exponents, e_len, moduli, m_len, signatures, s_len, n)"
+ startingLineNumber = "210"
+ endingLineNumber = "210"
+ landmarkName = "pairs_from_files(bases, b_len, exponents, e_len, moduli, m_len, signatures, s_len, n)"
landmarkType = "9">
<Locations>
<Location
@@ -1867,93 +1867,15 @@
filePath = "../source/rsa-test.c"
startingColumnNumber = "9223372036854775807"
endingColumnNumber = "9223372036854775807"
- startingLineNumber = "210"
- endingLineNumber = "210"
- landmarkName = "pairs_from_buffers(bases, b_len, exponents, e_len, moduli, m_len, signatures, s_len, n)"
+ startingLineNumber = "206"
+ endingLineNumber = "206"
+ landmarkName = "pairs_from_files(bases, b_len, exponents, e_len, moduli, m_len, signatures, s_len, n)"
landmarkType = "9">
</BreakpointContent>
</BreakpointProxy>
<BreakpointProxy
BreakpointExtensionID = "Xcode.Breakpoint.FileBreakpoint">
<BreakpointContent
- uuid = "C1D44B56-001C-4AD8-A710-1D82ADEC89C9"
- shouldBeEnabled = "No"
- ignoreCount = "0"
- continueAfterRunningActions = "No"
- filePath = "../source/rsa-test.c"
- startingColumnNumber = "9223372036854775807"
- endingColumnNumber = "9223372036854775807"
- startingLineNumber = "178"
- endingLineNumber = "178"
- landmarkName = "pairs_from_buffers(bases, b_len, exponents, e_len, moduli, m_len, signatures, s_len, n)"
- landmarkType = "9">
- <Locations>
- <Location
- uuid = "C1D44B56-001C-4AD8-A710-1D82ADEC89C9 - e8b83ddc77e639e1"
- shouldBeEnabled = "Yes"
- ignoreCount = "0"
- continueAfterRunningActions = "No"
- symbolName = "pairs_from_buffers"
- moduleName = "lib-gpu-verify"
- usesParentBreakpointCondition = "Yes"
- urlString = "file:///Users/cedriczwahlen/libgpuverify/source/rsa-test.c"
- startingColumnNumber = "9223372036854775807"
- endingColumnNumber = "9223372036854775807"
- startingLineNumber = "178"
- endingLineNumber = "178"
- offsetFromSymbolStart = "1400">
- </Location>
- <Location
- uuid = "C1D44B56-001C-4AD8-A710-1D82ADEC89C9 - e8b83ddc77e639e1"
- shouldBeEnabled = "Yes"
- ignoreCount = "0"
- continueAfterRunningActions = "No"
- symbolName = "pairs_from_buffers"
- moduleName = "lib-gpu-verify"
- usesParentBreakpointCondition = "Yes"
- urlString = "file:///Users/cedriczwahlen/libgpuverify/source/rsa-test.c"
- startingColumnNumber = "9223372036854775807"
- endingColumnNumber = "9223372036854775807"
- startingLineNumber = "178"
- endingLineNumber = "178"
- offsetFromSymbolStart = "1392">
- </Location>
- <Location
- uuid = "C1D44B56-001C-4AD8-A710-1D82ADEC89C9 - e8b83ddc77e639e1"
- shouldBeEnabled = "Yes"
- ignoreCount = "0"
- continueAfterRunningActions = "No"
- symbolName = "pairs_from_buffers"
- moduleName = "lib-gpu-verify"
- usesParentBreakpointCondition = "Yes"
- urlString = "file:///Users/cedriczwahlen/libgpuverify/source/rsa-test.c"
- startingColumnNumber = "9223372036854775807"
- endingColumnNumber = "9223372036854775807"
- startingLineNumber = "178"
- endingLineNumber = "178"
- offsetFromSymbolStart = "1431">
- </Location>
- <Location
- uuid = "C1D44B56-001C-4AD8-A710-1D82ADEC89C9 - e8b83ddc77e639e1"
- shouldBeEnabled = "Yes"
- ignoreCount = "0"
- continueAfterRunningActions = "No"
- symbolName = "pairs_from_buffers"
- moduleName = "lib-gpu-verify"
- usesParentBreakpointCondition = "Yes"
- urlString = "file:///Users/cedriczwahlen/libgpuverify/source/rsa-test.c"
- startingColumnNumber = "9223372036854775807"
- endingColumnNumber = "9223372036854775807"
- startingLineNumber = "178"
- endingLineNumber = "178"
- offsetFromSymbolStart = "1393">
- </Location>
- </Locations>
- </BreakpointContent>
- </BreakpointProxy>
- <BreakpointProxy
- BreakpointExtensionID = "Xcode.Breakpoint.FileBreakpoint">
- <BreakpointContent
uuid = "F9997565-9649-43A6-92BE-1BBC7B14FA12"
shouldBeEnabled = "No"
ignoreCount = "0"
@@ -1961,9 +1883,9 @@
filePath = "../source/rsa-test.c"
startingColumnNumber = "9223372036854775807"
endingColumnNumber = "9223372036854775807"
- startingLineNumber = "249"
- endingLineNumber = "249"
- landmarkName = "pairs_from_buffers(bases, b_len, exponents, e_len, moduli, m_len, signatures, s_len, n)"
+ startingLineNumber = "245"
+ endingLineNumber = "245"
+ landmarkName = "pairs_from_files(bases, b_len, exponents, e_len, moduli, m_len, signatures, s_len, n)"
landmarkType = "9">
</BreakpointContent>
</BreakpointProxy>
@@ -1977,9 +1899,9 @@
filePath = "../source/rsa-test.c"
startingColumnNumber = "9223372036854775807"
endingColumnNumber = "9223372036854775807"
- startingLineNumber = "201"
- endingLineNumber = "201"
- landmarkName = "pairs_from_buffers(bases, b_len, exponents, e_len, moduli, m_len, signatures, s_len, n)"
+ startingLineNumber = "197"
+ endingLineNumber = "197"
+ landmarkName = "pairs_from_files(bases, b_len, exponents, e_len, moduli, m_len, signatures, s_len, n)"
landmarkType = "9">
</BreakpointContent>
</BreakpointProxy>
@@ -1993,9 +1915,9 @@
filePath = "../source/rsa-test.c"
startingColumnNumber = "9223372036854775807"
endingColumnNumber = "9223372036854775807"
- startingLineNumber = "540"
- endingLineNumber = "540"
- landmarkName = "verify_pairs_with_opencl(bases, b_len, exponents, e_len, moduli, m_len, signatures, s_len, n, pks)"
+ startingLineNumber = "563"
+ endingLineNumber = "563"
+ landmarkName = "verify_pairs_with_opencl(bases, b_len, exponents, e_len, moduli, m_len, signatures, s_len, n, pks, result)"
landmarkType = "9">
</BreakpointContent>
</BreakpointProxy>
@@ -2009,8 +1931,8 @@
filePath = "../source/rsa-test.c"
startingColumnNumber = "9223372036854775807"
endingColumnNumber = "9223372036854775807"
- startingLineNumber = "622"
- endingLineNumber = "622"
+ startingLineNumber = "653"
+ endingLineNumber = "653"
landmarkName = "rsa_tests()"
landmarkType = "9">
</BreakpointContent>
@@ -2025,9 +1947,9 @@
filePath = "../source/rsa-test.c"
startingColumnNumber = "9223372036854775807"
endingColumnNumber = "9223372036854775807"
- startingLineNumber = "549"
- endingLineNumber = "549"
- landmarkName = "verify_pairs_with_opencl(bases, b_len, exponents, e_len, moduli, m_len, signatures, s_len, n, pks)"
+ startingLineNumber = "572"
+ endingLineNumber = "572"
+ landmarkName = "verify_pairs_with_opencl(bases, b_len, exponents, e_len, moduli, m_len, signatures, s_len, n, pks, result)"
landmarkType = "9">
</BreakpointContent>
</BreakpointProxy>
@@ -2041,8 +1963,8 @@
filePath = "../source/rsa-test.c"
startingColumnNumber = "9223372036854775807"
endingColumnNumber = "9223372036854775807"
- startingLineNumber = "610"
- endingLineNumber = "610"
+ startingLineNumber = "638"
+ endingLineNumber = "638"
landmarkName = "rsa_tests()"
landmarkType = "9">
<Locations>
@@ -2089,9 +2011,9 @@
filePath = "../source/rsa-test.c"
startingColumnNumber = "9223372036854775807"
endingColumnNumber = "9223372036854775807"
- startingLineNumber = "283"
- endingLineNumber = "283"
- landmarkName = "verify_pairs_with_opencl(bases, b_len, exponents, e_len, moduli, m_len, signatures, s_len, n, pks)"
+ startingLineNumber = "306"
+ endingLineNumber = "306"
+ landmarkName = "verify_pairs_with_opencl(bases, b_len, exponents, e_len, moduli, m_len, signatures, s_len, n, pks, result)"
landmarkType = "9">
</BreakpointContent>
</BreakpointProxy>
@@ -2105,9 +2027,9 @@
filePath = "../source/rsa-test.c"
startingColumnNumber = "9223372036854775807"
endingColumnNumber = "9223372036854775807"
- startingLineNumber = "244"
- endingLineNumber = "244"
- landmarkName = "pairs_from_buffers(bases, b_len, exponents, e_len, moduli, m_len, signatures, s_len, n)"
+ startingLineNumber = "239"
+ endingLineNumber = "239"
+ landmarkName = "pairs_from_files(bases, b_len, exponents, e_len, moduli, m_len, signatures, s_len, n)"
landmarkType = "9">
</BreakpointContent>
</BreakpointProxy>
@@ -2121,10 +2043,58 @@
filePath = "../source/rsa-test.c"
startingColumnNumber = "9223372036854775807"
endingColumnNumber = "9223372036854775807"
- startingLineNumber = "251"
- endingLineNumber = "251"
- landmarkName = "pairs_from_buffers(bases, b_len, exponents, e_len, moduli, m_len, signatures, s_len, n)"
+ startingLineNumber = "247"
+ endingLineNumber = "247"
+ landmarkName = "pairs_from_files(bases, b_len, exponents, e_len, moduli, m_len, signatures, s_len, n)"
+ landmarkType = "9">
+ </BreakpointContent>
+ </BreakpointProxy>
+ <BreakpointProxy
+ BreakpointExtensionID = "Xcode.Breakpoint.FileBreakpoint">
+ <BreakpointContent
+ uuid = "9303A078-7EDA-414A-8825-97250FC649BC"
+ shouldBeEnabled = "Yes"
+ ignoreCount = "0"
+ continueAfterRunningActions = "No"
+ filePath = "../source/rsa-test.c"
+ startingColumnNumber = "9223372036854775807"
+ endingColumnNumber = "9223372036854775807"
+ startingLineNumber = "611"
+ endingLineNumber = "611"
+ landmarkName = "rsa_tests()"
landmarkType = "9">
+ <Locations>
+ <Location
+ uuid = "9303A078-7EDA-414A-8825-97250FC649BC - b0b9078e770cf765"
+ shouldBeEnabled = "Yes"
+ ignoreCount = "0"
+ continueAfterRunningActions = "No"
+ symbolName = "rsa_tests"
+ moduleName = "lib-gpu-verify"
+ usesParentBreakpointCondition = "Yes"
+ urlString = "file:///Users/cedriczwahlen/libgpuverify/source/rsa-test.c"
+ startingColumnNumber = "9223372036854775807"
+ endingColumnNumber = "9223372036854775807"
+ startingLineNumber = "613"
+ endingLineNumber = "613"
+ offsetFromSymbolStart = "25">
+ </Location>
+ <Location
+ uuid = "9303A078-7EDA-414A-8825-97250FC649BC - b0b9078e770cf727"
+ shouldBeEnabled = "Yes"
+ ignoreCount = "0"
+ continueAfterRunningActions = "No"
+ symbolName = "rsa_tests"
+ moduleName = "lib-gpu-verify"
+ usesParentBreakpointCondition = "Yes"
+ urlString = "file:///Users/cedriczwahlen/libgpuverify/source/rsa-test.c"
+ startingColumnNumber = "9223372036854775807"
+ endingColumnNumber = "9223372036854775807"
+ startingLineNumber = "615"
+ endingLineNumber = "615"
+ offsetFromSymbolStart = "25">
+ </Location>
+ </Locations>
</BreakpointContent>
</BreakpointProxy>
</Breakpoints>
diff --git a/xcode/verify.cl b/xcode/verify.cl
@@ -1188,7 +1188,7 @@ int mpIsZero( const DIGIT_T *a, size_t ndigits)
void assert(bool precondition) {
- char str[] = "assert reached, also this message leaks memory";
+ char str[] = "";
if (!precondition)
mpFail(str);
@@ -1228,11 +1228,6 @@ __kernel void several(__global DIGIT_T* x, __global const unsigned long *s_len,
ndigits = max( max( n_len[index] - (index == 0 ? 0 : n_len[index - 1]) , mm_len[index] - (index == 0 ? 0 : mm_len[index - 1]) ), s_len[index] - (index == 0 ? 0 : s_len[index - 1]) );
edigits = e_len[index] - ( index == 0 ? 0 : e_len[index - 1] );
}
-
-
-
- // int ndigits = 64;
- // int edigits = 1;
// the result is copied in here, compare it to mm
DIGIT_T yout[MAX_ALLOC_SIZE *2];
@@ -1240,7 +1235,7 @@ __kernel void several(__global DIGIT_T* x, __global const unsigned long *s_len,
DIGIT_T mask;
unsigned long n;
- __global DIGIT_T * __private window_x;
+ __global DIGIT_T * __private window_x; // private scope pointers to global memory
__global DIGIT_T * __private window_e;
__global DIGIT_T * __private window_m;
__global DIGIT_T * __private window_mm;
@@ -1255,20 +1250,12 @@ __kernel void several(__global DIGIT_T* x, __global const unsigned long *s_len,
window_x = &x[index == 0 ? 0 : (s_len[index - 1])];
window_mm = &mm[index == 0 ? 0 : (mm_len[index - 1])];
-//
-// window_x = &x[0];
-// window_e = &e[0];
-// window_m = &m[0];
-// window_mm = &mm[0];
-//
+
// can probably be smaller
__private DIGIT_T t1[MAX_ALLOC_SIZE *2];
__private DIGIT_T t2[MAX_ALLOC_SIZE *2];
__private DIGIT_T y[MAX_ALLOC_SIZE *2];
- assert(ndigits <= MAX_FIXED_DIGITS);
- assert(ndigits != 0);
-
n = mpSizeof_g(window_e, edigits);
/* Catch e==0 => x^0=1 */
if (0 == n)
@@ -1309,16 +1296,14 @@ __kernel void several(__global DIGIT_T* x, __global const unsigned long *s_len,
int len = ( mm_len[index] - (index == 0 ? 0 : mm_len[index - 1]) );
-
- // MARK: valid cannot be written to by several at once (the same unit anyway)
- if (mpCompare_lg(yout,window_mm,len) == 0 && index == count - 1) {
- *valid |= 0x1 << index;
+
+ // MARK: I'm NOT 100% positive this is not a race condition
+ if (mpCompare_lg(yout,window_mm,len) == 0) {
+ *valid += 1;
}
}
- // if (index == 8) { *valid = 0xBA; }
-
}
