9
9
10
10
### XMethod implementations ###
11
11
12
+
12
13
class Accessor :
13
14
"""Generalized base class for buffer index calculation"""
14
15
@@ -31,15 +32,17 @@ def index(self, arg):
31
32
return int (arg )
32
33
# unwrap if inside item
33
34
try :
34
- arg = arg [' MImpl' ][ ' MIndex' ]
35
+ arg = arg [" MImpl" ][ " MIndex" ]
35
36
except :
36
37
pass
37
38
# https://github.com/intel/llvm/blob/97272b7ebd569bfa13811913a31e30f926559217/sycl/include/CL/sycl/accessor.hpp#L678-L690
38
39
result = 0
39
40
for dim in range (self .depth ):
40
- result = result * self .memory_range (dim ) + \
41
- self .offset (dim ) + \
42
- arg ['common_array' ][dim ]
41
+ result = (
42
+ result * self .memory_range (dim )
43
+ + self .offset (dim )
44
+ + arg ["common_array" ][dim ]
45
+ )
43
46
return result
44
47
45
48
def value (self , arg ):
@@ -50,16 +53,17 @@ class HostAccessor(Accessor):
50
53
"""For Host device memory layout"""
51
54
52
55
def payload (self ):
53
- return self .obj [' impl' ][ ' _M_ptr' ].dereference ()
56
+ return self .obj [" impl" ][ " _M_ptr" ].dereference ()
54
57
55
58
def memory_range (self , dim ):
56
- return self .payload ()[' MMemoryRange' ][ ' common_array' ][dim ]
59
+ return self .payload ()[" MMemoryRange" ][ " common_array" ][dim ]
57
60
58
61
def offset (self , dim ):
59
- return self .payload ()[' MOffset' ][ ' common_array' ][dim ]
62
+ return self .payload ()[" MOffset" ][ " common_array" ][dim ]
60
63
61
64
def data (self ):
62
- return self .payload ()['MData' ]
65
+ return self .payload ()["MData" ]
66
+
63
67
64
68
class HostAccessorLocal (HostAccessor ):
65
69
"""For Host device memory layout"""
@@ -68,27 +72,29 @@ def index(self, arg):
68
72
if arg .type .code == gdb .TYPE_CODE_INT :
69
73
return int (arg )
70
74
# https://github.com/intel/llvm/blob/97272b7ebd569bfa13811913a31e30f926559217/sycl/include/CL/sycl/accessor.hpp#L1049-L1053
71
- result = 0 ;
75
+ result = 0
72
76
for dim in range (self .depth ):
73
- result = result * \
74
- self .payload ()['MSize' ]['common_array' ][dim ] + \
75
- arg ['common_array' ][dim ];
76
- return result ;
77
+ result = (
78
+ result * self .payload ()["MSize" ]["common_array" ][dim ]
79
+ + arg ["common_array" ][dim ]
80
+ )
81
+ return result
77
82
78
83
def data (self ):
79
- return self .payload ()['MMem' ]
84
+ return self .payload ()["MMem" ]
85
+
80
86
81
87
class DeviceAccessor (Accessor ):
82
88
"""For CPU/GPU memory layout"""
83
89
84
90
def memory_range (self , dim ):
85
- return self .obj [' impl' ][ ' MemRange' ][ ' common_array' ][dim ]
91
+ return self .obj [" impl" ][ " MemRange" ][ " common_array" ][dim ]
86
92
87
93
def offset (self , dim ):
88
- return self .obj [' impl' ][ ' Offset' ][ ' common_array' ][dim ]
94
+ return self .obj [" impl" ][ " Offset" ][ " common_array" ][dim ]
89
95
90
96
def data (self ):
91
- return self .obj [' MData' ]
97
+ return self .obj [" MData" ]
92
98
93
99
94
100
class AccessorOpIndex (gdb .xmethod .XMethodWorker ):
@@ -106,12 +112,12 @@ def get_result_type(self, *args):
106
112
return self .result_type
107
113
108
114
def __call__ (self , obj , arg ):
109
- # No way to wasily figure out which devices is currently being used,
115
+ # No way to easily figure out which devices is currently being used,
110
116
# try all accessor implementations until one of them works:
111
117
accessors = [
112
118
DeviceAccessor (obj , self .result_type , self .depth ),
113
119
HostAccessor (obj , self .result_type , self .depth ),
114
- HostAccessorLocal (obj , self .result_type , self .depth )
120
+ HostAccessorLocal (obj , self .result_type , self .depth ),
115
121
]
116
122
for accessor in accessors :
117
123
try :
@@ -129,39 +135,41 @@ class AccessorOpIndex1D(AccessorOpIndex):
129
135
130
136
def get_arg_types (self ):
131
137
assert self .depth == 1
132
- return gdb .lookup_type ('size_t' )
138
+ return gdb .lookup_type ("size_t" )
139
+
133
140
134
141
class AccessorOpIndexItemTrue (AccessorOpIndex ):
135
142
"""Introduces an extra overload for item wrapper"""
136
143
137
144
def get_arg_types (self ):
138
145
return gdb .lookup_type ("cl::sycl::item<%s, true>" % self .depth )
139
146
147
+
140
148
class AccessorOpIndexItemFalse (AccessorOpIndex ):
141
149
"""Introduces an extra overload for item wrapper"""
142
150
143
151
def get_arg_types (self ):
144
152
return gdb .lookup_type ("cl::sycl::item<%s, false>" % self .depth )
145
153
154
+
146
155
class AccessorMatcher (gdb .xmethod .XMethodMatcher ):
147
156
"""Entry point for cl::sycl::accessor"""
157
+
148
158
def __init__ (self ):
149
- gdb .xmethod .XMethodMatcher .__init__ (self , ' AccessorMatcher' )
159
+ gdb .xmethod .XMethodMatcher .__init__ (self , " AccessorMatcher" )
150
160
151
161
def match (self , class_type , method_name ):
152
- if method_name != ' operator[]' :
162
+ if method_name != " operator[]" :
153
163
return None
154
164
155
- result = re .match (' ^cl::sycl::accessor<.+>$' , class_type .tag )
165
+ result = re .match (" ^cl::sycl::accessor<.+>$" , class_type .tag )
156
166
if result is None :
157
167
return None
158
168
159
169
depth = int (class_type .template_argument (1 ))
160
170
result_type = class_type .template_argument (0 )
161
171
162
- methods = [
163
- AccessorOpIndex (class_type , result_type , depth )
164
- ]
172
+ methods = [AccessorOpIndex (class_type , result_type , depth )]
165
173
try :
166
174
method = AccessorOpIndexItemTrue (class_type , result_type , depth )
167
175
method .get_arg_types ()
@@ -178,38 +186,49 @@ def match(self, class_type, method_name):
178
186
methods .append (AccessorOpIndex1D (class_type , result_type , depth ))
179
187
return methods
180
188
189
+
181
190
class PrivateMemoryOpCall (gdb .xmethod .XMethodWorker ):
182
191
"""Provides operator() overload for h_item argument"""
183
192
184
193
class ItemBase :
185
194
"""Wrapper for cl::sycl::detail::ItemBase which reimplements index calculation"""
186
195
187
- def __init__ (self , obj , ):
188
- result = re .match ('^cl::sycl::detail::ItemBase<(.+), (.+)>$' , str (obj .type ))
196
+ def __init__ (
197
+ self ,
198
+ obj ,
199
+ ):
200
+ result = re .match ("^cl::sycl::detail::ItemBase<(.+), (.+)>$" , str (obj .type ))
189
201
self .dim = int (result [1 ])
190
- self .with_offset = ( result [2 ] == ' true' )
202
+ self .with_offset = result [2 ] == " true"
191
203
self .obj = obj
192
204
193
205
def get_linear_id (self ):
194
- index = self .obj [' MIndex' ][ ' common_array' ]
195
- extent = self .obj [' MExtent' ][ ' common_array' ]
206
+ index = self .obj [" MIndex" ][ " common_array" ]
207
+ extent = self .obj [" MExtent" ][ " common_array" ]
196
208
197
209
if self .with_offset :
198
- offset = self .obj [' MOffset' ][ ' common_array' ]
210
+ offset = self .obj [" MOffset" ][ " common_array" ]
199
211
if self .dim == 1 :
200
212
return index [0 ] - offset [0 ]
201
213
elif self .dim == 2 :
202
214
return (index [0 ] - offset [0 ]) * extent [1 ] + (index [1 ] - offset [1 ])
203
215
else :
204
- return ((index [0 ] - offset [0 ]) * extent [1 ] * extent [2 ]) + \
205
- ((index [1 ] - offset [1 ]) * extent [2 ]) + (index [2 ] - offset [2 ])
216
+ return (
217
+ ((index [0 ] - offset [0 ]) * extent [1 ] * extent [2 ])
218
+ + ((index [1 ] - offset [1 ]) * extent [2 ])
219
+ + (index [2 ] - offset [2 ])
220
+ )
206
221
else :
207
222
if self .dim == 1 :
208
223
return index [0 ]
209
224
elif self .dim == 2 :
210
225
return index [0 ] * extent [1 ] + index [1 ]
211
226
else :
212
- return (index [0 ] * extent [1 ] * extent [2 ]) + (index [1 ] * extent [2 ]) + index [2 ]
227
+ return (
228
+ (index [0 ] * extent [1 ] * extent [2 ])
229
+ + (index [1 ] * extent [2 ])
230
+ + index [2 ]
231
+ )
213
232
214
233
def __init__ (self , result_type , dim ):
215
234
self .result_type = result_type
@@ -222,38 +241,41 @@ def get_result_type(self, *args):
222
241
return self .result_type
223
242
224
243
def __call__ (self , obj , * args ):
225
- if obj [' Val' ].type .tag .endswith (self .result_type ):
244
+ if obj [" Val" ].type .tag .endswith (self .result_type ):
226
245
# On device private_memory is a simple wrapper over actual value
227
- return obj [' Val' ]
246
+ return obj [" Val" ]
228
247
else :
229
248
# On host it wraps a unique_ptr to an array of items
230
- item_base = args [0 ][' localItem' ][ ' MImpl' ]
249
+ item_base = args [0 ][" localItem" ][ " MImpl" ]
231
250
item_base = self .ItemBase (item_base )
232
251
index = item_base .get_linear_id ()
233
- return obj ['Val' ]['_M_t' ]['_M_t' ]['_M_head_impl' ][index ]
252
+ return obj ["Val" ]["_M_t" ]["_M_t" ]["_M_head_impl" ][index ]
253
+
234
254
235
255
class PrivateMemoryMatcher (gdb .xmethod .XMethodMatcher ):
236
256
"""Entry point for cl::sycl::private_memory"""
237
257
238
258
def __init__ (self ):
239
- gdb .xmethod .XMethodMatcher .__init__ (self , ' PrivateMemoryMatcher' )
259
+ gdb .xmethod .XMethodMatcher .__init__ (self , " PrivateMemoryMatcher" )
240
260
241
261
def match (self , class_type , method_name ):
242
- if method_name != ' operator()' :
262
+ if method_name != " operator()" :
243
263
return None
244
264
245
- result = re .match ('^cl::sycl::private_memory<((cl::)?(sycl::)?id<.+>), (.+)>$' , class_type .tag )
265
+ result = re .match (
266
+ "^cl::sycl::private_memory<((cl::)?(sycl::)?id<.+>), (.+)>$" , class_type .tag
267
+ )
246
268
if result is None :
247
269
return None
248
270
return PrivateMemoryOpCall (result [1 ], result [4 ])
249
271
250
272
251
-
252
273
gdb .xmethod .register_xmethod_matcher (None , AccessorMatcher (), replace = True )
253
274
gdb .xmethod .register_xmethod_matcher (None , PrivateMemoryMatcher (), replace = True )
254
275
255
276
### Pretty-printer implementations ###
256
277
278
+
257
279
class SyclArrayPrinter :
258
280
"""Print an object deriving from cl::sycl::detail::array"""
259
281
@@ -275,7 +297,7 @@ def __next__(self):
275
297
elt = self .data [count ]
276
298
except :
277
299
elt = "<error reading variable>"
278
- return (' [%d]' % count , elt )
300
+ return (" [%d]" % count , elt )
279
301
280
302
def __init__ (self , value ):
281
303
if value .type .code == gdb .TYPE_CODE_REF :
@@ -288,7 +310,7 @@ def __init__(self, value):
288
310
289
311
def children (self ):
290
312
try :
291
- return self .ElementIterator (self .value [' common_array' ], self .dimensions )
313
+ return self .ElementIterator (self .value [" common_array" ], self .dimensions )
292
314
except :
293
315
# There is no way to return an error from this method. Return an
294
316
# empty iterable to make GDB happy and rely on to_string method
@@ -300,13 +322,14 @@ def to_string(self):
300
322
# Check if accessing array value will succeed and resort to
301
323
# error message otherwise. Individual array element access failures
302
324
# will be caught by iterator itself.
303
- _ = self .value [' common_array' ]
325
+ _ = self .value [" common_array" ]
304
326
return self .type .tag
305
327
except :
306
328
return "<error reading variable>"
307
329
308
330
def display_hint (self ):
309
- return 'array'
331
+ return "array"
332
+
310
333
311
334
class SyclBufferPrinter :
312
335
"""Print a cl::sycl::buffer"""
@@ -316,18 +339,22 @@ def __init__(self, value):
316
339
self .type = value .type .unqualified ().strip_typedefs ()
317
340
self .elt_type = value .type .template_argument (0 )
318
341
self .dimensions = value .type .template_argument (1 )
319
- self .typeregex = re .compile (' ^([a-zA-Z0-9_:]+)(<.*>)?$' )
342
+ self .typeregex = re .compile (" ^([a-zA-Z0-9_:]+)(<.*>)?$" )
320
343
321
344
def to_string (self ):
322
345
match = self .typeregex .match (self .type .tag )
323
346
if not match :
324
347
return "<error parsing type>"
325
- return ('%s<%s, %s> = {impl=%s}'
326
- % (match .group (1 ), self .elt_type , self .dimensions ,
327
- self .value ['impl' ].address ))
348
+ return "%s<%s, %s> = {impl=%s}" % (
349
+ match .group (1 ),
350
+ self .elt_type ,
351
+ self .dimensions ,
352
+ self .value ["impl" ].address ,
353
+ )
354
+
328
355
329
356
sycl_printer = gdb .printing .RegexpCollectionPrettyPrinter ("SYCL" )
330
- sycl_printer .add_printer ("cl::sycl::id" , ' ^cl::sycl::id<.*$' , SyclArrayPrinter )
331
- sycl_printer .add_printer ("cl::sycl::range" , ' ^cl::sycl::range<.*$' , SyclArrayPrinter )
332
- sycl_printer .add_printer ("cl::sycl::buffer" , ' ^cl::sycl::buffer<.*$' , SyclBufferPrinter )
357
+ sycl_printer .add_printer ("cl::sycl::id" , " ^cl::sycl::id<.*$" , SyclArrayPrinter )
358
+ sycl_printer .add_printer ("cl::sycl::range" , " ^cl::sycl::range<.*$" , SyclArrayPrinter )
359
+ sycl_printer .add_printer ("cl::sycl::buffer" , " ^cl::sycl::buffer<.*$" , SyclBufferPrinter )
333
360
gdb .printing .register_pretty_printer (None , sycl_printer , True )
0 commit comments