1 2014-11-05 Thomas Schwinge <thomas@codesourcery.com>
3 * testsuite/libgomp.oacc-c/cache-1.c: Remove directives that are
4 expected to fail, and rename the file to...
5 * testsuite/libgomp.oacc-c-c++-common/cache-1.c: ... this.
6 * testsuite/libgomp.oacc-c++/cache-1.C: Remove file.
8 2014-11-05 James Norris <jnorris@codesourcery.com>
10 * testsuite/libgomp.oacc-c/cache-1.c: New file.
11 * testsuite/libgomp.oacc-c++/cache-1.C: Likewise.
13 2014-11-05 Thomas Schwinge <thomas@codesourcery.com>
14 James Norris <jnorris@codesourcery.com>
16 * testsuite/libgomp.oacc-c++/c++.exp: Enable
17 libgomp.oacc-c-c++-common testing.
18 * testsuite/libgomp.oacc-c/c.exp: Likewise.
19 * testsuite/libgomp.oacc-c/abort-2.c: Rename to...
20 * testsuite/libgomp.oacc-c-c++-common/abort-2.c: ... this.
21 * testsuite/libgomp.oacc-c/abort.c: Rename to...
22 * testsuite/libgomp.oacc-c-c++-common/abort.c: ... this.
23 * testsuite/libgomp.oacc-c/acc_on_device-1.c: Rename to...
24 * testsuite/libgomp.oacc-c-c++-common/acc_on_device-1.c: ... this.
25 * testsuite/libgomp.oacc-c/clauses-1.c: Rename to...
26 * testsuite/libgomp.oacc-c-c++-common/clauses-1.c: ... this.
27 * testsuite/libgomp.oacc-c/clauses-2.c: Rename to...
28 * testsuite/libgomp.oacc-c-c++-common/clauses-2.c: ... this.
29 * testsuite/libgomp.oacc-c/context-1.c: Rename to...
30 * testsuite/libgomp.oacc-c-c++-common/context-1.c: ... this.
31 * testsuite/libgomp.oacc-c/context-2.c: Rename to...
32 * testsuite/libgomp.oacc-c-c++-common/context-2.c: ... this.
33 * testsuite/libgomp.oacc-c/context-3.c: Rename to...
34 * testsuite/libgomp.oacc-c-c++-common/context-3.c: ... this.
35 * testsuite/libgomp.oacc-c/context-4.c: Rename to...
36 * testsuite/libgomp.oacc-c-c++-common/context-4.c: ... this.
37 * testsuite/libgomp.oacc-c/data-1.c: Rename to...
38 * testsuite/libgomp.oacc-c-c++-common/data-1.c: ... this.
39 * testsuite/libgomp.oacc-c/data-2.c: Rename to...
40 * testsuite/libgomp.oacc-c-c++-common/data-2.c: ... this.
41 * testsuite/libgomp.oacc-c/data-3.c: Rename to...
42 * testsuite/libgomp.oacc-c-c++-common/data-3.c: ... this.
43 * testsuite/libgomp.oacc-c/deviceptr-1.c: Rename to...
44 * testsuite/libgomp.oacc-c-c++-common/deviceptr-1.c: ... this.
45 * testsuite/libgomp.oacc-c/if-1.c: Rename to...
46 * testsuite/libgomp.oacc-c-c++-common/if-1.c: ... this.
47 * testsuite/libgomp.oacc-c/kernels-1.c: Rename to...
48 * testsuite/libgomp.oacc-c-c++-common/kernels-1.c: ... this.
49 * testsuite/libgomp.oacc-c/lib-1.c: Rename to...
50 * testsuite/libgomp.oacc-c-c++-common/lib-1.c: ... this.
51 * testsuite/libgomp.oacc-c/lib-10.c: Rename to...
52 * testsuite/libgomp.oacc-c-c++-common/lib-10.c: ... this.
53 * testsuite/libgomp.oacc-c/lib-11.c: Rename to...
54 * testsuite/libgomp.oacc-c-c++-common/lib-11.c: ... this.
55 * testsuite/libgomp.oacc-c/lib-12.c: Rename to...
56 * testsuite/libgomp.oacc-c-c++-common/lib-12.c: ... this.
57 * testsuite/libgomp.oacc-c/lib-13.c: Rename to...
58 * testsuite/libgomp.oacc-c-c++-common/lib-13.c: ... this.
59 * testsuite/libgomp.oacc-c/lib-14.c: Rename to...
60 * testsuite/libgomp.oacc-c-c++-common/lib-14.c: ... this.
61 * testsuite/libgomp.oacc-c/lib-15.c: Rename to...
62 * testsuite/libgomp.oacc-c-c++-common/lib-15.c: ... this.
63 * testsuite/libgomp.oacc-c/lib-16.c: Rename to...
64 * testsuite/libgomp.oacc-c-c++-common/lib-16.c: ... this.
65 * testsuite/libgomp.oacc-c/lib-17.c: Rename to...
66 * testsuite/libgomp.oacc-c-c++-common/lib-17.c: ... this.
67 * testsuite/libgomp.oacc-c/lib-18.c: Rename to...
68 * testsuite/libgomp.oacc-c-c++-common/lib-18.c: ... this.
69 * testsuite/libgomp.oacc-c/lib-19.c: Rename to...
70 * testsuite/libgomp.oacc-c-c++-common/lib-19.c: ... this.
71 * testsuite/libgomp.oacc-c/lib-2.c: Rename to...
72 * testsuite/libgomp.oacc-c-c++-common/lib-2.c: ... this.
73 * testsuite/libgomp.oacc-c/lib-20.c: Rename to...
74 * testsuite/libgomp.oacc-c-c++-common/lib-20.c: ... this.
75 * testsuite/libgomp.oacc-c/lib-21.c: Rename to...
76 * testsuite/libgomp.oacc-c-c++-common/lib-21.c: ... this.
77 * testsuite/libgomp.oacc-c/lib-22.c: Rename to...
78 * testsuite/libgomp.oacc-c-c++-common/lib-22.c: ... this.
79 * testsuite/libgomp.oacc-c/lib-23.c: Rename to...
80 * testsuite/libgomp.oacc-c-c++-common/lib-23.c: ... this.
81 * testsuite/libgomp.oacc-c/lib-24.c: Rename to...
82 * testsuite/libgomp.oacc-c-c++-common/lib-24.c: ... this.
83 * testsuite/libgomp.oacc-c/lib-25.c: Rename to...
84 * testsuite/libgomp.oacc-c-c++-common/lib-25.c: ... this.
85 * testsuite/libgomp.oacc-c/lib-26.c: Rename to...
86 * testsuite/libgomp.oacc-c-c++-common/lib-26.c: ... this.
87 * testsuite/libgomp.oacc-c/lib-27.c: Rename to...
88 * testsuite/libgomp.oacc-c-c++-common/lib-27.c: ... this.
89 * testsuite/libgomp.oacc-c/lib-28.c: Rename to...
90 * testsuite/libgomp.oacc-c-c++-common/lib-28.c: ... this.
91 * testsuite/libgomp.oacc-c/lib-29.c: Rename to...
92 * testsuite/libgomp.oacc-c-c++-common/lib-29.c: ... this.
93 * testsuite/libgomp.oacc-c/lib-3.c: Rename to...
94 * testsuite/libgomp.oacc-c-c++-common/lib-3.c: ... this.
95 * testsuite/libgomp.oacc-c/lib-30.c: Rename to...
96 * testsuite/libgomp.oacc-c-c++-common/lib-30.c: ... this.
97 * testsuite/libgomp.oacc-c/lib-31.c: Rename to...
98 * testsuite/libgomp.oacc-c-c++-common/lib-31.c: ... this.
99 * testsuite/libgomp.oacc-c/lib-32.c: Rename to...
100 * testsuite/libgomp.oacc-c-c++-common/lib-32.c: ... this.
101 * testsuite/libgomp.oacc-c/lib-33.c: Rename to...
102 * testsuite/libgomp.oacc-c-c++-common/lib-33.c: ... this.
103 * testsuite/libgomp.oacc-c/lib-34.c: Rename to...
104 * testsuite/libgomp.oacc-c-c++-common/lib-34.c: ... this.
105 * testsuite/libgomp.oacc-c/lib-35.c: Rename to...
106 * testsuite/libgomp.oacc-c-c++-common/lib-35.c: ... this.
107 * testsuite/libgomp.oacc-c/lib-36.c: Rename to...
108 * testsuite/libgomp.oacc-c-c++-common/lib-36.c: ... this.
109 * testsuite/libgomp.oacc-c/lib-37.c: Rename to...
110 * testsuite/libgomp.oacc-c-c++-common/lib-37.c: ... this.
111 * testsuite/libgomp.oacc-c/lib-38.c: Rename to...
112 * testsuite/libgomp.oacc-c-c++-common/lib-38.c: ... this.
113 * testsuite/libgomp.oacc-c/lib-39.c: Rename to...
114 * testsuite/libgomp.oacc-c-c++-common/lib-39.c: ... this.
115 * testsuite/libgomp.oacc-c/lib-4.c: Rename to...
116 * testsuite/libgomp.oacc-c-c++-common/lib-4.c: ... this.
117 * testsuite/libgomp.oacc-c/lib-40.c: Rename to...
118 * testsuite/libgomp.oacc-c-c++-common/lib-40.c: ... this.
119 * testsuite/libgomp.oacc-c/lib-41.c: Rename to...
120 * testsuite/libgomp.oacc-c-c++-common/lib-41.c: ... this.
121 * testsuite/libgomp.oacc-c/lib-42.c: Rename to...
122 * testsuite/libgomp.oacc-c-c++-common/lib-42.c: ... this.
123 * testsuite/libgomp.oacc-c/lib-43.c: Rename to...
124 * testsuite/libgomp.oacc-c-c++-common/lib-43.c: ... this.
125 * testsuite/libgomp.oacc-c/lib-44.c: Rename to...
126 * testsuite/libgomp.oacc-c-c++-common/lib-44.c: ... this.
127 * testsuite/libgomp.oacc-c/lib-45.c: Rename to...
128 * testsuite/libgomp.oacc-c-c++-common/lib-45.c: ... this.
129 * testsuite/libgomp.oacc-c/lib-46.c: Rename to...
130 * testsuite/libgomp.oacc-c-c++-common/lib-46.c: ... this.
131 * testsuite/libgomp.oacc-c/lib-47.c: Rename to...
132 * testsuite/libgomp.oacc-c-c++-common/lib-47.c: ... this.
133 * testsuite/libgomp.oacc-c/lib-48.c: Rename to...
134 * testsuite/libgomp.oacc-c-c++-common/lib-48.c: ... this.
135 * testsuite/libgomp.oacc-c/lib-49.c: Rename to...
136 * testsuite/libgomp.oacc-c-c++-common/lib-49.c: ... this.
137 * testsuite/libgomp.oacc-c/lib-5.c: Rename to...
138 * testsuite/libgomp.oacc-c-c++-common/lib-5.c: ... this.
139 * testsuite/libgomp.oacc-c/lib-50.c: Rename to...
140 * testsuite/libgomp.oacc-c-c++-common/lib-50.c: ... this.
141 * testsuite/libgomp.oacc-c/lib-51.c: Rename to...
142 * testsuite/libgomp.oacc-c-c++-common/lib-51.c: ... this.
143 * testsuite/libgomp.oacc-c/lib-52.c: Rename to...
144 * testsuite/libgomp.oacc-c-c++-common/lib-52.c: ... this.
145 * testsuite/libgomp.oacc-c/lib-53.c: Rename to...
146 * testsuite/libgomp.oacc-c-c++-common/lib-53.c: ... this.
147 * testsuite/libgomp.oacc-c/lib-54.c: Rename to...
148 * testsuite/libgomp.oacc-c-c++-common/lib-54.c: ... this.
149 * testsuite/libgomp.oacc-c/lib-55.c: Rename to...
150 * testsuite/libgomp.oacc-c-c++-common/lib-55.c: ... this.
151 * testsuite/libgomp.oacc-c/lib-56.c: Rename to...
152 * testsuite/libgomp.oacc-c-c++-common/lib-56.c: ... this.
153 * testsuite/libgomp.oacc-c/lib-57.c: Rename to...
154 * testsuite/libgomp.oacc-c-c++-common/lib-57.c: ... this.
155 * testsuite/libgomp.oacc-c/lib-58.c: Rename to...
156 * testsuite/libgomp.oacc-c-c++-common/lib-58.c: ... this.
157 * testsuite/libgomp.oacc-c/lib-59.c: Rename to...
158 * testsuite/libgomp.oacc-c-c++-common/lib-59.c: ... this.
159 * testsuite/libgomp.oacc-c/lib-6.c: Rename to...
160 * testsuite/libgomp.oacc-c-c++-common/lib-6.c: ... this.
161 * testsuite/libgomp.oacc-c/lib-60.c: Rename to...
162 * testsuite/libgomp.oacc-c-c++-common/lib-60.c: ... this.
163 * testsuite/libgomp.oacc-c/lib-61.c: Rename to...
164 * testsuite/libgomp.oacc-c-c++-common/lib-61.c: ... this.
165 * testsuite/libgomp.oacc-c/lib-62.c: Rename to...
166 * testsuite/libgomp.oacc-c-c++-common/lib-62.c: ... this.
167 * testsuite/libgomp.oacc-c/lib-63.c: Rename to...
168 * testsuite/libgomp.oacc-c-c++-common/lib-63.c: ... this.
169 * testsuite/libgomp.oacc-c/lib-64.c: Rename to...
170 * testsuite/libgomp.oacc-c-c++-common/lib-64.c: ... this.
171 * testsuite/libgomp.oacc-c/lib-65.c: Rename to...
172 * testsuite/libgomp.oacc-c-c++-common/lib-65.c: ... this.
173 * testsuite/libgomp.oacc-c/lib-66.c: Rename to...
174 * testsuite/libgomp.oacc-c-c++-common/lib-66.c: ... this.
175 * testsuite/libgomp.oacc-c/lib-67.c: Rename to...
176 * testsuite/libgomp.oacc-c-c++-common/lib-67.c: ... this.
177 * testsuite/libgomp.oacc-c/lib-68.c: Rename to...
178 * testsuite/libgomp.oacc-c-c++-common/lib-68.c: ... this.
179 * testsuite/libgomp.oacc-c/lib-69.c: Rename to...
180 * testsuite/libgomp.oacc-c-c++-common/lib-69.c: ... this.
181 * testsuite/libgomp.oacc-c/lib-7.c: Rename to...
182 * testsuite/libgomp.oacc-c-c++-common/lib-7.c: ... this.
183 * testsuite/libgomp.oacc-c/lib-70.c: Rename to...
184 * testsuite/libgomp.oacc-c-c++-common/lib-70.c: ... this.
185 * testsuite/libgomp.oacc-c/lib-71.c: Rename to...
186 * testsuite/libgomp.oacc-c-c++-common/lib-71.c: ... this.
187 * testsuite/libgomp.oacc-c/lib-72.c: Rename to...
188 * testsuite/libgomp.oacc-c-c++-common/lib-72.c: ... this.
189 * testsuite/libgomp.oacc-c/lib-73.c: Rename to...
190 * testsuite/libgomp.oacc-c-c++-common/lib-73.c: ... this.
191 * testsuite/libgomp.oacc-c/lib-74.c: Rename to...
192 * testsuite/libgomp.oacc-c-c++-common/lib-74.c: ... this.
193 * testsuite/libgomp.oacc-c/lib-75.c: Rename to...
194 * testsuite/libgomp.oacc-c-c++-common/lib-75.c: ... this.
195 * testsuite/libgomp.oacc-c/lib-76.c: Rename to...
196 * testsuite/libgomp.oacc-c-c++-common/lib-76.c: ... this.
197 * testsuite/libgomp.oacc-c/lib-77.c: Rename to...
198 * testsuite/libgomp.oacc-c-c++-common/lib-77.c: ... this.
199 * testsuite/libgomp.oacc-c/lib-78.c: Rename to...
200 * testsuite/libgomp.oacc-c-c++-common/lib-78.c: ... this.
201 * testsuite/libgomp.oacc-c/lib-79.c: Rename to...
202 * testsuite/libgomp.oacc-c-c++-common/lib-79.c: ... this.
203 * testsuite/libgomp.oacc-c/lib-80.c: Rename to...
204 * testsuite/libgomp.oacc-c-c++-common/lib-80.c: ... this.
205 * testsuite/libgomp.oacc-c/lib-81.c: Rename to...
206 * testsuite/libgomp.oacc-c-c++-common/lib-81.c: ... this.
207 * testsuite/libgomp.oacc-c/lib-82.c: Rename to...
208 * testsuite/libgomp.oacc-c-c++-common/lib-82.c: ... this.
209 * testsuite/libgomp.oacc-c/lib-83.c: Rename to...
210 * testsuite/libgomp.oacc-c-c++-common/lib-83.c: ... this.
211 * testsuite/libgomp.oacc-c/lib-84.c: Rename to...
212 * testsuite/libgomp.oacc-c-c++-common/lib-84.c: ... this.
213 * testsuite/libgomp.oacc-c/lib-85.c: Rename to...
214 * testsuite/libgomp.oacc-c-c++-common/lib-85.c: ... this.
215 * testsuite/libgomp.oacc-c/lib-86.c: Rename to...
216 * testsuite/libgomp.oacc-c-c++-common/lib-86.c: ... this.
217 * testsuite/libgomp.oacc-c/lib-87.c: Rename to...
218 * testsuite/libgomp.oacc-c-c++-common/lib-87.c: ... this.
219 * testsuite/libgomp.oacc-c/lib-88.c: Rename to...
220 * testsuite/libgomp.oacc-c-c++-common/lib-88.c: ... this.
221 * testsuite/libgomp.oacc-c/lib-89.c: Rename to...
222 * testsuite/libgomp.oacc-c-c++-common/lib-89.c: ... this.
223 * testsuite/libgomp.oacc-c/lib-9.c: Rename to...
224 * testsuite/libgomp.oacc-c-c++-common/lib-9.c: ... this.
225 * testsuite/libgomp.oacc-c/lib-90.c: Rename to...
226 * testsuite/libgomp.oacc-c-c++-common/lib-90.c: ... this.
227 * testsuite/libgomp.oacc-c/lib-91.c: Rename to...
228 * testsuite/libgomp.oacc-c-c++-common/lib-91.c: ... this.
229 * testsuite/libgomp.oacc-c/lib-92.c: Rename to...
230 * testsuite/libgomp.oacc-c-c++-common/lib-92.c: ... this.
231 * testsuite/libgomp.oacc-c/nested-1.c: Rename to...
232 * testsuite/libgomp.oacc-c-c++-common/nested-1.c: ... this.
233 * testsuite/libgomp.oacc-c/nested-2.c: Rename to...
234 * testsuite/libgomp.oacc-c-c++-common/nested-2.c: ... this.
235 * testsuite/libgomp.oacc-c/offset-1.c: Rename to...
236 * testsuite/libgomp.oacc-c-c++-common/offset-1.c: ... this.
237 * testsuite/libgomp.oacc-c/parallel-1.c: Rename to...
238 * testsuite/libgomp.oacc-c-c++-common/parallel-1.c: ... this.
239 * testsuite/libgomp.oacc-c/pointer-align-1.c: Rename to...
240 * testsuite/libgomp.oacc-c-c++-common/pointer-align-1.c: ... this.
241 * testsuite/libgomp.oacc-c/present-1.c: Rename to...
242 * testsuite/libgomp.oacc-c-c++-common/present-1.c: ... this.
243 * testsuite/libgomp.oacc-c/present-2.c: Rename to...
244 * testsuite/libgomp.oacc-c-c++-common/present-2.c: ... this.
245 * testsuite/libgomp.oacc-c/reduction-1.c: Rename to...
246 * testsuite/libgomp.oacc-c-c++-common/reduction-1.c: ... this.
247 * testsuite/libgomp.oacc-c/reduction-2.c: Rename to...
248 * testsuite/libgomp.oacc-c-c++-common/reduction-2.c: ... this.
249 * testsuite/libgomp.oacc-c/reduction-3.c: Rename to...
250 * testsuite/libgomp.oacc-c-c++-common/reduction-3.c: ... this.
251 * testsuite/libgomp.oacc-c/reduction-4.c: Rename to...
252 * testsuite/libgomp.oacc-c-c++-common/reduction-4.c: ... this.
253 * testsuite/libgomp.oacc-c/reduction-5.c: Rename to...
254 * testsuite/libgomp.oacc-c-c++-common/reduction-5.c: ... this.
255 * testsuite/libgomp.oacc-c/reduction-initial-1.c: Rename to...
256 * testsuite/libgomp.oacc-c-c++-common/reduction-initial-1.c: ... this.
257 * testsuite/libgomp.oacc-c/subr.cu: Rename to...
258 * testsuite/libgomp.oacc-c-c++-common/subr.cu: ... this.
259 * testsuite/libgomp.oacc-c/subr.ptx: Rename to...
260 * testsuite/libgomp.oacc-c-c++-common/subr.ptx: ... this.
261 * testsuite/libgomp.oacc-c/timer.h: Rename to...
262 * testsuite/libgomp.oacc-c-c++-common/timer.h: ... this.
263 * testsuite/libgomp.oacc-c/update-1.c: Rename to...
264 * testsuite/libgomp.oacc-c-c++-common/update-1.c: ... this.
266 * libgomp.texi: Update for OpenACC.
268 * target.h (struct ACC_dispatch_t): Remove avail_func. Update all
270 * oacc-host.c (openacc_avail): Likewise.
271 * plugin-nvptx.c (PTX_avail, openacc_avail): Likewise.
272 * libgomp.texi (acc_get_num_devices): Revise.
274 * testsuite/libgomp.oacc-c/goacc_kernels.c: Remove file.
275 * testsuite/libgomp.oacc-c/goacc_parallel.c: Remove file.
277 2014-11-04 Cesar Philippidis <cesar@codesourcery.com>
279 * testsuite/libgomp.oacc-fortran/routine-1.f90: New test.
280 * testsuite/libgomp.oacc-fortran/routine-2.f90: New test.
281 * testsuite/libgomp.oacc-fortran/routine-3.f90: New test.
282 * testsuite/libgomp.oacc-fortran/routine-4.f90: New test.
284 2014-11-04 Thomas Schwinge <thomas@codesourcery.com>
286 * testsuite/libgomp.oacc-c/reduction-initial-1.c
287 [ACC_DEVICE_TYPE_host_nonshm]: Remove special-casing for N.
289 2014-11-03 Cesar Philippidis <cesar@codesourcery.com>
290 Thomas Schwinge <thomas@codesourcery.com>
292 * libgomp.map (GOACC_get_thread_num): Declare.
293 (GOACC_get_num_threads): Declare.
294 * libgomp_g.h (GOACC_get_thread_num): Declare.
295 (GOACC_get_num_threads): Declare.
296 * oacc-parallel.c (GOACC_parallel): Handle num_gangs.
297 (GOACC_get_num_threads): New function.
298 (GOACC_get_thread_num): New function.
299 * oacc-ptx.h: New file.
300 * plugin-nvptx.c (ABORT_PTX): Remove macro. Move to oacc-ptx.h.
301 (ACC_ON_DEVICE_PTX): Likewise.
302 (link_ptx): Also link ptx code defined by GOACC_INTERNAL_PTX.
303 (PTX_exec): Handle gangs/CTAs.
304 * testsuite/libgomp.oacc-c/reduction-1.c: New test.
305 * testsuite/libgomp.oacc-c/reduction-2.c: New test.
306 * testsuite/libgomp.oacc-c/reduction-3.c: New test.
307 * testsuite/libgomp.oacc-c/reduction-4.c: New test.
308 * testsuite/libgomp.oacc-c/reduction-5.: New test.
309 * testsuite/libgomp.oacc-c/reduction-initial-1.c: New test.
310 * testsuite/libgomp.oacc-fortran/reduction-1.f90: New test.
311 * testsuite/libgomp.oacc-fortran/reduction-2.f90: New test.
312 * testsuite/libgomp.oacc-fortran/reduction-3.f90: New test.
313 * testsuite/libgomp.oacc-fortran/reduction-4.f90: New test.
314 * testsuite/libgomp.oacc-fortran/reduction-5.f90: New test.
315 * testsuite/libgomp.oacc-fortran/reduction-6.f90: New test.
317 2014-11-03 Cesar Philippidis <cesar@codesourcery.com>
319 * libgomp.map (GOACC_enter_exit_data): Declare as global.
320 * libgomp_g.h (GOACC_enter_exit_data): Declare.
321 (GOACC_update): Declare.
322 (gomp_acc_insert_pointer): Declare.
323 (gomp_acc_remove_pointer): Declare.
324 * oacc-mem.c (gomp_acc_insert_pointer): New function.
325 (gomp_acc_remove_pointer): New function.
326 * oacc-parallel.c (find_pset): New function.
327 (GOACC_enter_exit_data): New function.
328 (GOACC_update): Handle GOMP_MAP_TO_PSET.
329 * testsuite/libgomp.oacc-c/data-2.c: New test.
330 * testsuite/libgomp.oacc-c/data-3.c: New test.
331 * testsuite/libgomp.oacc-fortran/data-1.f90: New test.
332 * testsuite/libgomp.oacc-fortran/data-2.f90: New test.
333 * testsuite/libgomp.oacc-fortran/data-3.f90: New test.
334 * testsuite/libgomp.oacc-fortran/data-4.f90: New test.
336 2014-10-31 Cesar Philippidis <cesar@codesourcery.com>
338 * testsuite/libgomp.oacc-fortran/asyncwait-1.f90: New test.
339 * testsuite/libgomp.oacc-fortran/asyncwait-2.f90: New test.
340 * testsuite/libgomp.oacc-fortran/asyncwait-3.f90: New test.
342 2014-10-31 Thomas Schwinge <thomas@codesourcery.com>
344 * configure.ac (CUDA_DRIVER_INCLUDE, CUDA_DRIVER_LIB): New
345 substituted variables.
346 (testsuite/libgomp-test-support.exp): Instantiate file.
347 * Makefile.in: Regenerate.
348 * configure: Likewise.
349 * testsuite/Makefile.in: Likewise.
350 * testsuite/libgomp-test-support.exp.in: New file.
351 * testsuite/lib/libgomp.exp: Load it.
352 (libgomp_init): For build-tree testing, add the CUDA paths to
353 ALWAYS_CFLAGS and always_ld_library_path.
355 2014-10-30 Julian Brown <julian@codesourcery.com>
357 * target.c (gomp_map_vars): Remove MM argument.
358 (GOMP_target, GOMP_target_data): Update calls to gomp_map_vars.
359 * oacc-mem.c (acc_map_data, present_create_copy): Update calls to
361 * oacc-parallel.c (GOACC_parallel, GOACC_data_start): Likewise.
362 * target.h (gomp_map_vars): Update prototype.
364 2014-10-30 Julian Brown <julian@codesourcery.com>
366 * oacc-host.c (DEBUG): Remove undefine.
367 * plugin-nvptx.c (DEBUG, DISABLE_ASYNC): Remove comment-out macro
369 * target.c (dump_mappings): Remove debugging function.
371 2014-10-30 Julian Brown <julian@codesourcery.com>
373 * target.c (gomp_load_plugin_for_device): Only call get_caps once.
374 (gomp_find_available_plugins): ...and don't call it again here.
376 2014-10-30 Julian Brown <julian@codesourcery.com>
378 * env.c (goacc_notify_var): New.
379 (initialize_env): Use above instead of
380 gomp_global_icv.acc_notify_var.
381 * error.c (gomp_vnotify): Use goacc_notify_var.
382 (gomp_notify): Fix formatting.
383 * libgomp.h (gomp_task_icv): Remove acc_notify_var field.
384 (goacc_notify_var): Add extern declaration.
386 2014-10-30 Julian Brown <julian@codesourcery.com>
388 * env.c (goacc_parse_device_num): Remove.
389 (initialize_env): Use parse_int instead of goacc_parse_device_num.
390 * oacc-init.c (lazy_open): Add bounds check for device number.
392 2014-10-30 Julian Brown <julian@codesourcery.com>
394 * env.c (oacc-int.h): Include.
395 * libgomp-plugin.h (GOMP_PLUGIN_async_unmap_vars): Don't declare
397 * libgomp.map (PLUGIN_1.0): Add GOMP_PLUGIN_acc_thread symbol.
398 * oacc-async.c (oacc-int.h): Include.
399 (acc_async_test, acc_async_test_all, acc_wait, acc_wait_async)
400 (acc_wait_all, acc_wait_all_async): Use base_dev not ACC_dev.
401 * oacc-cuda.c (oacc-int.h): Include.
402 (acc_get_current_cuda_device, acc_get_current_cuda_context)
403 (acc_get_cuda_stream): Use base_dev not ACC_dev.
404 (acc_set_cuda_stream): Call ACC_lazy_initialize, use base_dev not
406 * oacc-host.c (oacc-plugin.h, oacc-int.h): Include one or the
407 other, depending on HOST_NONSHM_PLUGIN setting.
408 (openacc_create_thread_data, openacc_destroy_thread_data): New.
409 (host_dispatch): Initialize create_thread_data_func,
410 destroy_thread_data_func hooks.
411 * oacc-init.c (oacc-int.h): Include.
412 (acc_device_lock): Make static.
413 (ACC_dev, ACC_handle, handle_num, struct ACC_context, ACC_contexts):
415 (base_dev, goacc_threads, goacc_thread_lock, goacc_cleanup_key):
417 (goacc_tls_data, goacc_tls_key): Define one or the other, depending
419 (dispatchers): Add comment.
420 (ACC_register): Only register 0th device.
421 (close_handle): Remove.
422 (_acc_init): Add comment. Don't use ACC_dev.
423 (goacc_new_thread, goacc_destroy_thread): New.
424 (lazy_open): Update comment. Remove code relating to contexts and
425 memory maps. Create thread-specific data if necessary.
426 (acc_init): Use base_dev not ACC_dev.
427 (_acc_shutdown): Don't call close_handle. Use base_dev not ACC_dev.
428 Free target-specific TLS data and close devices that are open at
429 the time of shutdown.
430 (lazy_init): Add comment. Use base_dev not ACC_dev.
431 (lazy_init_and_open): Add comment. Use base_dev not ACC_dev.
432 (acc_get_num_devices, acc_get_device_type, acc_get_device_num): Use
433 base_dev not ACC_dev.
434 (acc_set_device_num): Likewise. If switching device number,
435 de-associate previous device number from the current thread, but do
436 not close the device.
437 (acc_on_device): Check type for device for current thread, not
439 (ACC_runtime_initialize): Initialize goacc_tls_key,
440 goacc_cleanup_key, base_dev, goacc_threads, goacc_thread_lock. Don't
441 initialize ACC_contexts.
442 (saved_bound_dev): Remove.
443 (ACC_save_and_set_bind, ACC_restore_bind): Use goacc_thread.
444 (ACC_lazy_initialize): Use base_dev, goacc_thread instead of
446 * oacc-int.h (struct ACC_dispatch_t): Don't declare here.
447 (struct goacc_thread): New.
448 (goacc_tls_data): Define for HAVE_TLS.
449 (goacc_tls_key): Define for !HAVE_TLS.
450 (goacc_thread): New (versions with and without HAVE_TLS).
451 (ACC_mem_open, ACC_mem_close, ACC_resolve_device, ACC_dev)
452 (ACC_handle, struct memmap_t, ACC_memmap): Remove.
453 (base_dev): Add extern declaration.
454 * oacc-mem.c (oacc-int.h, assert.h): Include.
455 (ACC_memmap, ACC_mem_open, ACC_mem_close): Remove.
456 (lookup_host): Change memmap_t argument to struct
458 (lookup_dev): Change memmap_t argument to struct target_mem_desc.
459 Use list_count not refcount for iterating over mapped elements.
460 (acc_malloc): Use base_dev not ACC_dev.
461 (acc_free): Update call to lookup_dev. Use base_dev not ACC_dev.
462 (acc_memcpy_to_device, acc_memcpy_from_device): Use base_dev not
464 (acc_deviceptr, acc_is_present): Update call to lookup_host.
465 (acc_hostptr): Update call to lookup_dev.
466 (acc_map_data): Look up thread device instead of using ACC_dev,
467 update calls to lookup_host, lookup_dev. Use data environment in
469 (acc_unmap_data): Update call to lookup_host. Remove mapped data
470 from data environment not ACC_memmap.
471 (present_create_copy): Update call to lookup_host. Use data
472 environment instead of list in ACC_memmap.
473 (delete_copyout): Update call to lookup_host. Look up device in
474 current thread info instead of using ACC_dev.
475 (update_dev_host): Look up device in current thread info instead of
477 * oacc-parallel.c (oacc-int.h): Include.
478 (struct devgeom, devgeom, dump_devaddrs): Remove.
479 (select_acc_device): Call ACC_lazy_initialize earlier.
480 (GOACC_parallel): Use device for current thread instead of ACC_dev.
481 Use memory map from current device.
482 (GOACC_data_start): Likewise. Use thread info block for mapped data.
483 (GOACC_data_end): Use thread info block for mapped data.
484 (goacc_wait): Use device for current thread instead of ACC_dev.
485 (GOACC_update): Likewise. Formatting fixes.
486 * oacc-plugin.c (ACC_plugin_register): Remove.
487 (oacc-int.h): Include.
488 (GOMP_PLUGIN_acc_thread): New.
489 * oacc-plugin.h (target.h): Don't include.
490 (ACC_plugin_register): Remove.
491 (GOMP_PLUGIN_async_unmap_vars, GOMP_PLUGIN_acc_thread): Add extern
493 * plugin-nvptx.c (oacc-plugin.h): Include.
494 (current_stream, PTX_dev, PTX_devices): Remove.
495 (struct nvptx_thread): New.
496 (nvptx_thread): New function.
497 (select_stream_for_async): Locate ptx_dev in device-specific TLS
498 data instead of using TLS PTX_dev variable.
499 (PTX_init): Don't initialize PTX_devices.
500 (PTX_open_device): Remove PTX_devices list handling. Tweak context
502 (PTX_close_device): Remove PTX_devices list handling. Find PTX
503 device info via function argument instead of global TLS variable.
504 (PTX_get_num_devices): Make callable when backend has not been
506 (event_gc): Find PTX device info, current stream via nvptx_thread.
507 (event_add, PTX_exec, PTX_host2dev, PTX_dev2host)
508 (PTX_async_test_all, PTX_wait_all, PTX_wait_all_async)
509 (PTX_get_current_cuda_device, PTX_get_current_cuda_context)
510 (PTX_get_cuda_stream, PTX_set_cuda_stream, openacc_close_device)
511 (openacc_set_device_num, openacc_register_async_cleanup)
512 (openacc_async_set_async): Likewise.
513 (openacc_create_thread_data, openacc_destroy_thread_data): New.
514 * target.c (oacc-int.h): Include.
515 (gomp_fini_device): Split out memory-map freeing into...
516 (gomp_free_memmap): ...this new function.
517 (gomp_load_plugin_for_device): Initialize
518 openacc.create_thread_data_func, openacc.destroy_thread_data_func
520 (gomp_find_available_plugins): Initialize one target_device_descr
522 * target.h (oacc-int.h): Don't include.
523 (ACC_dispatch_t): Declare here. Add data_environ, ord fields.
524 Update comment for mem_map field.
525 (gomp_free_memmap): Add prototype.
527 2014-10-28 Thomas Schwinge <thomas@codesourcery.com>
529 * testsuite/libgomp.oacc-c/c.exp (ld_library_path, ALWAYS_CFLAGS):
530 Don't refer to CUDA installation in /opt/nvidia/cuda-5.5/.
532 * oacc-init.c: Don't use <sys/queue.h>'s SLIST_*.
533 * plugin-nvptx.c: Likewise.
535 2014-10-23 Thomas Schwinge <thomas@codesourcery.com>
537 * testsuite/libgomp.oacc-c/reduction-initial-1.c: New file.
539 2014-10-20 Cesar Philippidis <cesar@codesourcery.com>
541 * (GOACC_update): Declare.
543 2014-10-20 Julian Brown <julian@codesourcery.com>
545 * oacc-host.c (openacc_register_async_cleanup): New.
546 (host_dispatch): Initialise register_async_cleanup_func entry.
547 * oacc-int.h (struct ACC_dispatch_t): Add
548 register_async_cleanup_func hook.
549 * oacc-parallel.c (GOACC_parallel): Call
550 register_async_cleanup_func hook after queuing asynchronous
552 * plugin-nvptx.c (enum PTX_event_type): Add PTX_EVT_ASYNC_CLEANUP.
553 (struct PTX_event): Remove tgt field.
554 (event_gc): Don't do async cleanup in PTX_EVT_KNL, do it in
555 PTX_EVT_ASYNC_CLEANUP instead.
556 (event_add): Remove tgt argument. Support PTX_EVT_ASYNC_CLEANUP
558 (PTX_exec, PTX_host2dev, PTX_dev2host, PTX_wait_async)
559 (PTX_wait_all_async): Update calls to event_add.
560 (openacc_register_async_cleanup): New.
561 (PTX_async_test): Call event_gc on success path.
562 (PTX_async_test_all): Likewise.
563 * target.c (gomp_load_plugin_for_device): Initialise
564 register_async_cleanup hook.
566 2014-10-20 Julian Brown <julian@codesourcery.com>
568 * Makefile.am (AM_CPPFLAGS): Fix search path for locating
570 * Makefile.in: Regenerate.
572 2014-10-20 Julian Brown <julian@codesourcery.com>
574 * libgomp-plugin.c (gomp_plugin_*): Rename to...
575 (GOMP_PLUGIN_*): This.
576 * libgomp-plugin.h: Likewise.
577 * libgomp.map: Likewise.
578 * oacc-host.c (GOMP): Use GOMP_PLUGIN_ in macro expansion.
579 * oacc-plugin.c (gomp_plugin_*): Rename to...
580 (GOMP_PLUGIN_*): This.
581 * plugin-nvptx.c: Likewise.
583 2014-10-17 Thomas Schwinge <thomas@codesourcery.com>
584 James Norris <jnorris@codesourcery.com>
586 * libgomp.texi: Outline documentation for OpenACC.
588 2014-10-17 James Norris <jnorris@codesourcery.com>
589 Thomas Schwinge <thomas@codesourcery.com>
590 Tom de Vries <tom@codesourcery.com>
591 Cesar Philippidis <cesar@codesourcery.com>
593 * testsuite/Makefile.in: Regenerated.
594 * testsuite/lib/libgomp.exp
595 (check_effective_target_openacc_nvidia_accel_present)
596 (check_effective_target_openacc_nvidia_accel_selected): New
598 * testsuite/libgomp.oacc-fortran/fortran.exp: New exp file.
599 * testsuite/libgomp.oacc-fortran/*.f: New tests.
600 * testsuite/libgomp.oacc-fortran/*.f90: Likewise.
601 * testsuite/libgomp.oacc-c/c.exp: New exp file.
602 * testsuite/libgomp.oacc-c/subr.ptx: New file.
603 * testsuite/libgomp.oacc-c/subr.cu: New file.
604 * testsuite/libgomp.oacc-c/timer.h: New file.
605 * testsuite/libgomp.oacc-c/*.c: New tests.
606 * testsuite/libgomp.oacc-c++/c++.exp: New exp file.
608 2014-10-17 Nathan Sidwell <nathan@codesourcery.com>
609 James Norris <jnorris@codesourcery.com>
610 Thomas Schwinge <thomas@codesourcery.com>
611 Tom de Vries <tom@codesourcery.com>
612 Julian Brown <julian@codesourcery.com>
614 * Makefile.am (AM_CPPFLAGS): Search in ../include also.
615 (libgomp_plugin_nvptx_version_info, libgomp_plugin_nvptx_la_SOURCES)
616 (libgomp_plugin_nvptx_la_CPPFLAGS, libgomp_plugin_nvptx_la_LDFLAGS)
617 (libgomp_plugin_nvptx_la_LIBADD)
618 (libgomp_plugin_nvptx_la_LIBTOOLFLAGS): Set variables if
619 PLUGIN_NVPTX is defined.
620 (toolexeclib_LTLIBRARIES): Add nonshm-host
621 and (conditionally) nvidia plugins.
622 (libgomp_plugin_nonshm_host_version_info)
623 (libgomp_plugin_nonshm_host_la_SOURCES)
624 (libgomp_plugin_nonshm_host_la_CPPFLAGS)
625 (libgomp_plugin_nonshm_host_la_LDFLAGS)
626 (libgomp_plugin_nonshm_host_la_LIBTOOLFLAGS): Set variables.
627 (libgomp_la_SOURCES): Add oacc-parallel.c, splay-tree.c,
628 oacc-host.c, oacc-init.c, oacc-mem.c, oacc-async.c, oacc-plugin.c,
629 oacc-cuda.c, libgomp-plugin.c.
630 (nodist_libsubinclude_HEADERS): Add openacc.h,
631 ../include/gomp-constants.h.
632 * Makefile.in: Regenerate.
633 * config.h.in: Regenerate.
634 * configure.ac: Add TODOs for OpenACC in various places.
635 (CUDA_DRIVER_CPPFLAGS, CUDA_DRIVER_LDFLAGS): Initialize.
636 (--with-cuda-driver, --with-cuda-driver-include)
637 (--with-cuda-driver-lib, --enable-offload-targets): Implement new
639 (PLUGIN_NVPTX, PLUGIN_NVPTX_CPPFLAGS, PLUGIN_NVPTX_LDFLAGS)
640 (PLUGIN_NVPTX_LIBS): Initialize variables.
641 * configure: Regenerate.
642 * env.c (target.h): Include.
643 (goacc_device_num, goacc_device_type): New globals.
644 (goacc_parse_device_num, goacc_parse_device_type): New functions.
645 (initialize_env): Parse GCC_ACC_NOTIFY, ACC_DEVICE_TYPE,
646 ACC_DEVICE_NUM environment variables.
647 * error.c (gomp_verror, gomp_vfatal, gomp_vnotify, gomp_notify):
649 (gomp_fatal): Make global.
650 * libgomp.h (stdarg.h): Include.
651 (struct gomp_memory_mapping): Forward declaration.
652 (struct gomp_task_icv): Add acc_notify_var member.
653 (goacc_device_num, goacc_device_type): Add extern declarations.
654 (gomp_vnotify, gomp_notify, gomp_verror, gomp_vfatal): Add
656 (gomp_init_targets_once): Add prototype.
657 * libgomp.map (OACC_2.0): New symbol version. Add public acc_*
659 (PLUGIN_1.0): New symbol version. Add gomp plugin interface
661 * libgomp_g.h (GOACC_kernels, GOACC_parallel): Update prototypes.
662 (GOACC_wait): Add prototype.
663 * target.c (limits.h, stdbool.h, stdlib.h): Don't include.
664 (oacc-plugin.h, gomp-constants.h, stdio.h, assert.h): Include.
665 (splay_tree_node, splay_tree, splay_tree_key, target_mem_desc)
666 (splay_tree_key_s, enum target_type, gomp_device_descr): Don't
668 (splay-tree.h): Include.
670 (splay_compare): Change linkage to hidden not static.
671 (gomp_init_targets_once): New function.
672 (gomp_get_num_devices): Use above.
673 (dump_mappings): New function (for debugging).
674 (get_kind): New function.
675 (gomp_map_vars): Add gomp_memory_mapping (mm), is_openacc
676 parameters. Change KINDS to void *. Use lock from memory map
677 not device. Use macros from gomp-constants.h instead of
678 hard-coded values. Support OpenACC-specific mappings.
679 (gomp_copy_from_async): New function.
680 (gomp_unmap_vars): Add DO_COPYFROM argument. Only copy memory
681 back from device if it is true. Use lock from memory map not
683 (gomp_update): Add mm, is_openacc args. Use lock from
684 memory map not device. Use macros from gomp-constants.h not
686 (gomp_register_image_for_device): Add forward
688 (GOMP_offload_register): Change TARGET_DATA type to
689 void **. Check realloc result.
690 (gomp_init_device): Change linkage to hidden not static. Tweak mem
692 (gomp_fini_device): New function.
693 (GOMP_target): Adjust lazy initialization, check target
694 capabilities for OpenMP 4.0 support. Add locking around splay tree
695 lookup. Add new arg to gomp_unmap_vars call.
696 (GOMP_target_data): Tweak lazy initialization. Add new args to
697 gomp_map_vars, gomp_unmap_vars calls.
698 (GOMP_target_update): Tweak lazy initialization. Add new args to
700 (gomp_load_plugin_for_device): Initialize device_fini and
701 OpenACC-specific plugin hooks.
702 (gomp_register_images_for_device): Rename to...
703 (gomp_register_image_for_device): This, and register a single
704 device only, and only if it has not already had images registered.
705 (gomp_find_available_plugins): Rearrange to fix plugin loading and
706 initialization for OpenACC. Prefer a device with
707 TARGET_CAP_OPENMP_400 for OpenMP.
708 * target.h: New file.
709 * splay-tree.h: Move bulk of implementation to...
710 * splay-tree.c: New file.
711 * libgomp-plugin.c: New file.
712 * libgomp-plugin.h: New file.
713 * oacc-async.c: New file.
714 * oacc-cuda.c: New file.
715 * oacc-host.c: New file.
716 * oacc-init.c: New file.
717 * oacc-mem.c: New file.
718 * oacc-parallel.c: New file.
719 * oacc-plugin.c: New file.
720 * plugin-nvptx.c: New file.
721 * oacc-int.h: New file.
722 * openacc.f90: New file.
723 * openacc.h: New file.
724 * openacc_lib.h: New file.
726 2014-10-09 Thomas Schwinge <thomas@codesourcery.com>
728 * testsuite/libgomp.oacc-c/collapse-4.c: New file, moved from
729 gcc/testsuite/gcc.dg/goacc/collapse.c.
731 2014-09-23 Thomas Schwinge <thomas@codesourcery.com>
733 * libgomp.map (OACC_2.0): Add acc_on_device, acc_on_device_.
734 * fortran.c: Include "openacc.h".
735 (acc_on_device_): New function.
736 * oacc-parallel.c: Include "openacc.h".
737 (acc_on_device): New function.
738 * openacc.f90 (acc_device_kind, acc_device_none)
739 (acc_device_default, acc_device_host, acc_device_not_host): New
741 (acc_on_device): New function declaration.
742 * openacc_lib.h (acc_device_kind, acc_device_none)
743 (acc_device_default, acc_device_host, acc_device_not_host): New
745 (acc_on_device): New function declaration.
746 * openacc.h (acc_device_t): New enum.
747 (acc_on_device): New function declaration.
748 * testsuite/libgomp.oacc-c/acc_on_device-1.c: New file.
749 * testsuite/libgomp.oacc-fortran/acc_on_device-1-1.f90: Likewise.
750 * testsuite/libgomp.oacc-fortran/acc_on_device-1-2.f: Likewise.
751 * testsuite/libgomp.oacc-fortran/acc_on_device-1-3.f: Likewise.
753 2014-07-09 Thomas Schwinge <thomas@codesourcery.com>
754 Jakub Jelinek <jakub@redhat.com>
756 * target.c (gomp_map_vars, gomp_unmap_vars, gomp_update): Support
757 NULL mappings as well as mapping kind OMP_CLAUSE_MAP_TO_PSET.
758 Also, some code reformatting.
760 2014-06-06 Thomas Schwinge <thomas@codesourcery.com>
761 James Norris <jnorris@codesourcery.com>
763 * libgomp.map (GOACC_2.0): Add GOACC_update.
764 * oacc-parallel.c (GOACC_update): New function.
766 2014-03-18 Ilya Verbin <ilya.verbin@intel.com>
768 * libgomp.map (GOMP_4.0.1): New symbol version.
769 Add GOMP_offload_register.
770 * plugin-host.c (device_available): Replace with:
771 (get_num_devices): This.
773 (offload_register): Ditto.
774 (device_init): Ditto.
775 (device_get_table): Ditto.
777 * target.c (target_type): New enum.
778 (offload_image_descr): New struct.
779 (offload_images, num_offload_images): New globals.
780 (struct gomp_device_descr): Remove device_available_func.
781 Add type, is_initialized, get_type_func, get_num_devices_func,
782 offload_register_func, device_init_func, device_get_table_func,
784 (mapping_table): New struct.
785 (GOMP_offload_register): New function.
786 (gomp_init_device): Ditto.
787 (GOMP_target): Add device initialization and lookup for target fn.
788 (GOMP_target_data): Add device initialization.
789 (GOMP_target_update): Ditto.
790 (gomp_load_plugin_for_device): Take handles for get_type,
791 get_num_devices, offload_register, device_init, device_get_table,
792 device_run functions.
793 (gomp_register_images_for_device): New function.
794 (gomp_find_available_plugins): Add registration of offload images.
796 2014-03-18 Thomas Schwinge <thomas@codesourcery.com>
798 * libgomp_g.h (GOACC_kernels, GOACC_parallel): Add three
799 additional int arguments.
800 * oacc-parallel.c (GOACC_kernels, GOACC_parallel): Handle these.
801 * testsuite/libgomp.oacc-c/goacc_kernels.c: Adjust.
802 * testsuite/libgomp.oacc-c/goacc_parallel.c: Likewise.
804 2014-02-28 Thomas Schwinge <thomas@codesourcery.com>
806 * testsuite/libgomp.oacc-c/goacc_kernels.c: New file.
807 * testsuite/libgomp.oacc-c/kernels-1.c: Likewise.
808 * testsuite/libgomp.oacc-c/parallel-1.c: Add one missing test.
810 * libgomp.map (GOACC_2.0): Add GOACC_kernels.
811 * libgomp_g.h (GOACC_kernels): New prototype.
812 * oacc-parallel.c (GOACC_kernels): New function.
814 2014-02-21 Thomas Schwinge <thomas@codesourcery.com>
816 * testsuite/libgomp.oacc-c/data-1.c: New file.
818 * libgomp.map (GOACC_2.0): Add GOACC_data_end, GOACC_data_start.
819 * libgomp_g.h (GOACC_data_start, GOACC_data_end): New prototypes.
820 * oacc-parallel.c (GOACC_data_start, GOACC_data_end): New
823 2014-02-20 Thomas Schwinge <thomas@codesourcery.com>
825 * target.c (gomp_load_plugin_for_device): Don't call dlcose if
828 2014-02-20 Thomas Schwinge <thomas@codesourcery.com>
829 James Norris <jnorris@codesourcery.com>
831 * plugin-host.c: New file.
832 * target.c (struct gomp_device_descr): Add device_alloc_func,
833 device_free_func, device_dev2host_func, device_host2dev_func
835 (gomp_load_plugin_for_device): Load these.
836 (gomp_map_vars, gomp_unmap_tgt, gomp_unmap_vars, gomp_update): Use
838 (resolve_device, gomp_find_available_plugins): Remove ID 257 hack.
840 2014-02-17 Ilya Verbin <ilya.verbin@intel.com>
842 * target.c (gomp_map_vars_existing): Don't update copy_from for the
845 2014-01-28 Thomas Schwinge <thomas@codesourcery.com>
847 * testsuite/libgomp.oacc-c/parallel-1.c: Extend.
849 2013-12-18 Thomas Schwinge <thomas@codesourcery.com>
851 * libgomp_g.h (GOACC_parallel): Switch kinds array to unsigned
853 * oacc-parallel.c (GOACC_parallel): Likewise, and catch
856 2013-11-08 Thomas Schwinge <thomas@codesourcery.com>
859 * target.c (gomp_find_available_plugins): Don't skip device 257
862 2013-11-07 Thomas Schwinge <thomas@codesourcery.com>
864 * oacc-parallel.c: New file.
865 * Makefile.am (libgomp_la_SOURCES): Add it.
866 * Makefile.in: Regenerate.
867 * libgomp.map (GOACC_2.0): Add GOACC_parallel.
868 * libgomp_g.h (GOACC_parallel): New declaration.
869 * testsuite/libgomp.oacc-c/goacc_parallel.c: New file.
870 * testsuite/libgomp.oacc-c/parallel-1.c: New file.
872 * libgomp.map (GOACC_2.0): New symbol version.
874 * openacc.f90 (openacc_version): New integer parameter.
875 * openacc_lib.h (openacc_version): Likewise.
876 * testsuite/libgomp.oacc-fortran/openacc_version-1.f: New file.
877 * testsuite/libgomp.oacc-fortran/openacc_version-2.f90: Likewise.
879 * libgomp.map (OACC_2.0): New symbol version.
880 * Makefile.am (nodist_libsubinclude_HEADERS): Add openacc.h.
881 (nodist_finclude_HEADERS): Add openacc_lib.h, openacc.f90,
882 openacc.mod, and openacc_kinds.mod.
883 (openacc_kinds.mod): New target.
884 (%.mod): New target, generalized from omp_lib.mod.
885 * Makefile.in: Regenerate.
886 * openacc.f90: New file.
887 * openacc.h: Likewise.
888 * openacc_lib.h: Likewise.
889 * testsuite/libgomp.oacc-c++/c++.exp: Likewise.
890 * testsuite/libgomp.oacc-c/c.exp: Likewise.
891 * testsuite/libgomp.oacc-c/lib-1.c: Likewise.
892 * testsuite/libgomp.oacc-fortran/fortran.exp: Likewise.
893 * testsuite/libgomp.oacc-fortran/lib-1.f90: Likewise.
894 * testsuite/libgomp.oacc-fortran/lib-2.f: Likewise.
895 * testsuite/libgomp.oacc-fortran/lib-3.f: Likewise.
897 Backport from trunk, r204519:
899 * testsuite/lib/libgomp.exp (libgomp_init): Don't add -fopenmp to
901 * testsuite/libgomp.c++/c++.exp (ALWAYS_CFLAGS): Add -fopenmp.
902 * testsuite/libgomp.c/c.exp (ALWAYS_CFLAGS): Likewise.
903 * testsuite/libgomp.fortran/fortran.exp (ALWAYS_CFLAGS): Likewise.
904 * testsuite/libgomp.graphite/graphite.exp (ALWAYS_CFLAGS):
907 Backport from trunk, r204518:
909 * libgomp_g.h: Include <stddef.h> for size_t.
911 Backport from trunk, r204517:
913 * libgomp.spec.in: Update comment about libgomp's dependencies.
914 * configure.ac: Likewise.
915 * configure: Regenerate.
917 2013-10-10 Jakub Jelinek <jakub@redhat.com>
919 * testsuite/libgomp.c++/udr-6.C: Remove UDR + on type F.
921 2013-10-08 Jakub Jelinek <jakub@redhat.com>
923 * testsuite/libgomp.c/thread-limit-1.c (main): Check if
924 cnt isn't bigger than 6 at any point in time, sleep 10ms after
925 incrementing it and then atomically decrement.
926 * testsuite/libgomp.c/thread-limit-2.c (main): Likewise.
928 * libgomp.h (struct gomp_team): Add task_queued_count field.
929 Add comments about task_{,queued_,running_}count.
930 * team.c (gomp_new_team): Clear task_queued_count.
931 * task.c (GOMP_task): Increment task_queued_count.
932 (gomp_task_run_pre): Decrement task_queued_count. If it is
933 decremented to zero, call gomp_team_barrier_clear_task_pending.
934 (gomp_task_run_post_handle_dependers): Increment task_queued_count.
935 (gomp_barrier_handle_tasks): Don't call
936 gomp_team_barrier_clear_task_pending here.
938 * testsuite/libgomp.c/udr-2.c (main): Remove unused variable i.
939 * testsuite/libgomp.c/udr-3.c (main): Initialize i to 0.
940 * testsuite/libgomp.c++/udr-8.C (main): Likewise.
942 * testsuite/libgomp.c/affinity-1.c (min_cpusetsize): New variable.
943 (pthread_getaffinity_np): Set it when setting contig_cpucount.
944 (print_affinity): Use it for size, only use sysconf value if that is not
945 set and if it is smaller than sizeof (cpu_set_t), use sizeof (cpu_set_t).
947 2013-10-07 Jakub Jelinek <jakub@redhat.com>
949 * testsuite/libgomp.c/affinity-1.c (main): Fork even if
950 contig_cpucount is > 8.
953 * config/linux/proc.c: Include errno.h.
954 (gomp_get_cpuset_size): New variable.
955 (gomp_cpuset_popcount): Add cpusetsize argument, use it instead of
957 (gomp_init_num_threads): If CPU_ALLOC_SIZE is defined and
958 pthread_getaffinity_np returned EINVAL, increase gomp_cpuset_size
959 and retry. After successful pthread_getaffinity_np copy
960 gomp_cpuset_size to gomp_get_cpuset_size and try to find out
961 minimum gomp_cpuset_size that covers all the CPUs set in gomp_cpusetp.
962 (get_num_procs): Pass gomp_get_cpuset_size rather than gomp_cpuset_size
963 to pthread_getaffinity_np, adjust gomp_cpuset_popcount caller.
964 * config/linux/proc.h (gomp_cpuset_popcount): Add cpusetsize argument.
965 * config/linux/affinity.c (gomp_affinity_finalize_place_list,
966 gomp_affinity_init_level): Adjust gomp_cpuset_popcount callers.
967 * testsuite/libgomp.c/affinity-1.c (pthread_getaffinity_np): Set
968 contig_cpucount from the first successful pthread_getaffinity_np
969 call, rather than just first call.
971 2013-10-04 Jakub Jelinek <jakub@redhat.com>
973 * libgomp.h (struct gomp_task_icv): Add thread_limit_var.
974 (gomp_thread_limit_var, gomp_remaining_threads_count,
975 gomp_remaining_threads_lock): Remove.
976 (gomp_managed_threads_lock): New variable.
977 (struct gomp_thread_pool): Add threads_busy field.
978 (gomp_free_thread): New prototype.
979 * parallel.c (gomp_resolve_num_threads): Adjust for
980 thread_limit now being in icv->thread_limit_var. Use
981 UINT_MAX instead of ULONG_MAX as infinity. If not nested,
982 just return minimum of max_num_threads and icv->thread_limit_var
983 and if thr->thread_pool, set threads_busy to the returned value.
984 Otherwise, don't update atomically gomp_remaining_threads_count,
985 but instead thr->thread_pool->threads_busy.
986 (GOMP_parallel_end): Adjust for thread_limit now being in
987 icv->thread_limit_var. Use UINT_MAX instead of ULONG_MAX as
988 infinity. Adjust threads_busy in the pool rather than
989 gomp_remaining_threads_count. Remember team->nthreads and call
990 gomp_team_end before adjusting threads_busy, if not nested
991 afterwards, just set it to 1 non-atomically.
992 * team.c (gomp_thread_start): Clear thr->thread_pool and
993 thr->task before returning.
994 (gomp_free_pool_helper): Clear thr->thread_pool and thr->task
995 before calling pthread_exit.
996 (gomp_free_thread): No longer static. Use
997 gomp_managed_threads_lock instead of gomp_remaining_threads_lock.
998 (gomp_team_start): Set thr->thread_pool->threads_busy to
999 nthreads immediately after creating new pool. Use
1000 gomp_managed_threads_lock instead of gomp_remaining_threads_lock.
1001 (gomp_team_end): Use gomp_managed_threads_lock instead of
1002 gomp_remaining_threads_lock.
1003 (initialize_team): Don't call gomp_sem_init here.
1004 * env.c (gomp_global_icv): Initialize thread_limit_var field.
1005 (gomp_thread_limit_var, gomp_remaining_threads_count,
1006 gomp_remaining_threads_lock): Remove.
1007 (gomp_managed_threads_locks): New variable.
1008 (handle_omp_display_env): Adjust for thread_limit var being
1010 (initialize_env): Likewise. If user provided value is > INT_MAX,
1011 set gomp_global_icv.thread_limit_var to UINT_MAX (infinity).
1012 Initialize gomp_managed_threads_lock instead of
1013 gomp_remaining_threads_lock.
1014 (omp_get_thread_limit): Adjust for thread_limit var in
1015 icv->thread_limit_var.
1016 * target.c: Include limits.h.
1017 (resolve_device): If device_id is < -1, return NULL.
1018 (gomp_map_vars): Make sure tgt->array, tgt->to_free, tgt->tgt_start,
1019 tgt->tgt_end and k->copy_from aren't left uninitialized.
1020 (GOMP_target): Arrange for host callback to be performed in a
1021 separate initial thread and contention group, inheriting ICVs from
1022 gomp_global_icv etc.
1023 (GOMP_teams): Adjust thread_limit_var ICV.
1024 * testsuite/libgomp.c/affinity-1.c (main): Call omp_set_dynamic (0).
1025 * testsuite/libgomp.c/target-3.c: New test.
1026 * testsuite/libgomp.c/target-4.c: New test.
1027 * testsuite/libgomp.c/target-5.c: New test.
1028 * testsuite/libgomp.c/target-6.c: New test.
1029 * testsuite/libgomp.c/target-7.c: New test.
1030 * testsuite/libgomp.c/thread-limit-1.c: New test.
1031 * testsuite/libgomp.c/thread-limit-2.c: New test.
1032 * testsuite/libgomp.c/thread-limit-3.c: New test.
1034 * libgomp.h (struct gomp_team_state): Add place_partition_off
1035 and place_partition_len fields.
1036 (struct gomp_task_icv): Add bind_var field.
1037 (gomp_bind_var_list, gomp_bind_var_list_len, gomp_places_list,
1038 gomp_places_list_len): New extern decls.
1039 (struct gomp_thread): Add place field.
1040 (gomp_cpu_affinity, gomp_cpu_affinity_len): Remove.
1041 (gomp_init_thread_affinity): Add place argument.
1042 (gomp_affinity_alloc, gomp_affinity_init_place, gomp_affinity_add_cpus,
1043 gomp_affinity_remove_cpu, gomp_affinity_copy_place,
1044 gomp_affinity_same_place, gomp_affinity_finalize_place_list,
1045 gomp_affinity_init_level, gomp_affinity_print_place): New prototypes.
1046 (gomp_team_start): Add flags argument.
1047 * team.c (struct gomp_thread_start_data): Add place field.
1048 (gomp_thread_start): Initialize thr->place.
1049 (gomp_team_start): Add flags argument. Handle OpenMP 4.0 affinity.
1050 * env.c (gomp_global_icv): Initialize bind_var field.
1051 (gomp_cpu_affinity, gomp_cpu_affinity_len): Remove.
1052 (gomp_bind_var_list, gomp_bind_var_list_len, gomp_places_list,
1053 gomp_places_list_len): New variables.
1054 (parse_bind_var, parse_one_place, parse_places_var): New functions.
1055 (parse_affinity): Rewritten to construct OMP_PLACES list with unit
1057 (handle_omp_display_env): Remove proc_bind argument. Set display
1058 to true for OMP_DISPLAY_ENV=true. Print 201307 instead of 201107
1059 as _OPENMP. Enhance printing of OMP_PROC_BIND, add printing
1060 of OMP_PLACES, don't print anything for GOMP_CPU_AFFINITY.
1061 (initialize_env): Use parse_bind_var to parse OMP_PROC_BIND
1062 instead of parse_boolean. Use parse_places_var for OMP_PLACES
1063 parsing. Don't call parse_affinity if OMP_PLACES has been
1064 successfully parsed (and call gomp_init_affinity in that case).
1065 Adjust handle_omp_display_env caller.
1066 (omp_get_proc_bind): Return bind_var ICV.
1067 * config/posix/affinity.c (gomp_init_thread_affinity): Add place
1069 (gomp_affinity_alloc, gomp_affinity_init_place, gomp_affinity_add_cpus,
1070 gomp_affinity_remove_cpu, gomp_affinity_copy_place,
1071 gomp_affinity_same_place, gomp_affinity_finalize_place_list,
1072 gomp_affinity_init_level, gomp_affinity_print_place): New stubs.
1073 * config/linux/proc.c (gomp_cpuset_popcount): Fix up check extern decl.
1074 (gomp_init_num_threads): If HAVE_PTHREAD_AFFINITY_NP isn't defined,
1075 fix up sizeof argument. Free and clear gomp_cpusetp if it didn't
1076 contain any logical CPUs.
1077 (get_num_procs): Check gomp_places_list instead of gomp_cpu_affinity.
1078 * config/linux/affinity.c: Include errno.h, stdio.h and string.h.
1079 (affinity_counter): Remove.
1080 (CPU_CLR_S): Define if CPU_ALLOC_SIZE is not defined.
1081 (gomp_init_affinity): Rewritten, if gomp_places_list is NULL, try
1082 silently create OMP_PLACES=threads, if it is non-NULL afterwards,
1083 bind current thread to the first place.
1084 (gomp_init_thread_affinity): Rewritten. Add place argument, just
1085 pthread_setaffinity_np to gomp_places_list[place].
1086 (gomp_affinity_alloc, gomp_affinity_init_place, gomp_affinity_add_cpus,
1087 gomp_affinity_remove_cpu, gomp_affinity_copy_place,
1088 gomp_affinity_same_place, gomp_affinity_finalize_place_list,
1089 gomp_affinity_init_level, gomp_affinity_print_place): New functions.
1090 * parallel.c (GOMP_parallel_start): Adjust gomp_team_start caller.
1091 (GOMP_parallel): Likewise, pass through flags parameter to it.
1092 * sections.c (GOMP_parallel_sections_start): Adjust gomp_team_start
1094 (GOMP_parallel_sections): Likewise, pass through flags parameter to it.
1095 * loop.c (gomp_parallel_loop_start): Add flags argument, pass it through
1097 (GOMP_parallel_loop_static_start, GOMP_parallel_loop_dynamic_start,
1098 GOMP_parallel_loop_guided_start, GOMP_parallel_loop_runtime_start):
1099 Adjust gomp_parallel_loop_start callers.
1100 (GOMP_parallel_loop_static, GOMP_parallel_loop_dynamic,
1101 GOMP_parallel_loop_guided, GOMP_parallel_loop_runtime): Likewise, pass
1102 through flags parameter to it.
1103 * testsuite/libgomp.c/affinity-1.c: New test.
1104 * testsuite/libgomp.c++/affinity-1.C: New test.
1106 2013-10-01 Jakub Jelinek <jakub@redhat.com>
1109 * config/linux/proc.c (gomp_cpuset_size, gomp_cpusetp): New variables.
1110 (gomp_cpuset_popcount): Use CPU_COUNT_S if available, or CPU_COUNT if
1111 gomp_cpuset_size is sizeof (cpu_set_t). Use gomp_cpuset_size instead
1112 of sizeof (cpu_set_t) to determine number of iterations.
1113 (gomp_init_num_threads): Initialize gomp_cpuset_size and gomp_cpusetp
1114 here, use gomp_cpusetp instead of &cpuset and pass gomp_cpuset_size
1115 instead of sizeof (cpu_set_t) to pthread_getaffinity_np.
1116 (get_num_procs): Don't call pthread_getaffinity_np if gomp_cpusetp
1117 is NULL. Use gomp_cpusetp instead of &cpuset and pass gomp_cpuset_size
1118 instead of sizeof (cpu_set_t) to pthread_getaffinity_np.
1119 * config/linux/proc.h (gomp_cpuset_popcount): Add attribute_hidden.
1120 (gomp_cpuset_size, gomp_cpusetp): Declare.
1121 * config/linux/affinity.c (CPU_ISSET_S, CPU_ZERO_S, CPU_SET_S): Define
1122 if CPU_ALLOC_SIZE isn't defined.
1123 (gomp_init_affinity): Don't call pthread_getaffinity_np here, instead
1124 use gomp_cpusetp computed by gomp_init_num_threads. Use CPU_*_S
1125 variants of macros with gomp_cpuset_size as set size, for cpusetnew
1126 use alloca for it if CPU_ALLOC_SIZE is defined, otherwise local
1127 fixed size variable.
1128 (gomp_init_thread_affinity): Use CPU_*_S variants of macros with
1129 gomp_cpuset_size as set size, for cpuset use alloca for it if
1130 CPU_ALLOC_SIZE is defined, otherwise local fixed size variable.
1132 2013-09-27 Jakub Jelinek <jakub@redhat.com>
1134 * libgomp.h: Include stdlib.h.
1135 (struct gomp_task_depend_entry,
1136 struct gomp_dependers_vec): New types.
1137 (struct gomp_task): Add dependers, depend_hash, depend_count,
1138 num_dependees and depend fields.
1139 (struct gomp_taskgroup): Add num_children field.
1140 (gomp_finish_task): Free depend_hash if non-NULL.
1141 * libgomp_g.h (GOMP_task): Add depend argument.
1142 * hashtab.h: New file.
1143 * task.c: Include hashtab.h.
1144 (hash_entry_type): New typedef.
1145 (htab_alloc, htab_free, htab_hash, htab_eq): New inlines.
1146 (gomp_init_task): Clear dependers, depend_hash and depend_count
1148 (GOMP_task): Add depend argument, handle depend clauses. Increment
1149 num_children field in taskgroup.
1150 (gomp_task_run_pre): Don't increment task_running_count here,
1151 nor clear task_pending bit.
1152 (gomp_task_run_post_handle_depend_hash,
1153 gomp_task_run_post_handle_dependers,
1154 gomp_task_run_post_handle_depend): New functions.
1155 (gomp_task_run_post_remove_parent): Clear in_taskwait before
1156 signalling corresponding semaphore.
1157 (gomp_task_run_post_remove_taskgroup): Decrement num_children
1158 field and make the decrement to 0 MEMMODEL_RELEASE operation,
1159 rather than storing NULL to taskgroup->children. Clear
1160 in_taskgroup_wait before signalling corresponding semaphore.
1161 (gomp_barrier_handle_tasks): Move task_running_count increment
1162 and task_pending bit clearing here. Call
1163 gomp_task_run_post_handle_depend. If more than one new tasks
1164 have been queued, wake other threads if needed.
1165 (GOMP_taskwait): Call gomp_task_run_post_handle_depend. If more
1166 than one new tasks have been queued, wake other threads if needed.
1167 After waiting on taskwait_sem, enter critical section again.
1168 (GOMP_taskgroup_start): Initialize num_children field.
1169 (GOMP_taskgroup_end): Check num_children instead of children
1170 before critical section. If children is NULL, but num_children
1171 is non-zero, wait on taskgroup_sem. Call
1172 gomp_task_run_post_handle_depend. If more than one new tasks have
1173 been queued, wake other threads if needed. After waiting on
1174 taskgroup_sem, enter critical section again.
1175 * testsuite/libgomp.c/depend-1.c: New test.
1176 * testsuite/libgomp.c/depend-2.c: New test.
1177 * testsuite/libgomp.c/depend-3.c: New test.
1178 * testsuite/libgomp.c/depend-4.c: New test.
1180 2013-09-26 Jakub Jelinek <jakub@redhat.com>
1183 * testsuite/libgomp.c/simd-6.c: New test.
1184 * testsuite/libgomp.c++/simd-8.C: New test.
1186 2013-09-25 Jakub Jelinek <jakub@redhat.com>
1188 * parallel.c (GOMP_cancellation_point, GOMP_cancel): Handle
1189 GIMPLE_CANCEL_TASKGROUP cancellation.
1190 * libgomp.h (struct gomp_task): Add next_taskgroup, prev_taskgroup,
1191 taskgroup and copy_ctors_done fields.
1192 (struct gomp_taskgroup): New type.
1193 * task.c (gomp_init_task): Initialize copy_ctors_done and taskgroup
1195 (GOMP_task): Don't start a new thread also if it's taskgroup has
1196 been cancelled. Set copy_ctors_done field if needed. Initialize
1197 taskgroup field. If copy_ctors_done and already cancelled, don't
1198 discard the task. If taskgroup is non-NULL, enqueue the task
1199 into taskgroup queue.
1200 (gomp_task_run_pre, gomp_task_run_post_remove_parent,
1201 gomp_task_run_post_remove_taskgroup): New inline functions.
1202 (gomp_barrier_handle_tasks, GOMP_taskwait): Use them.
1203 (GOMP_taskgroup_start, GOMP_taskgroup_end): Implement taskgroup
1205 * testsuite/libgomp.c++/cancel-parallel-1.C: New test.
1206 * testsuite/libgomp.c++/cancel-parallel-2.C: New test.
1207 * testsuite/libgomp.c++/cancel-parallel-3.C: New test.
1208 * testsuite/libgomp.c++/cancel-for-1.C: New test.
1209 * testsuite/libgomp.c++/cancel-for-1.C: New test.
1210 * testsuite/libgomp.c++/cancel-taskgroup-1.C: New test.
1211 * testsuite/libgomp.c++/cancel-taskgroup-2.C: New test.
1212 * testsuite/libgomp.c++/cancel-taskgroup-3.C: New test.
1213 * testsuite/libgomp.c++/cancel-test.h: New file.
1214 * testsuite/libgomp.c++/cancel-sections-1.C: New test.
1215 * testsuite/libgomp.c++/taskgroup-1.C: New test.
1216 * testsuite/libgomp.c/cancel-taskgroup-1.c: New test.
1217 * testsuite/libgomp.c/cancel-taskgroup-2.c: New test.
1218 * testsuite/libgomp.c/taskgroup-1.c: New test.
1219 * testsuite/libgomp.c/cancel-parallel-3.c (do_some_work): Use
1222 2013-09-24 Richard Henderson <rth@redhat.com>
1223 Jakub Jelinek <jakub@redhat.com>
1225 * libgomp.map (GOMP_4.0): Export GOMP_barrier_cancel,
1226 GOMP_loop_end_cancel, GOMP_sections_end_cancel.
1227 * parallel.c (GOMP_cancellation_point): Add WHICH argument, do nothing
1228 if gomp_cancel_var is false, handle workshare and parallel
1229 cancellation point, add ialias.
1230 (GOMP_cancel): Add WHICH and DO_CANCEL arguments, do nothing if
1231 gomp_cancel_var is false, call GOMP_cancellation_point if DO_CANCEL
1232 is false, handle workshare and parallel cancellation.
1233 * config/posix/bar.c (gomp_barrier_init): Clear cancellable field.
1234 (gomp_team_barrier_wait_end): Clear BAR_CANCELLED from state.
1235 Set work_share_cancelled to 0 on last thread, use __atomic_load_n.
1236 (gomp_team_barrier_wait_cancel_end, gomp_team_barrier_wait_cancel,
1237 gomp_team_barrier_cancel): New functions.
1238 * config/posix/bar.h (gomp_barrier_t): Add cancellable field.
1239 (BAR_CANCELLED): Define.
1240 (BAR_INCR): Set to 8.
1241 (gomp_team_barrier_wait_cancel, gomp_team_barrier_wait_cancel_end,
1242 gomp_team_barrier_cancel): New prototypes.
1243 (gomp_barrier_wait_start): Preserve BAR_CANCELLED bit.
1244 (gomp_barrier_wait_cancel_start, gomp_team_barrier_wait_final,
1245 gomp_team_barrier_cancelled): New inline functions.
1246 * config/linux/bar.c (gomp_team_barrier_wait_end): Clear BAR_CANCELLED
1247 from state where needed. Set work_share_cancelled to 0 on last thread.
1248 (gomp_team_barrier_wait_final, gomp_team_barrier_wait_cancel_end,
1249 gomp_team_barrier_wait_cancel, gomp_team_barrier_cancel): New
1251 * config/linux/bar.h (gomp_barrier_t): Add awaited_final field.
1252 (BAR_CANCELLED): Define.
1253 (BAR_INCR): Set to 8.
1254 (gomp_barrier_init): Initialize awaited_final field.
1255 (gomp_team_barrier_wait_final, gomp_team_barrier_wait_cancel,
1256 gomp_team_barrier_wait_cancel_end, gomp_team_barrier_cancel): New
1258 (gomp_barrier_wait_start): Preserve BAR_CANCELLED bit.
1259 (gomp_barrier_wait_cancel_start, gomp_team_barrier_wait_final_start,
1260 gomp_team_barrier_cancelled): New inline functions.
1261 * work.c (gomp_work_share_end, gomp_work_share_end_nowait): Set
1262 team->work_shares_to_free to thr->ts.work_share before calling
1264 (gomp_work_share_end_cancel): New function.
1265 * team.c (gomp_thread_start): Use gomp_team_barrier_wait_final
1266 instead of gomp_team_barrier_wait.
1267 (gomp_new_team): Initialize work_shares_to_free, work_share_cancelled
1268 and team_cancelled fields.
1269 (gomp_team_end): Use gomp_team_barrier_wait_final instead of
1270 gomp_team_barrier_wait. If team->team_cancelled, call
1271 gomp_fini_worshare on ws chain starting at team->work_shares_to_free
1272 rather than thr->ts.work_share.
1273 * env.c (gomp_global_icv): Initialize target_data field.
1274 (gomp_cancel_var): New global variable.
1275 (handle_omp_display_env): Print cancel-var ICV.
1276 (initialize_env): Parse OMP_CANCELLATION env var.
1277 (omp_get_cancellation): Return gomp_cancel_var.
1278 * barrier.c (GOMP_barrier_cancel): New function.
1279 * libgomp.h (gomp_cancel_var): New extern decl.
1280 (struct gomp_team): Add work_shares_to_free, work_share_cancelled
1281 and team_cancelled fields.
1282 (gomp_cancel_kind): New enum.
1283 (gomp_work_share_end_cancel): New prototype.
1284 * libgomp_g.h (GOMP_barrier_cancel, GOMP_loop_end_cancel,
1285 GOMP_sections_end_cancel): New prototypes.
1286 (GOMP_cancel, GOMP_cancellation_point): Adjust prototypes.
1287 * task.c (GOMP_task): If gomp_team_barrier_cancelled, don't
1288 queue or start new tasks.
1289 (gomp_barrier_handle_tasks): If gomp_team_barrier_cancelled,
1290 don't start any new tasks, just free all tasks.
1291 * sections.c (GOMP_sections_end_cancel): New function.
1292 * loop.c (GOMP_loop_end_cancel): New function.
1293 * testsuite/libgomp.c/cancel-parallel-1.c: New test.
1294 * testsuite/libgomp.c/cancel-parallel-2.c: New test.
1295 * testsuite/libgomp.c/cancel-parallel-3.c: New test.
1296 * testsuite/libgomp.c/cancel-for-1.c: New test.
1297 * testsuite/libgomp.c/cancel-for-2.c: New test.
1298 * testsuite/libgomp.c/cancel-sections-1.c: New test.
1300 2013-09-23 Michael Zolotukhin <michael.v.zolotukhin@intel.com>
1302 * configure.ac: Add checks for plugins support.
1303 * config.h.in: Regenerated.
1304 * configure: Regenerated.
1305 * target.c (struct target_mem_desc): Add device_descr field.
1308 (struct gomp_device_descr): New.
1309 (gomp_get_num_devices): Call gomp_target_init.
1310 (resolve_device): Return device_descr instead of int.
1311 (gomp_map_vars): Add devicep argument and update the function
1313 (gomp_unmap_tgt): Likewise.
1314 (gomp_unmap_vars): Likewise.
1315 (gomp_update): Likewise.
1316 (GOMP_target): Use device_descr struct.
1317 (GOMP_target_data): Likewise.
1318 (GOMP_target_update): Likewise.
1319 (gomp_check_plugin_file_name): New.
1320 (gomp_load_plugin_for_device): New.
1321 (gomp_find_available_plugins): New.
1322 (gomp_target_init): New.
1324 2013-09-19 Jakub Jelinek <jakub@redhat.com>
1326 * libgomp_g.h (GOMP_target): Change 3rd argument to const void *.
1327 (GOMP_target_data, GOMP_target_update): Add openmp_target argument.
1328 * target.c (GOMP_target): Change fnname argument to openmp_target,
1329 with const void * type.
1330 (GOMP_target_data, GOMP_target_update): Add openmp_target argument.
1332 * testsuite/libgomp.c/simd-4.c: New test.
1333 * testsuite/libgomp.c/simd-5.c: New test.
1334 * testsuite/libgomp.c/udr-1.c: New test.
1335 * testsuite/libgomp.c/udr-2.c: New test.
1336 * testsuite/libgomp.c/udr-3.c: New test.
1337 * testsuite/libgomp.c++/udr-9.C: New test.
1339 2013-09-18 Jakub Jelinek <jakub@redhat.com>
1341 * testsuite/libgomp.c++/simd-4.C: New test.
1342 * testsuite/libgomp.c++/simd-5.C: New test.
1343 * testsuite/libgomp.c++/simd-6.C: New test.
1344 * testsuite/libgomp.c++/simd-7.C: New test.
1345 * testsuite/libgomp.c++/udr-1.C: New test.
1346 * testsuite/libgomp.c++/udr-2.C: New test.
1347 * testsuite/libgomp.c++/udr-3.C: New test.
1348 * testsuite/libgomp.c++/udr-4.C: New test.
1349 * testsuite/libgomp.c++/udr-5.C: New test.
1350 * testsuite/libgomp.c++/udr-6.C: New test.
1351 * testsuite/libgomp.c++/udr-7.C: New test.
1352 * testsuite/libgomp.c++/udr-8.C: New test.
1354 2013-09-16 Jakub Jelinek <jakub@redhat.com>
1356 * splay-tree.h: New file.
1357 * target.c: Include stdbool.h.
1358 (splay_tree_node, splay_tree, splay_tree_key): New typedefs.
1359 (struct target_mem_desc, struct splay_tree_key_s): New structures.
1360 (splay_compare): New inline function.
1361 (gomp_get_num_devices): New function.
1362 (resolve_device): Use default_device_var ICV. Add temporarily
1363 magic testing device number 257.
1364 (dev_splay_tree, dev_env_lock): New variables.
1365 (gomp_map_vars_existing, gomp_map_vars, gomp_unmap_tgt,
1366 gomp_unmap_vars, gomp_update): New functions.
1367 (GOMP_target, GOMP_target_data, GOMP_target_end_data,
1368 GOMP_target_update): Add support for magic testing device number 257.
1369 * libgomp.h (struct target_mem_desc): Forward declare.
1370 (struct gomp_task_icv): Add default_device_var and target_data.
1371 (gomp_get_num_devices): New prototype.
1372 * env.c (gomp_global_icv): Add default_device_var initializer.
1373 (parse_int): New function.
1374 (handle_omp_display_env): Print OMP_DEFAULT_DEVICE.
1375 (initialize_env): Initialize default_device_var.
1376 (omp_set_default_device): Set default_device_var ICV.
1377 (omp_get_default_device): Query default_device_var ICV.
1378 (omp_get_num_devices): Call gomp_get_num_devices.
1379 (omp_get_num_teams, omp_get_team_num, omp_is_initial_device): Add
1382 2013-09-13 Jakub Jelinek <jakub@redhat.com>
1384 PR tree-optimization/58392
1385 * testsuite/libgomp.c/pr58392.c: New test.
1387 2013-09-06 Jakub Jelinek <jakub@redhat.com>
1389 * testsuite/libgomp.c/target-2.c: New test.
1390 * testsuite/libgomp.c++/target-3.C: New test.
1392 2013-09-05 Jakub Jelinek <jakub@redhat.com>
1394 * testsuite/libgomp.c++/target-2.C: New test.
1395 * testsuite/libgomp.c++/target-2-aux.cc: New file.
1397 * libgomp.map (GOMP_4.0): Add GOMP_target, GOMP_target_data,
1398 GOMP_target_end_data, GOMP_target_update and GOMP_teams.
1399 * Makefile.am (libgomp_la_SOURCES): Add target.c.
1400 * Makefile.in: Regenerated.
1401 * target.c: New file.
1402 * libgomp_g.h (GOMP_target, GOMP_target_data,
1403 GOMP_target_end_data, GOMP_target_update, GOMP_teams): New prototypes.
1404 * testsuite/libgomp.c++/for-11.C (main): Uncomment
1405 #pragma omp target teams directive.
1406 * testsuite/libgomp.c/for-3.c (main): Likewise.
1407 * testsuite/libgomp.c++/target-1.C: New test.
1408 * testsuite/libgomp.c/target-1.c: New test.
1410 2013-07-06 Jakub Jelinek <jakub@redhat.com>
1412 * testsuite/libgomp.c/for-3.c: New test.
1413 * testsuite/libgomp.c/simd-1.c: New test.
1414 * testsuite/libgomp.c/simd-2.c: New test.
1415 * testsuite/libgomp.c/simd-3.c: New test.
1417 2013-07-05 Richard Henderson <rth@redhat.com>
1419 * config/linux/bar.c (gomp_team_barrier_wait_end): Remove one more
1420 hard-coded constant.
1422 2013-07-03 Richard Henderson <rth@redhat.com>
1424 * config/linux/bar.h (BAR_TASK_PENDING, BAR_WAS_LAST,
1425 BAR_WAITING_FOR_TASK, BAR_INCR): New.
1426 (gomp_barrier_wait_start, gomp_barrier_last_thread,
1427 gomp_team_barrier_set_task_pending,
1428 gomp_team_barrier_clear_task_pending,
1429 gomp_team_barrier_set_waiting_for_tasks,
1430 gomp_team_barrier_waiting_for_tasks,
1431 gomp_team_barrier_done): Use them.
1432 * config/linux/bar.c (gomp_barrier_wait_end,
1433 gomp_barrier_wait_last, gomp_team_barrier_wait_end): Likewise.
1434 * config/posix/bar.h, config/posix/bar.c: Similarly.
1436 2013-07-03 Jakub Jelinek <jakub@redhat.com>
1438 * testsuite/libgomp.c++/atomic-14.C: Adjust for stricter atomic
1440 * testsuite/libgomp.c++/atomic-15.C: Likewise.
1441 * testsuite/libgomp.c/atomic-17.c: Likewise.
1443 2013-06-28 Jakub Jelinek <jakub@redhat.com>
1445 * testsuite/libgomp.c++/simd-1.C: New test.
1446 * testsuite/libgomp.c++/simd-2.C: New test.
1447 * testsuite/libgomp.c++/simd-3.C: New test.
1449 2013-06-21 Jakub Jelinek <jakub@redhat.com>
1451 * testsuite/libgomp.c/for-1.h: New file.
1452 * testsuite/libgomp.c/for-2.h: New file.
1453 * testsuite/libgomp.c/for-1.c: New test.
1454 * testsuite/libgomp.c/for-2.c: New test.
1455 * testsuite/libgomp.c++/for-9.C: New test.
1456 * testsuite/libgomp.c++/for-10.C: New test.
1457 * testsuite/libgomp.c++/for-11.C: New test.
1459 2013-06-12 Jakub Jelinek <jakub@redhat.com>
1461 * fortran.c (omp_is_initial_device): Add ialias_redirect.
1462 (omp_is_initial_device_): New function.
1463 * omp_lib.f90.in (omp_is_initial_device): New interface.
1464 * omp.h.in (omp_is_initial_device): New prototype.
1465 * libgomp.map (omp_is_initial_device, omp_is_initial_device_):
1467 * env.c (omp_is_initial_device): New function. Add ialias for it.
1468 * omp_lib.h.in (omp_is_initial_device): New external.
1470 * omp_lib.f90.in (omp_get_dynamic, omp_get_nested,
1471 omp_in_parallel, omp_get_max_threads, omp_get_num_procs,
1472 omp_get_num_threads, omp_get_thread_num, omp_get_thread_limit,
1473 omp_set_max_active_levels, omp_get_max_active_levels,
1474 omp_get_level, omp_get_ancestor_thread_num,
1475 omp_get_team_size, omp_get_active_level, omp_in_final,
1476 omp_get_cancellation, omp_get_default_device,
1477 omp_get_num_devices, omp_get_num_teams, omp_get_team_num): Remove
1478 useless use omp_lib_kinds.
1480 2013-04-30 Jakub Jelinek <jakub@redhat.com>
1482 * testsuite/libgomp.c/atomic-17.c: New test.
1483 * testsuite/libgomp.c++/atomic-14.C: New test.
1484 * testsuite/libgomp.c++/atomic-15.C: New test.
1486 2013-04-10 Jakub Jelinek <jakub@redhat.com>
1488 * libgomp.map (omp_get_cancellation, omp_get_cancellation_,
1489 omp_get_proc_bind, omp_get_proc_bind_, omp_set_default_device,
1490 omp_set_default_device_, omp_set_default_device_8_,
1491 omp_get_default_device, omp_get_default_device_,
1492 omp_get_num_devices, omp_get_num_devices_, omp_get_num_teams,
1493 omp_get_num_teams_, omp_get_team_num, omp_get_team_num_): Export
1495 (GOMP_cancel, GOMP_cancellation_point, GOMP_parallel_loop_dynamic,
1496 GOMP_parallel_loop_guided, GOMP_parallel_loop_runtime,
1497 GOMP_parallel_loop_static, GOMP_parallel_sections, GOMP_parallel,
1498 GOMP_taskgroup_start, GOMP_taskgroup_end): Export @@GOMP_4.0.
1499 * parallel.c (GOMP_parallel_end): Add ialias.
1500 (GOMP_parallel, GOMP_cancel, GOMP_cancellation_point): New
1502 * omp.h.in (omp_proc_bind_t): New typedef.
1503 (omp_get_cancellation, omp_get_proc_bind, omp_set_default_device,
1504 omp_get_default_device, omp_get_num_devices, omp_get_num_teams,
1505 omp_get_team_num): New prototypes.
1506 * env.c (omp_get_cancellation, omp_get_proc_bind,
1507 omp_set_default_device, omp_get_default_device, omp_get_num_devices,
1508 omp_get_num_teams, omp_get_team_num): New functions.
1509 * fortran.c (ULP, STR1, STR2, ialias_redirect): Removed.
1510 (omp_get_cancellation_, omp_get_proc_bind_, omp_set_default_device_,
1511 omp_set_default_device_8_, omp_get_default_device_,
1512 omp_get_num_devices_, omp_get_num_teams_, omp_get_team_num_): New
1514 * libgomp.h (ialias_ulp, ialias_str1, ialias_str2, ialias_redirect,
1515 ialias_call): Define.
1516 * libgomp_g.h (GOMP_parallel_loop_static, GOMP_parallel_loop_dynamic,
1517 GOMP_parallel_loop_guided, GOMP_parallel_loop_runtime, GOMP_parallel,
1518 GOMP_cancel, GOMP_cancellation_point, GOMP_taskgroup_start,
1519 GOMP_taskgroup_end, GOMP_parallel_sections): New prototypes.
1520 * task.c (GOMP_taskgroup_start, GOMP_taskgroup_end): New functions.
1521 * sections.c (GOMP_parallel_sections): New function.
1522 * loop.c (GOMP_parallel_loop_static, GOMP_parallel_loop_dynamic,
1523 GOMP_parallel_loop_guided, GOMP_parallel_loop_runtime): New
1525 (GOMP_parallel_end): Add ialias_redirect.
1526 * omp_lib.f90.in (omp_proc_bind_kind, omp_proc_bind_false,
1527 omp_proc_bind_true, omp_proc_bind_master, omp_proc_bind_close,
1528 omp_proc_bind_spread): New params.
1529 (omp_get_cancellation, omp_get_proc_bind, omp_set_default_device,
1530 omp_get_default_device, omp_get_num_devices, omp_get_num_teams,
1531 omp_get_team_num): New interfaces.
1532 * omp_lib.h.in (omp_proc_bind_kind, omp_proc_bind_false,
1533 omp_proc_bind_true, omp_proc_bind_master, omp_proc_bind_close,
1534 omp_proc_bind_spread): New params.
1535 (omp_get_cancellation, omp_get_proc_bind, omp_set_default_device,
1536 omp_get_default_device, omp_get_num_devices, omp_get_num_teams,
1537 omp_get_team_num): New externals.
1539 2013-03-20 Tobias Burnus <burnus@net-b.de>
1541 * libgomp.texi (Environment Variables): Minor cleanup,
1542 update section refs to OpenMP 4.0rc2.
1543 (OMP_DISPLAY_ENV, GOMP_SPINCOUNT): Document these
1544 environment variables.
1546 2013-03-20 Tobias Burnus <burnus@net-b.de>
1548 * env.c (handle_omp_display_env): New function.
1549 (initialize_env): Use it.
1551 2013-03-20 Jakub Jelinek <jakub@redhat.com>
1553 * testsuite/libgomp.c/atomic-14.c: Add parens to make it valid.
1554 * testsuite/libgomp.c/atomic-15.c: New test.
1555 * testsuite/libgomp.c/atomic-16.c: New test.
1557 * testsuite/libgomp.c++/atomic-10.C: New test.
1558 * testsuite/libgomp.c++/atomic-11.C: New test.
1559 * testsuite/libgomp.c++/atomic-12.C: New test.
1560 * testsuite/libgomp.c++/atomic-13.C: New test.
1562 Copyright (C) 2013-2014 Free Software Foundation, Inc.
1564 Copying and distribution of this file, with or without modification,
1565 are permitted in any medium without royalty provided the copyright
1566 notice and this notice are preserved.