Skip to content

Commit f3956a4

Browse files
committed
More fixes for Windows CUDA examples
For some stupid reason, plot3, plotting and surface CUDA examples are not compiling on WINDOWS stating CUDA kernel has no visiblity to the global constant variables.
1 parent 6354af4 commit f3956a4

File tree

3 files changed

+18
-16
lines changed

3 files changed

+18
-16
lines changed

examples/cuda/plot3.cu

+2-2
Original file line numberDiff line numberDiff line change
@@ -88,7 +88,7 @@ int main(void)
8888

8989

9090
__global__
91-
void gen_curve(float t, float dx, float* out)
91+
void gen_curve(float t, float dx, float* out, const float ZMIN, const size_t ZSIZE)
9292
{
9393
int offset = blockIdx.x * blockDim.x + threadIdx.x;
9494

@@ -110,5 +110,5 @@ void kernel(float t, float dx, float* dev_out)
110110
static const dim3 threads(1024);
111111
dim3 blocks(divup(ZSIZE, 1024));
112112

113-
gen_curve<<< blocks, threads >>>(t, dx, dev_out);
113+
gen_curve<<< blocks, threads >>>(t, dx, dev_out, ZMIN, ZSIZE);
114114
}

examples/cuda/plotting.cu

+6-6
Original file line numberDiff line numberDiff line change
@@ -10,10 +10,10 @@ const unsigned DIMY = 800;
1010
const unsigned WIN_ROWS = 2;
1111
const unsigned WIN_COLS = 2;
1212

13-
const float dx = 0.1;
14-
const float FRANGE_START = 0.f;
15-
const float FRANGE_END = 2 * 3.141592f;
16-
const size_t DATA_SIZE = ( FRANGE_END - FRANGE_START ) / dx;
13+
static const float dx = 0.1;
14+
static const float FRANGE_START = 0.f;
15+
static const float FRANGE_END = 2 * 3.141592f;
16+
static const size_t DATA_SIZE = ( FRANGE_END - FRANGE_START ) / dx;
1717

1818
void kernel(float* dev_out);
1919

@@ -97,7 +97,7 @@ int main(void)
9797

9898

9999
__global__
100-
void simple_sinf(float* out)
100+
void simple_sinf(float* out, const size_t DATA_SIZE, const float dx)
101101
{
102102
int x = blockIdx.x * blockDim.x + threadIdx.x;
103103

@@ -112,5 +112,5 @@ void kernel(float* dev_out)
112112
static const dim3 threads(DATA_SIZE);
113113
dim3 blocks(1);
114114

115-
simple_sinf<<< blocks, threads >>>(dev_out);
115+
simple_sinf << < blocks, threads >> >(dev_out, DATA_SIZE, dx);
116116
}

examples/cuda/surface.cu

+10-8
Original file line numberDiff line numberDiff line change
@@ -8,10 +8,10 @@
88
const unsigned DIMX = 1000;
99
const unsigned DIMY = 800;
1010

11-
static const float XMIN = -1.0f;
12-
static const float XMAX = 2.f;
13-
static const float YMIN = -1.0f;
14-
static const float YMAX = 1.f;
11+
const float XMIN = -1.0f;
12+
const float XMAX = 2.f;
13+
const float YMIN = -1.0f;
14+
const float YMAX = 1.f;
1515

1616
const float DX = 0.01;
1717
const size_t XSIZE = (XMAX-XMIN)/DX+1;
@@ -86,13 +86,15 @@ int main(void)
8686

8787

8888
__global__
89-
void sincos_surf(float t, float dx, float* out)
89+
void sincos_surf(float t, float dx, float* out,
90+
const float XMIN, const float YMIN,
91+
const size_t XSIZE, const size_t YSIZE)
9092
{
9193
int i = blockIdx.x * blockDim.x + threadIdx.x;
9294
int j = blockIdx.y * blockDim.y + threadIdx.y;
9395

94-
float x=XMIN+i*dx;
95-
float y=YMIN+j*dx;
96+
float x= ::XMIN + i*dx;
97+
float y= ::YMIN + j*dx;
9698
if (i<XSIZE && j<YSIZE) {
9799
int offset = j + i * YSIZE;
98100
out[ 3 * offset ] = x;
@@ -112,5 +114,5 @@ void kernel(float t, float dx, float* dev_out)
112114
dim3 blocks(divup(XSIZE, threads.x),
113115
divup(YSIZE, threads.y));
114116

115-
sincos_surf<<< blocks, threads >>>(t, dx, dev_out);
117+
sincos_surf<<< blocks, threads >>>(t, dx, dev_out, XMIN, YMIN, XSIZE, YSIZE);
116118
}

0 commit comments

Comments
 (0)