-
Notifications
You must be signed in to change notification settings - Fork 0
/
interface.cu
378 lines (311 loc) · 10.7 KB
/
interface.cu
1
2
3
4
5
6
7
8
9
10
11
12
13
14
15
16
17
18
19
20
21
22
23
24
25
26
27
28
29
30
31
32
33
34
35
36
37
38
39
40
41
42
43
44
45
46
47
48
49
50
51
52
53
54
55
56
57
58
59
60
61
62
63
64
65
66
67
68
69
70
71
72
73
74
75
76
77
78
79
80
81
82
83
84
85
86
87
88
89
90
91
92
93
94
95
96
97
98
99
100
101
102
103
104
105
106
107
108
109
110
111
112
113
114
115
116
117
118
119
120
121
122
123
124
125
126
127
128
129
130
131
132
133
134
135
136
137
138
139
140
141
142
143
144
145
146
147
148
149
150
151
152
153
154
155
156
157
158
159
160
161
162
163
164
165
166
167
168
169
170
171
172
173
174
175
176
177
178
179
180
181
182
183
184
185
186
187
188
189
190
191
192
193
194
195
196
197
198
199
200
201
202
203
204
205
206
207
208
209
210
211
212
213
214
215
216
217
218
219
220
221
222
223
224
225
226
227
228
229
230
231
232
233
234
235
236
237
238
239
240
241
242
243
244
245
246
247
248
249
250
251
252
253
254
255
256
257
258
259
260
261
262
263
264
265
266
267
268
269
270
271
272
273
274
275
276
277
278
279
280
281
282
283
284
285
286
287
288
289
290
291
292
293
294
295
296
297
298
299
300
301
302
303
304
305
306
307
308
309
310
311
312
313
314
315
316
317
318
319
320
321
322
323
324
325
326
327
328
329
330
331
332
333
334
335
336
337
338
339
340
341
342
343
344
345
346
347
348
349
350
351
352
353
354
355
356
357
358
359
360
361
362
363
364
365
366
367
368
369
370
371
372
373
374
375
376
377
378
#include <iostream>
#include "src/common.hpp"
#include "src/scf.hpp"
#include "src/integrate.hpp"
#include <thrust/host_vector.h>
#include <thrust/device_vector.h>
#include <thrust/inner_product.h>
using namespace std;
using namespace etics;
Integrator IntegratorObj;
// GLOBAL VARIABLES
Real ConstantStep = 0.001953125;
Real T, Step, dT1, dT2, Tcrit;
int N;
extern Real mass;
extern int k3gs, k3bs, k4gs, k4bs;
/*extern*/ Particle *hostP;
thrust::host_vector<Particle> PPP;
/*extern*/ thrust::device_vector<Particle> PPPPP;
// /*extern*/ thrust::device_vector<vec3> F0xxxxxx;
// /*extern*/ thrust::device_vector<Real> PotPotPot; // ugly name
// /*extern*/ thrust::device_vector<vec3> F1xxxxxx;
/*extern*/ vec3 *F1_ptr;
// extern Particle *P_h;
// extern thrust::device_vector<Particle> P;
//
// extern thrust::device_vector<vec3> F0;
// extern thrust::device_vector<Real> Potential;
// extern thrust::device_vector<vec3> F1;
void CommitParticles();
// void InitSCF(int N);
// void ForceSCF(int N, Real *Potential, Particle *PPPPP, vec3 *F);
void DriftStep();
void KickStep();
void CommitForces();
int InitilizeIntegratorMemory();
#define PTR(x) (thrust::raw_pointer_cast((x).data()))
int initialize_code() {
#warning initscf should be here!!! just the problem is that N is requires, so fix it
return 0;
}
int recommit_parameters() {
return 0;
}
int commit_parameters() {
return 0;
}
int new_particle(int *id, double mass, double x, double y, double z, double vx, double vy, double vz, double radius) {
Particle p;
p.m = mass;
p.pos = vec3(x, y, z);
p.vel = vec3(vx, vy, vz);
PPP.push_back(p);
*id = N;
N++;
return 0;
}
int commit_particles() {
// cerr << "calling commit_particles" << endl;
cerr << "we did commit_particles()" << endl;
etics::scf::Init(N, 180, 64, 2605, 384);
#warning hardcoded launch configuration
IntegratorObj = Integrator(&PPP[0], N);
return 0;
}
struct CenterMassFunctor {
Real ConstantMass;
__host__ __device__ CenterMassFunctor(Real _ConstantMass) : ConstantMass(_ConstantMass) {}
__host__ __device__ vec3 operator() (const Particle &p) const {return p.pos;}
};
struct ShiftFunctor {
vec3 Shift;
__host__ __device__ ShiftFunctor(vec3 _Shift) : Shift(_Shift) {}
__host__ __device__ Particle operator() (Particle &p) const {
p.pos += Shift;
p.CalculateR2();
return p;
}
};
int counttt = 0;
bool FirstStep = true;
int evolve_model(double t) {
// PPPPP = PPP;
cerr << "call evolve_model t_end = " << t << " dt = " << t - T << "****************" << endl;
// vec3 CenterMass = thrust::transform_reduce(PPPPP.begin(), PPPPP.end(), CenterMassFunctor(mass), vec3(0,0,0), thrust::plus<vec3>());
// CenterMass = CenterMass * (1.0/N); //ugly should divide by the total mass
// cerr << "CENTER OF MASS " << CenterMass.x << endl;
//
// // thrust::transform(PPPPP.begin(), PPPPP.end(), PPPPP.begin(), ShiftFunctor(-CenterMass));
//
// vec3 CenterMass2 = thrust::transform_reduce(PPPPP.begin(), PPPPP.end(), CenterMassFunctor(mass), vec3(0,0,0), thrust::plus<vec3>());
// CenterMass2 = CenterMass2 * (1.0/N); //ugly should divide by the total mass
// cerr << "CENTER OF MASS after correction " << CenterMass2.x << endl;
//
Step = ConstantStep;
while (T <= t) {
// Take the drift step.
IntegratorObj.DriftStep(Step);
// Calculate the forces in the new positions.
// ForceSCF(N, PTR(PotPotPot), PTR(PPPPP), PTR(F1xxxxxx));
IntegratorObj.CalculateGravity();
// Finish by taking the kick step.
// The kick functor also "commits" the predicted forces into the "acc" member.
IntegratorObj.KickStep(Step);
// N particles were implicitly propagated in this iteration.
// Advance global time.
T += Step;
}
//
// vec3 CenterMass3 = thrust::transform_reduce(PPPPP.begin(), PPPPP.end(), CenterMassFunctor(mass), vec3(0,0,0), thrust::plus<vec3>());
// CenterMass3 = CenterMass3 * (1.0/N); //ugly should divide by the total mass
// cerr << "CENTER OF MASS after evolve " << CenterMass3.x << endl;
//
// cerr << "done evolve; transform" << endl;
// // thrust::transform(PPPPP.begin(), PPPPP.end(), PPPPP.begin(), ShiftFunctor(+CenterMass)); // antishift
//
// vec3 CenterMass4 = thrust::transform_reduce(PPPPP.begin(), PPPPP.end(), CenterMassFunctor(mass), vec3(0,0,0), thrust::plus<vec3>());
// CenterMass4 = CenterMass4 * (1.0/N); //ugly should divide by the total mass
// cerr << "CENTER OF MASS after antishift " << CenterMass4.x << endl;
//
// cerr << "done transform; download to RAM" << endl;
IntegratorObj.CopyParticlesToHost(&PPP[0]);
//
// cerr << "done download; return" << endl;
return 0;
}
int set_begin_time(double time_begin) {
// cerr << "called set_begin_time(" << time_begin << endl;
return 0;
}
int get_begin_time(double *time_begin) {
*time_begin = 0;
return 0;
}
int get_mass(int index_of_the_particle, double *mass) {
*mass = PPP[index_of_the_particle].m;
return 0;
}
int get_time(double *time) {
*time = T;
return 0;
}
int set_mass(int index_of_the_particle, double mass) {
// cerr << "calling set_mass" << endl;
PPP[index_of_the_particle].m = mass;
return 0;
}
int get_index_of_first_particle(int *index_of_the_particle) {
// cerr << "calling get_index_of_first_particle" << endl;
*index_of_the_particle = 0;
return 0;
}
int get_total_radius(double *radius) {
return -2;
}
int get_potential_at_point(double soft, double x, double y, double z, double *phi) {
return -2;
}
int get_total_mass(double *mass) {
return -2;
}
int set_eps2(double epsilon_squared) {
return -1;
}
int get_eps2(double *epsilon_squared) {
*epsilon_squared = 0;
return -1;
}
int get_number_of_particles(int *number_of_particles) {
// cerr << "calling get_number_of_particles" << endl;
*number_of_particles = PPP.size();
return 0;
}
int get_index_of_next_particle(int index_of_the_particle, int *index_of_the_next_particle) {
*index_of_the_next_particle = index_of_the_particle + 1;
return 0;
}
int delete_particle(int index_of_the_particle) {
return -2;
}
int get_potential(int index_of_the_particle, double *potential) {
return -2;
}
int synchronize_model() {
// cerr << "calling synchronize_model" << endl;
return 0;
}
int set_state(int index_of_the_particle, double mass, double radius, double x, double y, double z, double vx, double vy, double vz) {
cerr << "calling set_state" << endl;
// cerr << "calling set_state" << endl;
PPP[index_of_the_particle].pos = vec3(x, y, z);
PPP[index_of_the_particle].vel = vec3(vx, vy, vz);
return 0;
}
int get_state(int index_of_the_particle, double *mass, double *radius, double *x, double *y, double *z, double *vx, double *vy, double *vz) {
// cerr << "calling get_state" << endl;
Particle p = PPP[index_of_the_particle];
*mass = index_of_the_particle;
*x = p.pos.x;
*y = p.pos.y;
*z = p.pos.z;
*vx = p.vel.x;
*vy = p.vel.y;
*vz = p.vel.z;
return 0;
}
int get_time_step(double *time_step) {
// cerr << "calling get_time_step" << endl;
*time_step = ConstantStep;
return 0;
}
int set_time_step(double time_step) {
cerr << "calling set_time_step" << endl;
ConstantStep = time_step;
return 0;
}
int get_launch_config(int **launch_config) {
return 0;
}
int set_launch_config(int *launch_config) {
// k3gs = launch_config[0];
// k3bs = launch_config[1];
// k4gs = launch_config[2];
// k4bs = launch_config[3];
return -2;
}
int recommit_particles() {
// cerr << "calling recommit_particles" << endl;
#warning put something here
cerr << "hhhhhhhhhhhhhhhhhhhhhhhhhhhhhh" << endl;
PPPPP = PPP;
return -2;
}
int set_acceleration(int index_of_the_particle, double ax, double ay, double az) {
return -2;
}
int get_center_of_mass_position(double *x, double *y, double *z) {
// vec3 CenterMass = thrust::transform_reduce(PPPPP.begin(), PPPPP.end(), CenterMassFunctor(mass), vec3(0,0,0), thrust::plus<vec3>());
// CenterMass = CenterMass * (1.0/N); //ugly should divide by the total mass
// *x = CenterMass.x;
// *y = CenterMass.y;
// *z = CenterMass.z;
return 0;
}
int get_center_of_mass_velocity(double *vx, double *vy, double *vz) {
return -2;
}
int get_radius(int index_of_the_particle, double *radius) {
*radius = 0;
return 0;
}
int set_radius(int index_of_the_particle, double radius) {
// should store the radius somewhere but completely ignored by code
// cerr << "calling set_radius" << endl;
return 0;
}
int cleanup_code() {
IntegratorObj.~Integrator();
cerr << "bye" << endl;
return 0;
}
int get_gravity_at_point(double soft, double x, double y, double z, double *forcex, double *forcey, double *forcez) {
return -2;
}
int get_velocity(int index_of_the_particle, double *vx, double *vy, double *vz) {
*vx = PPP[index_of_the_particle].vel.x;
*vy = PPP[index_of_the_particle].vel.y;
*vz = PPP[index_of_the_particle].vel.z;
return 0;
}
int get_position(int index_of_the_particle, double *x, double *y, double *z) {
*x = PPP[index_of_the_particle].pos.x;
*y = PPP[index_of_the_particle].pos.y;
*z = PPP[index_of_the_particle].pos.z;
return 0;
}
bool already_printed = false;
int set_position(int index_of_the_particle, double x, double y, double z) {
if (already_printed == false) {
cerr << "calling set_position" << endl;
cerr << "---------index_of_the_particle=" << index_of_the_particle << endl;
cerr << "--------- x" << PPP[index_of_the_particle].pos.x << "--->" << x << endl;
cerr << "--------- y" << PPP[index_of_the_particle].pos.y << "--->" << y << endl;
cerr << "--------- z" << PPP[index_of_the_particle].pos.z << "--->" << z << endl;
already_printed = true;
}
PPP[index_of_the_particle].pos = vec3(x, y, z);
counttt++;
return 0;
}
int get_acceleration(int index_of_the_particle, double *ax, double *ay, double *az) {
return -2;
}
int set_velocity(int index_of_the_particle, double vx, double vy, double vz) {
// cerr << "calling set_velocity" << endl;
PPP[index_of_the_particle].vel = vec3(vx, vy, vz);
return 0;
}
int get_kinetic_energy(double *kinetic_energy) {
*kinetic_energy = IntegratorObj.KineticEnergy();
return 0;
}
int get_potential_energy(double *potential_energy) {
*potential_energy = IntegratorObj.PotentialEnergy();
return 0;
}
int update_force_potential_arrays(double tttt) {
#warning time shouldnt be a parameter to this one
// ForceSCF(N, PTR(PotPotPot), PTR(PPPPP), PTR(F0xxxxxx));
return 0;
}