Swarm-NG  1.1
gpulog_write.h
1 //
2 // !!! MACHINE GENERATED FILE; DO NOT EDIT !!!
3 // Generated using ../scripts/gen_gpulog_write.pl write
4 //
5 template<typename T1>
6 #ifdef __CUDACC__
7  __device__
8 #endif
9  inline PTR_T(SCALAR(T1)) write(const int recid, const T1 &v1)
10  {
11  typedef internal::pktsize<header, T1> P;
12  //P::dump();
13 
14  // allocate and test for end-of-buffer
15  int len = P::len_with_padding(v1);
16  int at = A::atomicAdd(this->at, len);
17  if(has_overflowed(at + len))
18  {
19  A::atomicAdd(this->at, -len);
20  return NULL;
21  }
22  char *ptr = buffer + at;
23 
24  // write
25  header v0(recid, len);
26  P::IO0::put(ptr, v0, P::begin0, P::len0);
27  P::IO1::put(ptr, v1, P::begin1, P::len1);
28 
29 #if ARGINFO
30  P::store_arginfo(ptr, v1);
31 #endif
32 
33  DHOST( std::cerr << "Total packet len = " << len << "\n"; )
34  return (SCALAR(T1)*)(ptr + P::begin1);
35  }
36 
37 template<typename T1, typename T2>
38 #ifdef __CUDACC__
39  __device__
40 #endif
41  inline PTR_T(SCALAR(T2)) write(const int recid, const T1 &v1, const T2 &v2)
42  {
43  typedef internal::pktsize<header, T1, T2> P;
44  //P::dump();
45 
46  // allocate and test for end-of-buffer
47  int len = P::len_with_padding(v2);
48  int at = A::atomicAdd(this->at, len);
49  if(has_overflowed(at + len))
50  {
51  A::atomicAdd(this->at, -len);
52  return NULL;
53  }
54  char *ptr = buffer + at;
55 
56  // write
57  header v0(recid, len);
58  P::IO0::put(ptr, v0, P::begin0, P::len0);
59  P::IO1::put(ptr, v1, P::begin1, P::len1);
60  P::IO2::put(ptr, v2, P::begin2, P::len2);
61 
62 #if ARGINFO
63  P::store_arginfo(ptr, v2);
64 #endif
65 
66  DHOST( std::cerr << "Total packet len = " << len << "\n"; )
67  return (SCALAR(T2)*)(ptr + P::begin2);
68  }
69 
70 template<typename T1, typename T2, typename T3>
71 #ifdef __CUDACC__
72  __device__
73 #endif
74  inline PTR_T(SCALAR(T3)) write(const int recid, const T1 &v1, const T2 &v2, const T3 &v3)
75  {
76  typedef internal::pktsize<header, T1, T2, T3> P;
77  //P::dump();
78 
79  // allocate and test for end-of-buffer
80  int len = P::len_with_padding(v3);
81  int at = A::atomicAdd(this->at, len);
82  if(has_overflowed(at + len))
83  {
84  A::atomicAdd(this->at, -len);
85  return NULL;
86  }
87  char *ptr = buffer + at;
88 
89  // write
90  header v0(recid, len);
91  P::IO0::put(ptr, v0, P::begin0, P::len0);
92  P::IO1::put(ptr, v1, P::begin1, P::len1);
93  P::IO2::put(ptr, v2, P::begin2, P::len2);
94  P::IO3::put(ptr, v3, P::begin3, P::len3);
95 
96 #if ARGINFO
97  P::store_arginfo(ptr, v3);
98 #endif
99 
100  DHOST( std::cerr << "Total packet len = " << len << "\n"; )
101  return (SCALAR(T3)*)(ptr + P::begin3);
102  }
103 
104 template<typename T1, typename T2, typename T3, typename T4>
105 #ifdef __CUDACC__
106  __device__
107 #endif
108  inline PTR_T(SCALAR(T4)) write(const int recid, const T1 &v1, const T2 &v2, const T3 &v3, const T4 &v4)
109  {
110  typedef internal::pktsize<header, T1, T2, T3, T4> P;
111  //P::dump();
112 
113  // allocate and test for end-of-buffer
114  int len = P::len_with_padding(v4);
115  int at = A::atomicAdd(this->at, len);
116  if(has_overflowed(at + len))
117  {
118  A::atomicAdd(this->at, -len);
119  return NULL;
120  }
121  char *ptr = buffer + at;
122 
123  // write
124  header v0(recid, len);
125  P::IO0::put(ptr, v0, P::begin0, P::len0);
126  P::IO1::put(ptr, v1, P::begin1, P::len1);
127  P::IO2::put(ptr, v2, P::begin2, P::len2);
128  P::IO3::put(ptr, v3, P::begin3, P::len3);
129  P::IO4::put(ptr, v4, P::begin4, P::len4);
130 
131 #if ARGINFO
132  P::store_arginfo(ptr, v4);
133 #endif
134 
135  DHOST( std::cerr << "Total packet len = " << len << "\n"; )
136  return (SCALAR(T4)*)(ptr + P::begin4);
137  }
138 
139 template<typename T1, typename T2, typename T3, typename T4, typename T5>
140 #ifdef __CUDACC__
141  __device__
142 #endif
143  inline PTR_T(SCALAR(T5)) write(const int recid, const T1 &v1, const T2 &v2, const T3 &v3, const T4 &v4, const T5 &v5)
144  {
145  typedef internal::pktsize<header, T1, T2, T3, T4, T5> P;
146  //P::dump();
147 
148  // allocate and test for end-of-buffer
149  int len = P::len_with_padding(v5);
150  int at = A::atomicAdd(this->at, len);
151  if(has_overflowed(at + len))
152  {
153  A::atomicAdd(this->at, -len);
154  return NULL;
155  }
156  char *ptr = buffer + at;
157 
158  // write
159  header v0(recid, len);
160  P::IO0::put(ptr, v0, P::begin0, P::len0);
161  P::IO1::put(ptr, v1, P::begin1, P::len1);
162  P::IO2::put(ptr, v2, P::begin2, P::len2);
163  P::IO3::put(ptr, v3, P::begin3, P::len3);
164  P::IO4::put(ptr, v4, P::begin4, P::len4);
165  P::IO5::put(ptr, v5, P::begin5, P::len5);
166 
167 #if ARGINFO
168  P::store_arginfo(ptr, v5);
169 #endif
170 
171  DHOST( std::cerr << "Total packet len = " << len << "\n"; )
172  return (SCALAR(T5)*)(ptr + P::begin5);
173  }
174 
175 template<typename T1, typename T2, typename T3, typename T4, typename T5, typename T6>
176 #ifdef __CUDACC__
177  __device__
178 #endif
179  inline PTR_T(SCALAR(T6)) write(const int recid, const T1 &v1, const T2 &v2, const T3 &v3, const T4 &v4, const T5 &v5, const T6 &v6)
180  {
181  typedef internal::pktsize<header, T1, T2, T3, T4, T5, T6> P;
182  //P::dump();
183 
184  // allocate and test for end-of-buffer
185  int len = P::len_with_padding(v6);
186  int at = A::atomicAdd(this->at, len);
187  if(has_overflowed(at + len))
188  {
189  A::atomicAdd(this->at, -len);
190  return NULL;
191  }
192  char *ptr = buffer + at;
193 
194  // write
195  header v0(recid, len);
196  P::IO0::put(ptr, v0, P::begin0, P::len0);
197  P::IO1::put(ptr, v1, P::begin1, P::len1);
198  P::IO2::put(ptr, v2, P::begin2, P::len2);
199  P::IO3::put(ptr, v3, P::begin3, P::len3);
200  P::IO4::put(ptr, v4, P::begin4, P::len4);
201  P::IO5::put(ptr, v5, P::begin5, P::len5);
202  P::IO6::put(ptr, v6, P::begin6, P::len6);
203 
204 #if ARGINFO
205  P::store_arginfo(ptr, v6);
206 #endif
207 
208  DHOST( std::cerr << "Total packet len = " << len << "\n"; )
209  return (SCALAR(T6)*)(ptr + P::begin6);
210  }
211 
212 template<typename T1, typename T2, typename T3, typename T4, typename T5, typename T6, typename T7>
213 #ifdef __CUDACC__
214  __device__
215 #endif
216  inline PTR_T(SCALAR(T7)) write(const int recid, const T1 &v1, const T2 &v2, const T3 &v3, const T4 &v4, const T5 &v5, const T6 &v6, const T7 &v7)
217  {
218  typedef internal::pktsize<header, T1, T2, T3, T4, T5, T6, T7> P;
219  //P::dump();
220 
221  // allocate and test for end-of-buffer
222  int len = P::len_with_padding(v7);
223  int at = A::atomicAdd(this->at, len);
224  if(has_overflowed(at + len))
225  {
226  A::atomicAdd(this->at, -len);
227  return NULL;
228  }
229  char *ptr = buffer + at;
230 
231  // write
232  header v0(recid, len);
233  P::IO0::put(ptr, v0, P::begin0, P::len0);
234  P::IO1::put(ptr, v1, P::begin1, P::len1);
235  P::IO2::put(ptr, v2, P::begin2, P::len2);
236  P::IO3::put(ptr, v3, P::begin3, P::len3);
237  P::IO4::put(ptr, v4, P::begin4, P::len4);
238  P::IO5::put(ptr, v5, P::begin5, P::len5);
239  P::IO6::put(ptr, v6, P::begin6, P::len6);
240  P::IO7::put(ptr, v7, P::begin7, P::len7);
241 
242 #if ARGINFO
243  P::store_arginfo(ptr, v7);
244 #endif
245 
246  DHOST( std::cerr << "Total packet len = " << len << "\n"; )
247  return (SCALAR(T7)*)(ptr + P::begin7);
248  }
249 
250 template<typename T1, typename T2, typename T3, typename T4, typename T5, typename T6, typename T7, typename T8>
251 #ifdef __CUDACC__
252  __device__
253 #endif
254  inline PTR_T(SCALAR(T8)) write(const int recid, const T1 &v1, const T2 &v2, const T3 &v3, const T4 &v4, const T5 &v5, const T6 &v6, const T7 &v7, const T8 &v8)
255  {
256  typedef internal::pktsize<header, T1, T2, T3, T4, T5, T6, T7, T8> P;
257  //P::dump();
258 
259  // allocate and test for end-of-buffer
260  int len = P::len_with_padding(v8);
261  int at = A::atomicAdd(this->at, len);
262  if(has_overflowed(at + len))
263  {
264  A::atomicAdd(this->at, -len);
265  return NULL;
266  }
267  char *ptr = buffer + at;
268 
269  // write
270  header v0(recid, len);
271  P::IO0::put(ptr, v0, P::begin0, P::len0);
272  P::IO1::put(ptr, v1, P::begin1, P::len1);
273  P::IO2::put(ptr, v2, P::begin2, P::len2);
274  P::IO3::put(ptr, v3, P::begin3, P::len3);
275  P::IO4::put(ptr, v4, P::begin4, P::len4);
276  P::IO5::put(ptr, v5, P::begin5, P::len5);
277  P::IO6::put(ptr, v6, P::begin6, P::len6);
278  P::IO7::put(ptr, v7, P::begin7, P::len7);
279  P::IO8::put(ptr, v8, P::begin8, P::len8);
280 
281 #if ARGINFO
282  P::store_arginfo(ptr, v8);
283 #endif
284 
285  DHOST( std::cerr << "Total packet len = " << len << "\n"; )
286  return (SCALAR(T8)*)(ptr + P::begin8);
287  }
288 
289 template<typename T1, typename T2, typename T3, typename T4, typename T5, typename T6, typename T7, typename T8, typename T9>
290 #ifdef __CUDACC__
291  __device__
292 #endif
293  inline PTR_T(SCALAR(T9)) write(const int recid, const T1 &v1, const T2 &v2, const T3 &v3, const T4 &v4, const T5 &v5, const T6 &v6, const T7 &v7, const T8 &v8, const T9 &v9)
294  {
295  typedef internal::pktsize<header, T1, T2, T3, T4, T5, T6, T7, T8, T9> P;
296  //P::dump();
297 
298  // allocate and test for end-of-buffer
299  int len = P::len_with_padding(v9);
300  int at = A::atomicAdd(this->at, len);
301  if(has_overflowed(at + len))
302  {
303  A::atomicAdd(this->at, -len);
304  return NULL;
305  }
306  char *ptr = buffer + at;
307 
308  // write
309  header v0(recid, len);
310  P::IO0::put(ptr, v0, P::begin0, P::len0);
311  P::IO1::put(ptr, v1, P::begin1, P::len1);
312  P::IO2::put(ptr, v2, P::begin2, P::len2);
313  P::IO3::put(ptr, v3, P::begin3, P::len3);
314  P::IO4::put(ptr, v4, P::begin4, P::len4);
315  P::IO5::put(ptr, v5, P::begin5, P::len5);
316  P::IO6::put(ptr, v6, P::begin6, P::len6);
317  P::IO7::put(ptr, v7, P::begin7, P::len7);
318  P::IO8::put(ptr, v8, P::begin8, P::len8);
319  P::IO9::put(ptr, v9, P::begin9, P::len9);
320 
321 #if ARGINFO
322  P::store_arginfo(ptr, v9);
323 #endif
324 
325  DHOST( std::cerr << "Total packet len = " << len << "\n"; )
326  return (SCALAR(T9)*)(ptr + P::begin9);
327  }
328 
329 template<typename T1, typename T2, typename T3, typename T4, typename T5, typename T6, typename T7, typename T8, typename T9, typename T10>
330 #ifdef __CUDACC__
331  __device__
332 #endif
333  inline PTR_T(SCALAR(T10)) write(const int recid, const T1 &v1, const T2 &v2, const T3 &v3, const T4 &v4, const T5 &v5, const T6 &v6, const T7 &v7, const T8 &v8, const T9 &v9, const T10 &v10)
334  {
335  typedef internal::pktsize<header, T1, T2, T3, T4, T5, T6, T7, T8, T9, T10> P;
336  //P::dump();
337 
338  // allocate and test for end-of-buffer
339  int len = P::len_with_padding(v10);
340  int at = A::atomicAdd(this->at, len);
341  if(has_overflowed(at + len))
342  {
343  A::atomicAdd(this->at, -len);
344  return NULL;
345  }
346  char *ptr = buffer + at;
347 
348  // write
349  header v0(recid, len);
350  P::IO0::put(ptr, v0, P::begin0, P::len0);
351  P::IO1::put(ptr, v1, P::begin1, P::len1);
352  P::IO2::put(ptr, v2, P::begin2, P::len2);
353  P::IO3::put(ptr, v3, P::begin3, P::len3);
354  P::IO4::put(ptr, v4, P::begin4, P::len4);
355  P::IO5::put(ptr, v5, P::begin5, P::len5);
356  P::IO6::put(ptr, v6, P::begin6, P::len6);
357  P::IO7::put(ptr, v7, P::begin7, P::len7);
358  P::IO8::put(ptr, v8, P::begin8, P::len8);
359  P::IO9::put(ptr, v9, P::begin9, P::len9);
360  P::IO10::put(ptr, v10, P::begin10, P::len10);
361 
362 #if ARGINFO
363  P::store_arginfo(ptr, v10);
364 #endif
365 
366  DHOST( std::cerr << "Total packet len = " << len << "\n"; )
367  return (SCALAR(T10)*)(ptr + P::begin10);
368  }
369