FK20 CUDA
g1test_fft.cu File Reference
#include <stdint.h>
#include <stdio.h>
#include <stdlib.h>
#include <string.h>
#include <unistd.h>
#include "fr.cuh"
#include "g1.cuh"
#include "g1test.cuh"
Include dependency graph for g1test_fft.cu:

Go to the source code of this file.

Macros

#define SET_SHAREDMEM(SZ, FN)
 

Functions

__global__ void g1p_add_wrapper (g1p_t *sum, int count, const g1p_t *x, const g1p_t *y)
 Kernel wrapper for device addition. More...
 
__global__ void g1p_mul_wrapper (g1p_t *q, int count, const fr_t *x, const g1p_t *p)
 Kernel wrapper for device multiplication. More...
 
__global__ void g1p_fr2g1p_wrapper (g1p_t *g1, int count, const fr_t *fr)
 Kernel wrapper for device multiplication computing k*G from k. More...
 
void G1TestFFT (unsigned rows)
 Test for FFT and IFFT of points on the G1 curve. Checks self consistency with the following properties: More...
 

Variables

__managed__ fr_t X [512 *512]
 
__managed__ fr_t Y [512 *512]
 
__managed__ fr_t Z [512 *512]
 
__managed__ g1p_t P [512 *512]
 
__managed__ g1p_t Q [512 *512]
 
__managed__ g1p_t R [512 *512]
 
__managed__ g1p_t S [512 *512]
 
__managed__ g1p_t T [512 *512]
 
__managed__ uint8_t cmp [512 *512]
 

Macro Definition Documentation

◆ SET_SHAREDMEM

#define SET_SHAREDMEM (   SZ,
  FN 
)
Value:
err = cudaFuncSetAttribute(FN, cudaFuncAttributeMaxDynamicSharedMemorySize, SZ); \
cudaDeviceSynchronize(); \
if (err != cudaSuccess) printf("Error cudaFuncSetAttribute: %s:%d, error %d (%s)\n", __FILE__, __LINE__, err, cudaGetErrorName(err));

Definition at line 22 of file g1test_fft.cu.

Function Documentation

◆ g1p_add_wrapper()

__global__ void g1p_add_wrapper ( g1p_t sum,
int  count,
const g1p_t x,
const g1p_t y 
)

Kernel wrapper for device addition.

Parameters
[out]sumResult array
[in]countNumber of array elements
[in]xArray of points on the curve
[in]yArray of points on the curve
Returns
void

Definition at line 38 of file g1test_fft.cu.

Here is the call graph for this function:

◆ g1p_fr2g1p_wrapper()

__global__ void g1p_fr2g1p_wrapper ( g1p_t g1,
int  count,
const fr_t fr 
)

Kernel wrapper for device multiplication computing k*G from k.

Parameters
[out]g1Array of points on the curve
[in]countNumber of array elements
[in]frArray of residues in Fr
Returns
void

Definition at line 99 of file g1test_fft.cu.

◆ g1p_mul_wrapper()

__global__ void g1p_mul_wrapper ( g1p_t q,
int  count,
const fr_t x,
const g1p_t p 
)

Kernel wrapper for device multiplication.

Parameters
[out]qProduct, array of G1 points
[in]countNumber of array elements
[in]xArray of factors
[in]pArray of factors
Returns
void

Definition at line 69 of file g1test_fft.cu.

◆ G1TestFFT()

void G1TestFFT ( unsigned  rows)

Test for FFT and IFFT of points on the G1 curve. Checks self consistency with the following properties:

IFT(FFT(P)) == P FFT(IFT(P)) == P FFT(P+Q) == FFT(P) + FFT(Q) IFT(P+Q) == IFT(P) + IFT(Q) FFT(x*P) == x*FFT(P) IFT(x*P) == x*IFT(P) FFT(G*X) == G*FFT(X) (FFT commutes with mapping from Fr to G1) IFT(G*X) == G*IFT(X) (IFT commutes with mapping from Fr to G1)

Parameters
rows

Definition at line 136 of file g1test_fft.cu.

Here is the call graph for this function:
Here is the caller graph for this function:

Variable Documentation

◆ cmp

__managed__ uint8_t cmp[512 *512]

Definition at line 18 of file g1test_fft.cu.

◆ P

__managed__ g1p_t P[512 *512]

Definition at line 16 of file g1test_fft.cu.

◆ Q

__managed__ g1p_t Q[512 *512]

Definition at line 16 of file g1test_fft.cu.

◆ R

__managed__ g1p_t R[512 *512]

Definition at line 16 of file g1test_fft.cu.

◆ S

__managed__ g1p_t S[512 *512]

Definition at line 16 of file g1test_fft.cu.

◆ T

__managed__ g1p_t T[512 *512]

Definition at line 16 of file g1test_fft.cu.

◆ X

__managed__ fr_t X[512 *512]

Definition at line 15 of file g1test_fft.cu.

◆ Y

__managed__ fr_t Y[512 *512]

Definition at line 15 of file g1test_fft.cu.

◆ Z

__managed__ fr_t Z[512 *512]

Definition at line 15 of file g1test_fft.cu.