def test_DatabaseObserver(db: grewe_features_db.Database):
with db.Session(commit=True) as s:
s.add(
_StaticFeatures("benchmarks_suite:benchmark:kernel", "kernel void A() {}")
)
observer = run_gpgpu_benchmarks.DatabaseObserver(db)
observer.OnBenchmarkRun(
gpgpu_pb2.GpgpuBenchmarkRun(
benchmark_suite="suite",
benchmark_name="benchmark",
dataset_name="dataset",
hostname="hostname",
run=libcecl_pb2.LibceclExecutableRun(
returncode=0,
device=clinfo_pb2.OpenClDevice(name="opencl_env"),
kernel_invocation=[
libcecl_pb2.OpenClKernelInvocation(
kernel_name="kernel",
global_size=128,
local_size=64,
transferred_bytes=123,
transfer_time_ns=3000,
kernel_time_ns=1000,
)
],
),
)
)
with db.Session() as s:
assert s.query(grewe_features_db.DynamicFeatures).count() == 0
observer.CommitRecords()
with db.Session() as s:
assert s.query(grewe_features_db.DynamicFeatures).count() == 1
record = s.query(grewe_features_db.DynamicFeatures).one()
assert record.driver == grewe_features_db.DynamicFeaturesDriver.LIBCECL
assert (
record.static_features_id
== s.query(grewe_features_db.StaticFeatures.id).one()
)
assert record.opencl_env == "opencl_env"
assert record.hostname == "hostname"
assert record.outcome == "PASS"
assert record.gsize == 128
assert record.wgsize == 64
assert record.work_item_local_mem_size == None
assert record.work_item_private_mem_size == None
assert record.transferred_bytes == 123
assert record.transfer_time_ns == 3000
assert record.kerenl_time_ns == 1000
def ProcessBatch(
batch: typing.List[typing.Tuple[str, ]],
db: grewe_features_db.Database,
pool: multiprocessing.Pool,
):
app.Log(1, "Formatting files")
srcs = pool.imap_unordered(_PrettifySource, [row[0] for row in batch])
srcs = [s for s in srcs if s]
with tempfile.TemporaryDirectory(prefix="phd_clgen_import_") as d:
d = pathlib.Path(d)
paths = [d / f"{i}.cl" for i in range(1, len(srcs) + 1)]
for path, src in zip(paths, srcs):
with open(path, "w") as f:
f.write(src)
app.Log(1, "Importing files")
success_count, new_row_count = db.ImportStaticFeaturesFromPaths(
paths, FLAGS.origin, pool)
app.Log(
1,
"Extracted features from %d of %d kernels, %d new rows",
success_count,
len(batch),
new_row_count,
)
def test_DriveKernelAndRecordResults_input_insensitive(
db: grewe_features_db.Database,
env: cldrive_env.OpenCLEnvironment,
dynamic_params,
num_runs: int,
):
drive_with_cldrive.DriveKernelAndRecordResults(
db,
10,
"kernel void A(global int* a) { a[get_global_id(0)] = 0; }",
env,
dynamic_params,
num_runs,
)
with db.Session() as s:
assert s.query(
grewe_features_db.DynamicFeatures).count() == (len(dynamic_params))
records = s.query(grewe_features_db.DynamicFeatures).all()
for i, record in enumerate(records):
assert record.static_features_id == 10
assert record.opencl_env == env.name
assert record.hostname == system.HOSTNAME
assert record.outcome == "INPUT_INSENSITIVE"
assert record.gsize == dynamic_params[i].global_size_x
assert record.wgsize == dynamic_params[i].local_size_x
assert record.work_item_local_mem_size == 0
assert record.work_item_private_mem_size == 0
assert record.transferred_bytes is None
assert record.transfer_time_ns is None
assert record.kernel_time_ns is None
def test_DriveKernelAndRecordResults_broken_kernel(
db: grewe_features_db.Database,
env: cldrive_env.OpenCLEnvironment,
dynamic_params,
num_runs: int,
):
drive_with_cldrive.DriveKernelAndRecordResults(db, 10, "invalid kernel",
env, dynamic_params,
num_runs)
with db.Session() as s:
assert s.query(grewe_features_db.DynamicFeatures).count() == 1
record = s.query(grewe_features_db.DynamicFeatures).first()
assert record.static_features_id == 10
assert record.opencl_env == env.name
assert record.hostname == system.HOSTNAME
assert record.outcome == "PROGRAM_COMPILATION_FAILURE"
assert record.gsize is None
assert record.wgsize is None
assert record.work_item_local_mem_size is None
assert record.work_item_private_mem_size is None
assert record.transferred_bytes is None
assert record.transfer_time_ns is None
assert record.kernel_time_ns is None
def test_GetBatchOfKernelsToDrive_overlap(db: grewe_features_db.Database,
env: cldrive_env.OpenCLEnvironment):
# Add a dynamic features to the database.
with db.Session() as s:
features = s.query(grewe_features_db.StaticFeatures).first()
s.add(_DynamicFeatures(features, env))
s.flush()
batch = drive_with_cldrive.GetBatchOfKernelsToDrive(s, env, 16)
assert len(batch) == 2
assert len(set(b.src for b in batch)) == 2
def Sample(
instance: clgen.Instance,
db: grewe_features_db.Database,
profiler: prof.AutoCsvProfiler,
pool: multiprocessing.Pool,
):
observer = SampleObserver()
with profiler.Profile(f"Create {FLAGS.batch_size} samples"):
samples = instance.model.Sample(instance.sampler, [observer])
prefix = "phd_experimental_deeplearning_"
with tempfile.TemporaryDirectory(prefix=prefix) as d:
d = pathlib.Path(d)
with profiler.Profile(f"Create {FLAGS.batch_size} tempfiles"):
paths_to_import = [
CreateTempFileFromSample(d, sample, i)
for i, sample in enumerate(observer.samples)
]
with profiler.Profile() as p:
num_successes = db.ImportStaticFeaturesFromPaths(
paths_to_import, FLAGS.origin)
p.name = f"Import {num_successes} / {FLAGS.batch_size} samples"
def test_DriveKernelAndRecordResults(
db: grewe_features_db.Database,
env: cldrive_env.OpenCLEnvironment,
dynamic_params,
num_runs: int,
):
drive_with_cldrive.DriveKernelAndRecordResults(
db,
0,
"kernel void A(global int* a) { a[get_global_id(0)] *= 2; }",
env,
dynamic_params,
num_runs,
)
with db.Session() as s:
assert s.query(
grewe_features_db.DynamicFeatures).count() == (len(dynamic_params))
records = s.query(grewe_features_db.DynamicFeatures).all()
for i, record in enumerate(records):
print("Testing record", i, record)
assert record.static_features_id == 0
assert record.opencl_env == env.name
assert record.hostname == system.HOSTNAME
assert record.outcome == "PASS"
assert record.run_count == num_runs
assert record.gsize == dynamic_params[i].global_size_x
assert record.wgsize == dynamic_params[i].local_size_x
assert record.work_item_local_mem_size == 0
assert record.work_item_private_mem_size == 0
assert record.transferred_bytes == (
2 * 4 * dynamic_params[i].global_size_x)
assert record.transfer_time_ns >= 100 # Flaky but likely.
assert record.kernel_time_ns >= 100 # Flaky but likely.
def __init__(self, database: db.Database):
self.session = database.MakeSession()
self.records = []
self.origin_to_features_id_map = {}
def DriveKernelAndRecordResults(
database: db.Database,
static_features_id: int,
src: str,
env: cldrive_env.OpenCLEnvironment,
dynamic_params: typing.List[cldrive_pb2.DynamicParams],
num_runs: int,
) -> None:
"""Drive a single kernel and record results."""
def ErrorFeatures(outcome: str) -> db.DynamicFeatures:
return db.DynamicFeatures(
static_features_id=static_features_id,
driver=db.DynamicFeaturesDriver.CLDRIVE,
opencl_env=env.name,
hostname=system.HOSTNAME,
outcome=outcome,
run_count=0,
)
try:
df = cldrive.DriveToDataFrame(
cldrive_pb2.CldriveInstances(instance=[
cldrive_pb2.CldriveInstance(
device=env.proto,
opencl_src=src,
dynamic_params=dynamic_params,
min_runs_per_kernel=num_runs,
)
]),
timeout_seconds=FLAGS.cldrive_timeout_seconds,
)
# Record programs which contain no kernels.
if not len(df):
with database.Session(commit=True) as session:
session.add(ErrorFeatures("NO_KERNELS"))
return
# Remove the columns which are not exported to the database:
# 'instance' is not used since we only drive a single instance at a time.
# 'build_opts' is never changed. 'kernel' is not needed because each static
# features entry is a single kernel.
df.drop(columns=["instance", "build_opts", "kernel"], inplace=True)
# Fix the naming differences between cldrive and the database.
df.rename(
columns={
"device": "opencl_env",
"global_size": "gsize",
"local_size": "wgsize",
},
inplace=True,
)
# NaN values are excluded in groupby statements, and we need to groupby
# columns that may be NaN (gsize and wgsize). Replace NaN with -1 since all
# integer column values are >= 0, so this value will never occur normally.
# See: https://github.com/pandas-dev/pandas/issues/3729
nan_placeholder = -1
df[["gsize", "wgsize"]] = df[["gsize",
"wgsize"]].fillna(nan_placeholder)
# Aggregate runtimes and append run_count.
groupby_columns = ["opencl_env", "gsize", "wgsize", "outcome"]
run_counts = df.groupby(groupby_columns).count()["kernel_time_ns"]
df = df.groupby(groupby_columns).mean()
df["run_count"] = run_counts
df.reset_index(inplace=True)
# Now that we have done the groupby, replace the NaN placeholder values
# with true NaN.
df[["gsize",
"wgsize"]] = df[["gsize",
"wgsize"]].replace(nan_placeholder, np.nan)
# Add missing columns.
df["static_features_id"] = static_features_id
df["driver"] = db.DynamicFeaturesDriver.CLDRIVE
df["hostname"] = system.HOSTNAME
# Import the dataframe into the SQL table.
df.to_sql(
db.DynamicFeatures.__tablename__,
con=database.engine,
if_exists="append",
index=False,
dtype={"driver": sql.Enum(db.DynamicFeaturesDriver)},
)
app.Log(1, "Imported %d dynamic features", len(df))
except cldrive.CldriveCrash:
with database.Session(commit=True) as session:
session.add(ErrorFeatures("DRIVER_CRASH"))
app.Log(1, "Driver crashed")
except pbutil.ProtoWorkerTimeoutError:
with database.Session(commit=True) as session:
session.add(ErrorFeatures("DRIVER_TIMEOUT"))
app.Log(1, "Driver timed out")
def ImportFromLegacyGpgpu(
database: db.Database,
logs_zip: pathlib.Path,
expected_devtype: str,
expected_device_name: str,
opencl_env: str,
hostname: str,
) -> None:
"""Import legacy GPGPU logs to database."""
failures = []
with database.Session() as session:
origin_to_features_id_map = {}
rows = []
with tempfile.TemporaryDirectory(prefix="phd_gpgpu_") as d:
with zipfile.ZipFile(logs_zip) as zipf:
zipf.extractall(d)
logs_dir = pathlib.Path(d) / "logs"
assert logs_dir.is_dir()
# Directory structure of archive is:
# logs/<run_num>/<device_name>/<benchmark_log>.
app.Log(1, "Importing logs from %d runs",
len(list(logs_dir.iterdir())))
for run_dir in logs_dir.iterdir():
for device_dir in run_dir.iterdir():
for benchmark_log in device_dir.iterdir():
# Re-construct the benchmark name from the file path.
log_filename_components = benchmark_log.name.split("-")
benchmark_suite = "-".join(
log_filename_components[:-1])
if benchmark_suite == "npb-3.3":
benchmark_name, _ = log_filename_components[
-1].split(".")
else:
benchmark_name = log_filename_components[-1]
if (benchmark_suite == "nvidia-4.2"
and benchmark_name.startswith("ocl")):
benchmark_name = benchmark_name[len("ocl"):]
log_lines = fs.Read(benchmark_log).split("\n")
kernel_invocations = libcecl_runtime.KernelInvocationsFromCeclLog(
log_lines,
# Legacy libcecl converted the nanoseconds returned by OpenCL API
# to milliseconds float. See:
# https://github.com/ChrisCummins/phd/blob/6e30717650380d7733d53984189ae579e19dc273/gpu/libcecl/libcecl.c#L153-L164
nanosecond_identity=lambda s: int(float(s) * 1e6),
expected_devtype=expected_devtype,
expected_device_name=expected_device_name,
)
for kernel_invocation in kernel_invocations:
origin = (
f"benchmarks_{benchmark_suite}:{benchmark_name}:"
f"{kernel_invocation.kernel_name}")
if origin in origin_to_features_id_map:
static_features_id = origin_to_features_id_map[
origin]
else:
static_features_id = GetStaticFeaturesId(
session, origin,
kernel_invocation.kernel_name, log_lines)
origin_to_features_id_map[
origin] = static_features_id
if static_features_id:
rows.append(
DynamicFeaturesFromKernelInvocation(
kernel_invocation, static_features_id))
df = pd.DataFrame(rows)
# Aggregate runtimes and append run_count.
groupby_columns = ["static_features_id", "gsize", "wgsize"]
run_counts = (
df[groupby_columns +
["run_count"]].groupby(groupby_columns).sum()["run_count"])
df = df.groupby(groupby_columns).mean()
df["run_count"] = run_counts
df.reset_index(inplace=True)
# Add missing columns.
df["driver"] = db.DynamicFeaturesDriver.LIBCECL
df["outcome"] = "PASS"
df["opencl_env"] = opencl_env
df["hostname"] = hostname
df.to_sql(
db.DynamicFeatures.__tablename__,
con=database.engine,
if_exists="append",
index=False,
dtype={"driver": sql.Enum(db.DynamicFeaturesDriver)},
)
app.Log(1, "Failed to find features for %d kernels", len(failures))
app.Log(
1,
"Imported %d records from %d kernels (%d distinct feature vectors)",
len(df),
len(origin_to_features_id_map),
len(set(origin_to_features_id_map.values())),
)
def test_GetBatchOfKernelsToDrive(db: grewe_features_db.Database,
env: cldrive_env.OpenCLEnvironment):
with db.Session() as s:
batch = drive_with_cldrive.GetBatchOfKernelsToDrive(s, env, 16)
assert len(batch) == 3
assert len(set(b.src for b in batch)) == 3